Merge pull request #685 from ROCm-Developer-Tools/hip-trig-return

Improve hip_trig test case

[ROCm/hip commit: 5c665acf74]

projects/hip/tests/src/deviceLib/hip_trig.cpp

@@ -32,21 +32,26 @@ THE SOFTWARE.
 #include "test_common.h"
 #include <hip/device_functions.h>
 
+#define HIP_ASSERT(x) (assert((x) == hipSuccess))
+
 #define LEN 512
 #define SIZE LEN << 2
 
+#define TEST_DEBUG (0)
+
 __global__ void kernel_trig(hipLaunchParm lp, float* In, float* sin_d, float* cos_d, float* tan_d,
                             float* sin_pd, float* cos_pd) {
     int tid = threadIdx.x + blockIdx.x * blockDim.x;
-    sin_d[tid] = __sinf(In[tid]);
-    cos_d[tid] = __cosf(In[tid]);
-    tan_d[tid] = __tanf(In[tid]);
-    __sincosf(In[tid], &sin_pd[tid], &cos_pd[tid]);
+    sin_d[tid] = sinf(In[tid]);
+    cos_d[tid] = cosf(In[tid]);
+    tan_d[tid] = tanf(In[tid]);
+    sincosf(In[tid], &sin_pd[tid], &cos_pd[tid]);
 }
 
 int main() {
     float *In, *sin_h, *cos_h, *tan_h, *sin_ph, *cos_ph;
     float *In_d, *sin_d, *cos_d, *tan_d, *sin_pd, *cos_pd;
+    int errors = 0;
     In = new float[LEN];
     sin_h = new float[LEN];
     cos_h = new float[LEN];
@@ -61,14 +66,16 @@ int main() {
         sin_ph[i] = 0.0f;
         cos_ph[i] = 0.0f;
     }
-    hipMalloc((void**)&In_d, SIZE);
-    hipMalloc((void**)&sin_d, SIZE);
-    hipMalloc((void**)&cos_d, SIZE);
-    hipMalloc((void**)&tan_d, SIZE);
-    hipMalloc((void**)&sin_pd, SIZE);
-    hipMalloc((void**)&cos_pd, SIZE);
+    HIP_ASSERT(hipMalloc((void**)&In_d, SIZE));
+    HIP_ASSERT(hipMalloc((void**)&sin_d, SIZE));
+    HIP_ASSERT(hipMalloc((void**)&cos_d, SIZE));
+    HIP_ASSERT(hipMalloc((void**)&tan_d, SIZE));
+    HIP_ASSERT(hipMalloc((void**)&sin_pd, SIZE));
+    HIP_ASSERT(hipMalloc((void**)&cos_pd, SIZE));
+
     hipMemcpy(In_d, In, SIZE, hipMemcpyHostToDevice);
-    hipLaunchKernel(kernel_trig, dim3(LEN, 1, 1), dim3(1, 1, 1), 0, 0, In_d, sin_d, cos_d, tan_d,
+    hipLaunchKernel(kernel_trig, dim3(LEN, 1, 1), dim3(1, 1, 1), 0, 0,
+                    In_d, sin_d, cos_d, tan_d,
                     sin_pd, cos_pd);
     hipMemcpy(sin_h, sin_d, SIZE, hipMemcpyDeviceToHost);
     hipMemcpy(cos_h, cos_d, SIZE, hipMemcpyDeviceToHost);
@@ -77,8 +84,28 @@ int main() {
     hipMemcpy(cos_ph, cos_pd, SIZE, hipMemcpyDeviceToHost);
     for (int i = 0; i < LEN; i++) {
         if (sin_h[i] != sin_ph[i] || cos_h[i] != cos_ph[i] || tan_h[i] * cos_h[i] != sin_h[i]) {
-            std::cout << "Failed!" << std::endl;
+            errors++;
+#if TEST_DEBUG
+            std::cout << "Check Failed!" << std::endl;
+            std::cout << " sin_h: " << sin_h[i] << " sin_ph: " << sin_ph[i] << "\n"
+                      << " cos_h: " << cos_h[i] << " cos_ph:" << cos_ph[i] << "\n"
+                      << " tan_h * cos_h: " << tan_h[i] * cos_h[i] << " sin_h[i]: " << sin_h[i] << "\n";
+#endif
         }
     }
-    passed();
+
+    HIP_ASSERT(hipFree(In_d));
+    HIP_ASSERT(hipFree(sin_d));
+    HIP_ASSERT(hipFree(cos_d));
+    HIP_ASSERT(hipFree(tan_d));
+    HIP_ASSERT(hipFree(sin_pd));
+    HIP_ASSERT(hipFree(cos_pd));
+
+    if (errors != 0) {
+        std::cout << "hip_trig FAILED!" << std::endl;
+        return -1;
+    } else {
+        std::cout << "hip_trig PASSED!" << std::endl;
+    }
+    return errors;
 }