Add direct test for hipMemsetD32 and hipMemsetD32Async

tests/src/runtimeApi/memory/hipMemset.cpp

@@ -26,11 +26,11 @@ THE SOFTWARE.
  * BUILD: %t %s ../../test_common.cpp
  * RUN: %t
  * //Small copy
- * RUN: %t -N 10    --memsetval 0x42
+ * RUN: %t -N 10    --memsetval 0x42 --memsetD32val 0x101
  * // Oddball size
- * RUN: %t -N 10013 --memsetval 0x5a
+ * RUN: %t -N 10013 --memsetval 0x5a --memsetD32val 0xDEADBEEF
  * // Big copy
- * RUN: %t -N 256M  --memsetval 0xa6
+ * RUN: %t -N 256M  --memsetval 0xa6 --memsetD32val 0xCAFEBABE
  * HIT_END
  */
 
@@ -62,6 +62,30 @@ bool testhipMemset(int memsetval,int p_gpuDevice)
     return testResult;
 }
 
+bool testhipMemsetD32(int memsetD32val,int p_gpuDevice)
+{
+    size_t Nbytes = N*sizeof(int);
+    printf ("testhipMemsetD32 N=%zu  memsetD32val=%8x device=%d\n", N, memsetD32val, p_gpuDevice);
+    int *A_d;
+    int *A_h;
+    bool testResult = true;
+
+    HIPCHECK ( hipMalloc(&A_d, Nbytes) );
+    A_h = (int*)malloc(Nbytes);
+    HIPCHECK ( hipMemsetD32(A_d, memsetD32val, N) );
+    HIPCHECK ( hipMemcpy(A_h, A_d, Nbytes, hipMemcpyDeviceToHost));
+
+    for (int i=0; i<N; i++) {
+        if (A_h[i] != memsetD32val) {
+            testResult = false;            printf("mismatch at index:%d computed:%08x, memsetD32val:%08x\n", i, A_h[i], memsetD32val);
+            break;
+        }
+    }
+    HIPCHECK(hipFree(A_d));
+    free(A_h);
+    return testResult;
+}
+
 bool testhipMemsetAsync(int memsetval,int p_gpuDevice)
 {
     size_t Nbytes = N*sizeof(int);
@@ -91,6 +115,35 @@ bool testhipMemsetAsync(int memsetval,int p_gpuDevice)
     return testResult;
 }
 
+bool testhipMemsetD32Async(int memsetD32val,int p_gpuDevice)
+{
+    size_t Nbytes = N*sizeof(int);
+    printf ("testhipMemsetD32Async N=%zu  memsetval=%8x device=%d\n", N, memsetD32val, p_gpuDevice);
+    int *A_d;
+    int *A_h;
+    bool testResult = true;
+
+    HIPCHECK ( hipMalloc((void**)&A_d, Nbytes) );
+    A_h = (int*)malloc(Nbytes);
+    hipStream_t stream;
+    HIPCHECK(hipStreamCreate(&stream));
+    HIPCHECK ( hipMemsetD32Async(A_d, memsetD32val, N, stream ));
+    HIPCHECK ( hipStreamSynchronize(stream));
+    HIPCHECK ( hipMemcpy(A_h, (void*)A_d, Nbytes, hipMemcpyDeviceToHost));
+
+    for (int i=0; i<N; i++) {
+        if (A_h[i] != memsetD32val) {
+            testResult = false;
+            printf("mismatch at index:%d computed:%02x, memsetD32val:%02x\n", i, A_h[i], memsetD32val);
+            break;
+        }
+    }
+    HIPCHECK(hipFree((void*)A_d));
+    HIPCHECK(hipStreamDestroy(stream));
+    free(A_h);
+    return testResult;
+}
+
 int main(int argc, char *argv[])
 {
     HipTest::parseStandardArguments(argc, argv, true);
@@ -98,6 +151,8 @@ int main(int argc, char *argv[])
     HIPCHECK(hipSetDevice(p_gpuDevice));
     testResult &= testhipMemset(memsetval, p_gpuDevice);
     testResult &= testhipMemsetAsync(memsetval, p_gpuDevice);
+    testResult &= testhipMemsetD32(memsetD32val, p_gpuDevice);
+    testResult &= testhipMemsetD32Async(memsetD32val, p_gpuDevice);
     if (testResult) passed();
     failed("Output Mismatch\n"); 
 }