SWDEV-347670 - StreamWait test issues

SWDEV-439486 - Fix Unit_hipStreamValue_Wait32_NonBlocking_Mask_Gte

Change-Id: Iab61cacaa72ac88905a2befe3aff931f8216d152

catch/hipTestMain/config/config_amd_windows

@@ -63,6 +63,7 @@
         "Unit_hipDeviceGetCacheConfig_Positive_Basic",
         "Unit_hipDeviceGetCacheConfig_Positive_Basic",
         "Unit_hipDeviceGetCacheConfig_Positive_Threaded",
+        "SWDEV-347670 - blocking tests have TDR, causing hangs",
         "Unit_hipStreamValue_Wait32_Blocking_Mask_Gte",
         "Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_1",
         "Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_2",
@@ -77,6 +78,8 @@
         "Unit_hipStreamValue_Wait64_Blocking_Mask_And",
         "Unit_hipStreamValue_Wait64_Blocking_NoMask_Gte",
         "Unit_hipStreamValue_Wait64_Blocking_NoMask_Eq",
+        "Unit_hipStreamValue_Wait32_Blocking_NoMask_Nor",
+        "Unit_hipStreamValue_Wait64_Blocking_NoMask_And",
         "Unit_hipGetDeviceFlags_Positive_Context",
         "Unit_hipIpcCloseMemHandle_Negative_Close_In_Originating_Process",
         "Unit_hipIpcOpenMemHandle_Negative_Open_In_Creating_Process",

catch/unit/stream/hipStreamValue.cc

@@ -165,6 +165,8 @@ template <PtrType type, typename UIntT> auto allocMem() {
     *signalPtr = 0;
 
     auto freeStuff = [](uint64_t* sPtr) { HIP_CHECK(hipFree(sPtr)); };
+    // sync ensures the memory intialization is complete before using the memory
+    HIP_CHECK(hipDeviceSynchronize());
     return TestPtr<type, UIntT, std::unique_ptr<uint64_t, decltype(freeStuff)>>{
         signalPtr, std::unique_ptr<uint64_t, decltype(freeStuff)>(signalPtr, freeStuff)};
   } else if constexpr (type == PtrType::DevicePtrToHost) {
@@ -183,7 +185,8 @@ template <PtrType type, typename UIntT> auto allocMem() {
       HIP_CHECK(hipHostUnregister(ptr));
       delete[] ptr;
     };
-
+    // sync ensures the memory intialization is complete before using the memory
+    HIP_CHECK(hipDeviceSynchronize());
     return TestPtr<type, UIntT, std::unique_ptr<UIntT[], decltype(freeStuff)>>{
         devicePtr, std::unique_ptr<UIntT[], decltype(freeStuff)>(hostPtr, freeStuff)};
   } else if constexpr (type == PtrType::HostPtr) {
@@ -199,7 +202,8 @@ template <PtrType type, typename UIntT> auto allocMem() {
       HIP_CHECK(hipHostUnregister(ptr));
       delete[] ptr;
     };
-
+    // sync ensures the memory intialization is complete before using the memory
+    HIP_CHECK(hipDeviceSynchronize());
     return TestPtr<type, UIntT, std::unique_ptr<UIntT[], decltype(freeStuff)>>{
         hostPtr, std::unique_ptr<UIntT[], decltype(freeStuff)>(hostPtr, freeStuff)};
   } else {
@@ -208,6 +212,8 @@ template <PtrType type, typename UIntT> auto allocMem() {
     HIP_CHECK(hipMalloc(&devicePtr, sizeof(UIntT) * arraySize));
     HIP_CHECK(hipMemset(devicePtr, 0, sizeof(UIntT) * arraySize));
     auto freeStuff = [](UIntT* ptr) { HIP_CHECK(hipFree(ptr)); };
+    // sync ensures the memory intialization is complete before using the memory
+    HIP_CHECK(hipDeviceSynchronize());
     return TestPtr<type, UIntT, std::unique_ptr<UIntT, decltype(freeStuff)>>{
         devicePtr, std::unique_ptr<UIntT, decltype(freeStuff)>(devicePtr, freeStuff)};
   }

catch/unit/vulkan_interop/hipDestroyExternalSemaphore.cc

@@ -21,8 +21,6 @@ THE SOFTWARE.
 
 #include "vulkan_test.hh"
 
-constexpr bool enable_validation = false;
-
 TEST_CASE("Unit_hipDestroyExternalSemaphore_Vulkan_Negative_Parameters") {
   SECTION("extSem == nullptr") {
     HIP_CHECK_ERROR(hipDestroyExternalSemaphore(nullptr), hipErrorInvalidValue);
@@ -30,6 +28,7 @@ TEST_CASE("Unit_hipDestroyExternalSemaphore_Vulkan_Negative_Parameters") {
 
 // Segfaults in Nvidia and Amd
 #if 0
+  constexpr bool enable_validation = false;
   SECTION("Double free") {
     VulkanTest vkt(enable_validation);
     const auto ext_semaphore = ImportBinarySemaphore(vkt);