SWDEV-299127 - Merge 'develop' into 'amd-staging'

Change-Id: I7c7b322e43bada968c3da96456ed0e5a3c1d4041

tests/catch/hipTestMain/config/config_amd_windows_MI2xx.json

@@ -71,6 +71,23 @@
 		"Unit_hipStreamSynchronize_FinishWork",
 		"Unit_hipStreamSynchronize_NullStreamAndStreamPerThread",
 		"Unit_hipMultiThreadDevice_NearZero",
-		"Unit_hipStreamPerThread_DeviceReset_1"
+		"Unit_hipStreamPerThread_DeviceReset_1",
+		"Unit_hipStreamValue_Wait32_Blocking_Mask_Gte",
+		"Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_1",
+		"Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_2",
+		"Unit_hipStreamValue_Wait32_Blocking_Mask_And",
+		"Unit_hipStreamValue_Wait32_Blocking_NoMask_Eq",
+		"Unit_hipStreamValue_Wait32_Blocking_NoMask_Gte",
+		"Unit_hipStreamValue_Wait32_Blocking_NoMask_And",
+		"Unit_hipStreamValue_Wait32_Blocking_NoMask_Nor",
+		"Unit_hipStreamValue_Wait64_Blocking_Mask_Gte_1",
+		"Unit_hipStreamValue_Wait64_Blocking_Mask_Gte_2",
+		"Unit_hipStreamValue_Wait64_Blocking_Mask_Eq_1",
+		"Unit_hipStreamValue_Wait64_Blocking_Mask_Eq_2",
+		"Unit_hipStreamValue_Wait64_Blocking_Mask_And",
+		"Unit_hipStreamValue_Wait64_Blocking_NoMask_Gte",
+		"Unit_hipStreamValue_Wait64_Blocking_NoMask_Eq",
+		"Unit_hipStreamValue_Wait64_Blocking_NoMask_And",
+		"Unit_hipStreamValue_Wait64_Blocking_NoMask_Nor" 
 	]
 }

tests/catch/hipTestMain/config/config_amd_windows_common.json

@@ -78,6 +78,23 @@
 	   "Unit_hipGraphMemcpyNodeSetParamsToSymbol_Functional",
 	   "Unit_hipStreamQuery_WithPendingWork",
 	   "Unit_hipStreamWaitEvent_DifferentStreams",
-	   "Unit_hipStreamQuery_WithFinishedWork"
+	   "Unit_hipStreamQuery_WithFinishedWork",
+	   "Unit_hipStreamValue_Wait32_Blocking_Mask_Gte",
+	   "Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_1",
+	   "Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_2",
+	   "Unit_hipStreamValue_Wait32_Blocking_Mask_And",
+	   "Unit_hipStreamValue_Wait32_Blocking_NoMask_Eq",
+	   "Unit_hipStreamValue_Wait32_Blocking_NoMask_Gte",
+	   "Unit_hipStreamValue_Wait32_Blocking_NoMask_And",
+	   "Unit_hipStreamValue_Wait32_Blocking_NoMask_Nor",
+	   "Unit_hipStreamValue_Wait64_Blocking_Mask_Gte_1",
+	   "Unit_hipStreamValue_Wait64_Blocking_Mask_Gte_2",
+	   "Unit_hipStreamValue_Wait64_Blocking_Mask_Eq_1",
+	   "Unit_hipStreamValue_Wait64_Blocking_Mask_Eq_2",
+	   "Unit_hipStreamValue_Wait64_Blocking_Mask_And",
+	   "Unit_hipStreamValue_Wait64_Blocking_NoMask_Gte",
+	   "Unit_hipStreamValue_Wait64_Blocking_NoMask_Eq",
+	   "Unit_hipStreamValue_Wait64_Blocking_NoMask_And",
+	   "Unit_hipStreamValue_Wait64_Blocking_NoMask_Nor"
 	]
 }

tests/catch/include/hip_test_common.hh

@@ -397,7 +397,6 @@ static inline void runKernelForDuration(std::chrono::milliseconds duration,
   static size_t ticksPerSecond = findTicksPerSecond();
   const auto millis = duration.count();
   hipLaunchKernelGGL(waitKernel, dim3(1), dim3(1), 0, stream, ticksPerSecond * millis / 1000);
-  HIP_CHECK(hipGetLastError());
 }
 
 }  // namespace HipTest

tests/catch/unit/event/Unit_hipEventIpc.cc

@@ -91,9 +91,13 @@ TEST_CASE("Unit_hipEventIpc") {
     hipEvent_t ipc_event;
     hipError_t err = hipIpcOpenEventHandle(&ipc_event, ipc_handle);
 
+    #if HT_WIN
+    // always different process Id on Windows
+    HIP_CHECK(err);
+    #else
     // hipIpcOpenEventHandle() should be called in a different process, hence it should fail here
     REQUIRE(err == hipErrorInvalidContext);
-
+    #endif
     HIP_CHECK(hipEventDestroy(start));
     HIP_CHECK(hipEventDestroy(stop));
 

tests/catch/unit/event/hipEventSynchronize.cc

@@ -110,13 +110,13 @@ TEST_CASE("Unit_hipEventSynchronize_NoEventRecord_Positive") {
 
   // Record the end_event
   HIP_CHECK(hipEventRecord(end_event, NULL));
+  // End event has not been completed
+  HIP_CHECK_ERROR(hipEventQuery(end_event), hipErrorNotReady);
 
   // When hipEventSynchronized is called on event that has not been recorded,
   // the function returns immediately
   HIP_CHECK(hipEventSynchronize(dummy_event));
 
-  // End event has not been completed
-  HIP_CHECK_ERROR(hipEventQuery(end_event), hipErrorNotReady);
   // Wait for end_event to complete
   HIP_CHECK(hipEventSynchronize(end_event));