SWDEV-491252 - Add stream capture testcases to host allocation APIs (#590)

* SWDEV-491252 - Add stream capture testcases to host allocation APIs

* SWDEV-491252 - Add stream capture behavior testcase for hipFreeHost

* SWDEV-491252 - Refactor capture testcases

* SWDEV-491252 - Run clang-format

---------

Co-authored-by: Marko Arandjelovic <Marko.Arandjelovic@amd.com>

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

@@ -104,3 +104,16 @@ TEST_CASE("Unit_hipFreeHost_Multithreading") {
   }
   HIP_CHECK_THREAD_FINALIZE();
 }
+
+TEST_CASE("Unit_hipFreeHost_Capture") {
+  void* ptr = nullptr;
+  constexpr size_t kPtrSize = 1024;
+
+  HIP_CHECK(hipHostMalloc(&ptr, kPtrSize));
+
+  hipError_t capture_error = hipSuccess;
+  constexpr bool kRelaxedModeAllowed = true;
+  BEGIN_CAPTURE_SYNC(capture_error, kRelaxedModeAllowed);
+  HIP_CHECK_ERROR(hipFreeHost(ptr), capture_error);
+  END_CAPTURE_SYNC(capture_error);
+}

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

@@ -384,3 +384,22 @@ TEST_CASE("Unit_hipHostAlloc_ArgValidation") {
     REQUIRE(ptr == nullptr);
   }
 }
+
+TEST_CASE("Unit_hipHostAlloc_Capture") {
+  int* host_memory = nullptr;
+  int flags = get_flags();
+
+  hipError_t capture_error = hipSuccess;
+  constexpr bool kRelaxedModeAllowed = true;
+  BEGIN_CAPTURE_SYNC(capture_error, kRelaxedModeAllowed);
+
+  HIP_CHECK_ERROR(hipHostAlloc(reinterpret_cast<void**>(&host_memory), sizeof(int), flags),
+                  capture_error);
+
+  END_CAPTURE_SYNC(capture_error);
+
+  if (capture_error == hipSuccess) {
+    REQUIRE(host_memory != nullptr);
+    HIP_CHECK(hipFreeHost(host_memory));
+  }
+}

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

@@ -104,3 +104,15 @@ TEST_CASE("Unit_hipHostFree_Multithreading") {
   }
   HIP_CHECK_THREAD_FINALIZE();
 }
+
+TEST_CASE("Unit_hipHostFree_Capture") {
+  void* host_ptr = nullptr;
+  constexpr size_t kAllocSize = 1024;
+  HIP_CHECK(hipHostMalloc(&host_ptr, kAllocSize));
+
+  hipError_t capture_error = hipSuccess;
+  constexpr bool kRelaxedModeAllowed = true;
+  BEGIN_CAPTURE_SYNC(capture_error, kRelaxedModeAllowed);
+  HIP_CHECK_ERROR(hipHostFree(host_ptr), capture_error);
+  END_CAPTURE_SYNC(capture_error);
+}

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

@@ -99,4 +99,26 @@ TEST_CASE("Unit_hipHostGetDevicePointer_UseCase") {
   }
 
   HIP_CHECK(hipHostFree(hPtr));
-}
+}
+
+TEST_CASE("Unit_hipHostGetDevicePointer_Capture") {
+  if (!DeviceAttributesSupport(0, hipDeviceAttributeCanMapHostMemory)) {
+    HipTest::HIP_SKIP_TEST("Device does not support mapping host memory");
+    return;
+  }
+
+  int* host_ptr = nullptr;
+  int* device_ptr = nullptr;
+  HIP_CHECK(hipHostMalloc(&host_ptr, sizeof(int)));
+
+  hipStream_t stream = nullptr;
+  HIP_CHECK(hipStreamCreate(&stream));
+
+  GENERATE_CAPTURE();
+  BEGIN_CAPTURE(stream);
+  HIP_CHECK(hipHostGetDevicePointer(reinterpret_cast<void**>(&device_ptr), host_ptr, 0));
+  END_CAPTURE(stream);
+
+  HIP_CHECK(hipStreamDestroy(stream));
+  HIP_CHECK(hipHostFree(host_ptr));
+}

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

@@ -220,3 +220,22 @@ TEST_CASE("Unit_hipHostGetFlags_InvalidArgs") {
     }
   }
 }
+
+TEST_CASE("Unit_hipHostGetFlags_Capture") {
+  unsigned int host_flags = 0;
+  void* host_ptr = nullptr;
+  constexpr size_t kAllocSize = 1024;
+
+  HIP_CHECK(hipHostMalloc(reinterpret_cast<void**>(&host_ptr), kAllocSize));
+
+  hipStream_t stream = nullptr;
+  HIP_CHECK(hipStreamCreate(&stream));
+
+  GENERATE_CAPTURE();
+  BEGIN_CAPTURE(stream);
+  HIP_CHECK(hipHostGetFlags(&host_flags, host_ptr));
+  END_CAPTURE(stream);
+
+  HIP_CHECK(hipHostFree(host_ptr));
+  HIP_CHECK(hipStreamDestroy(stream));
+}

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

