SWDEV-491246 - add stream capture tests for Memset sync APIs (#1776)

projects/hip-tests/catch/unit/memory/hipMemset.cc

@@ -269,3 +269,51 @@ TEST_CASE("Unit_hipMemset_2AsyncOperations") {
   HIP_CHECK(hipFree(p2));
   HIP_CHECK(hipStreamDestroy(s));
 }
+
+/**
+ * Test Description
+ * ------------------------
+ *    - Test hipMemset while stream is capturing.
+ * Test source
+ * ------------------------
+ *    - unit/memory/hipMemset.cc
+ * Test requirements
+ * ------------------------
+ *    - HIP_VERSION >= 6.0
+ */
+TEST_CASE("Unit_hipMemset_Capture") {
+  const size_t N = 1024;
+  void* dst = nullptr;
+  HIP_CHECK(hipMalloc(&dst, N));
+
+  hipError_t memcpy_err = hipSuccess;
+  BEGIN_CAPTURE_SYNC(memcpy_err, false);
+  HIP_CHECK_ERROR(hipMemset(dst, 0xAB, N), memcpy_err);
+  END_CAPTURE_SYNC(memcpy_err);
+
+  HIP_CHECK(hipFree(dst));
+}
+
+/**
+ * Test Description
+ * ------------------------
+ *    - Test hipMemsetD8 while stream is capturing.
+ * Test source
+ * ------------------------
+ *    - unit/memory/hipMemset.cc
+ * Test requirements
+ * ------------------------
+ *    - HIP_VERSION >= 6.0
+ */
+TEST_CASE("Unit_hipMemsetD8_Capture") {
+  const size_t N = 512;
+  void* dst = nullptr;
+  HIP_CHECK(hipMalloc(&dst, N  * sizeof(uint8_t)));
+
+  hipError_t memcpy_err = hipSuccess;
+  BEGIN_CAPTURE_SYNC(memcpy_err, false);
+  HIP_CHECK_ERROR(hipMemsetD8(dst, 0xCD, N), memcpy_err);
+  END_CAPTURE_SYNC(memcpy_err);
+
+  HIP_CHECK(hipFree(dst));
+}

projects/hip-tests/catch/unit/memory/hipMemset2D.cc

@@ -227,3 +227,35 @@ TEST_CASE("Unit_hipMemset2DAsync_capturehipMemset2DAsync") {
   free(A_h);
   free(B_h);
 }
+
+/**
+ * Test Description
+ * ------------------------
+ *    - Test hipMemset2D while stream is capturing.
+ * Test source
+ * ------------------------
+ *    - unit/memory/hipMemset2D.cc
+ * Test requirements
+ * ------------------------
+ *    - HIP_VERSION >= 6.0
+ */
+TEST_CASE("Unit_hipMemset2D_Capture") {
+  CHECK_IMAGE_SUPPORT
+
+  constexpr int memsetval = 0x24;
+  constexpr size_t numH = 256;
+  constexpr size_t numW = 256;
+  size_t pitch_A;
+  size_t width = numW * sizeof(char);
+  void* dst = nullptr;
+
+  HIP_CHECK(hipMallocPitch(reinterpret_cast<void**>(&dst), &pitch_A, width,
+                          numH));
+
+  hipError_t memcpy_err = hipSuccess;
+  BEGIN_CAPTURE_SYNC(memcpy_err, false);
+  HIP_CHECK_ERROR(hipMemset2D(dst, pitch_A, memsetval, numW, numH), memcpy_err);
+  END_CAPTURE_SYNC(memcpy_err);
+
+  HIP_CHECK(hipFree(dst));
+}

projects/hip-tests/catch/unit/memory/hipMemset3D.cc

@@ -184,3 +184,36 @@ TEST_CASE("Unit_hipMemset3DAsync_capturehipMemset3DAsync") {
   HIP_CHECK(hipFree(A_d.ptr));
   free(A_h);
 }
+
+/**
+ * Test Description
+ * ------------------------
+ *    - Test hipMemset3D while stream is capturing.
+ * Test source
+ * ------------------------
+ *    - unit/memory/hipMemset3D.cc
+ * Test requirements
+ * ------------------------
+ *    - HIP_VERSION >= 6.0
+ */
+TEST_CASE("Unit_hipMemset3D_Capture") {
+  CHECK_IMAGE_SUPPORT
+
+  constexpr int memsetval = 0x22;
+  constexpr size_t numH = 256;
+  constexpr size_t numW = 256;
+  constexpr size_t depth = 10;
+  size_t width = numW * sizeof(char);
+
+  hipExtent extent = make_hipExtent(width, numH, depth);
+  hipPitchedPtr devPitchedPtr;
+
+  HIP_CHECK(hipMalloc3D(&devPitchedPtr, extent));
+
+  hipError_t memcpy_err = hipSuccess;
+  BEGIN_CAPTURE_SYNC(memcpy_err, false);
+  HIP_CHECK_ERROR(hipMemset3D(devPitchedPtr, memsetval, extent), memcpy_err);
+  END_CAPTURE_SYNC(memcpy_err);
+
+  HIP_CHECK(hipFree(devPitchedPtr.ptr));
+}

projects/hip-tests/catch/unit/memory/hipMemsetD16.cc

@@ -195,3 +195,27 @@ TEST_CASE("Unit_hipMemsetD16_KernelBuffer") {
 
   REQUIRE(result == true);
 }
+
+/**
+ * Test Description
+ * ------------------------
+ *    - Test hipMemsetD16 while stream is capturing.
+ * Test source
+ * ------------------------
+ *    - unit/memory/hipMemsetD16.cc
+ * Test requirements
+ * ------------------------
+ *    - HIP_VERSION >= 6.0
+ */
+TEST_CASE("Unit_hipMemsetD16_Capture") {
+  const size_t N = 256;
+  void* dst = nullptr;
+  HIP_CHECK(hipMalloc(&dst, N * sizeof(uint16_t)));
+
+  hipError_t memcpy_err = hipSuccess;
+  BEGIN_CAPTURE_SYNC(memcpy_err, false);
+  HIP_CHECK_ERROR(hipMemsetD16(dst, 0xAB, N), memcpy_err);
+  END_CAPTURE_SYNC(memcpy_err);
+
+  HIP_CHECK(hipFree(dst));
+}

projects/hip-tests/catch/unit/memory/hipMemsetD32.cc

@@ -195,3 +195,27 @@ TEST_CASE("Unit_hipMemsetD32_KernelBuffer") {
 
   REQUIRE(result == true);
 }
+
+/**
+ * Test Description
+ * ------------------------
+ *    - Test hipMemsetD32 while stream is capturing.
+ * Test source
+ * ------------------------
+ *    - unit/memory/hipMemsetD32.cc
+ * Test requirements
+ * ------------------------
+ *    - HIP_VERSION >= 6.0
+ */
+TEST_CASE("Unit_hipMemsetD32_Capture") {
+  const size_t N = 128;
+  void* dst = nullptr;
+  HIP_CHECK(hipMalloc(&dst, N * sizeof(uint32_t)));
+
+  hipError_t memcpy_err = hipSuccess;
+  BEGIN_CAPTURE_SYNC(memcpy_err, false);
+  HIP_CHECK_ERROR(hipMemsetD32(dst, 0xAB, N), memcpy_err);
+  END_CAPTURE_SYNC(memcpy_err);
+
+  HIP_CHECK(hipFree(dst));
+}