@@ -297,3 +297,17 @@ TEST_CASE("Unit_hipHostMalloc_AllocateUseMoreThanAvailGPUMemory") {
   }
 }
 #endif
+
+TEST_CASE("Unit_hipHostMalloc_Capture") {
+  int* host_ptr = nullptr;
+  hipError_t capture_error = hipSuccess;
+
+  constexpr bool kRelaxedModeAllowed = true;
+  BEGIN_CAPTURE_SYNC(capture_error, kRelaxedModeAllowed);
+  HIP_CHECK_ERROR(hipHostMalloc(reinterpret_cast<void**>(&host_ptr), sizeBytes), capture_error);
+  END_CAPTURE_SYNC(capture_error);
+
+  if (host_ptr != nullptr) {
+    HIP_CHECK(hipHostFree(host_ptr));
+  }
+}

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

@@ -980,6 +980,21 @@ TEMPLATE_TEST_CASE("Unit_hipHostRegister_Negative", "", int, float, double) {
   }
 }
 
+TEST_CASE("Unit_hipHostRegister_Capture") {
+  constexpr size_t kBufferSize = 1024;
+  auto buffer = std::make_unique<int[]>(kBufferSize);
+  hipError_t capture_error = hipSuccess;
+
+  constexpr bool kRelaxedModeAllowed = true;
+  BEGIN_CAPTURE_SYNC(capture_error, kRelaxedModeAllowed);
+  HIP_CHECK_ERROR(hipHostRegister(buffer.get(), kBufferSize, 0), capture_error);
+  END_CAPTURE_SYNC(capture_error);
+
+  if (capture_error == hipSuccess) {
+    HIP_CHECK(hipHostUnregister(buffer.get()));
+  }
+}
+
 /**
  * End doxygen group MemoryTest.
  * @}

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

@@ -94,4 +94,21 @@ TEST_CASE("Unit_hipHostUnregister_AlreadyUnregisteredPointer") {
   }
 }
 
+TEST_CASE("Unit_hipHostUnregister_Capture") {
+  constexpr size_t kBufferSize = 1024;
+  auto buffer = std::make_unique<int[]>(kBufferSize);
+  hipError_t capture_error = hipSuccess;
+
+  HIP_CHECK_ERROR(hipHostRegister(buffer.get(), kBufferSize, 0), capture_error);
+
+  constexpr bool kRelaxedModeAllowed = true;
+  BEGIN_CAPTURE_SYNC(capture_error, kRelaxedModeAllowed);
+  HIP_CHECK_ERROR(hipHostUnregister(buffer.get()), capture_error);
+  END_CAPTURE_SYNC(capture_error);
+
+  if (capture_error != hipSuccess) {
+    HIP_CHECK(hipHostUnregister(buffer.get()));
+  }
+}
+
 }  // namespace hipHostUnregisterTests

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

@@ -71,3 +71,20 @@ TEST_CASE("Unit_hipMallocHost_Negative") {
     HIP_CHECK_ERROR(hipMallocHost(reinterpret_cast<void**>(&host_memory), -1), hipErrorOutOfMemory);
   }
 }
+
+TEST_CASE("Unit_hipMallocHost_Capture") {
+  int* host_memory = nullptr;
+
+  hipError_t capture_error = hipSuccess;
+  constexpr bool kRelaxedModeAllowed = true;
+
+  BEGIN_CAPTURE_SYNC(capture_error, kRelaxedModeAllowed);
+  HIP_CHECK_ERROR(hipMallocHost(reinterpret_cast<void**>(&host_memory), sizeof(int)),
+                  capture_error);
+  END_CAPTURE_SYNC(capture_error);
+
+  if (capture_error == hipSuccess) {
+    REQUIRE(host_memory != nullptr);
+    HIP_CHECK(hipHostFree(host_memory));
+  }
+}

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

@@ -130,3 +130,19 @@ TEST_CASE("Unit_hipMemAllocHost_VerifyAccess") {
     HIP_CHECK(hipCtxDestroy(devices_ctxs[device_index]));
   }
 }
+
+TEST_CASE("Unit_hipMemAllocHost_Capture") {
+  int* host_memory = nullptr;
+
+  hipError_t capture_error = hipSuccess;
+  constexpr bool kRelaxedModeAllowed = true;
+  BEGIN_CAPTURE_SYNC(capture_error, kRelaxedModeAllowed);
+  HIP_CHECK_ERROR(hipMemAllocHost(reinterpret_cast<void**>(&host_memory), sizeof(int)),
+                  capture_error);
+  END_CAPTURE_SYNC(capture_error);
+
+  if (capture_error == hipSuccess) {
+    REQUIRE(host_memory != nullptr);
+    HIP_CHECK(hipHostFree(host_memory));
+  }
+}