From d4e5dfa9429338de4f7750b830eaf5ae0df31544 Mon Sep 17 00:00:00 2001
From: Jaydeep Patel <jaydeepkumar.patel@amd.com>
Date: Tue, 17 Oct 2023 17:25:59 +0000
Subject: [PATCH] SWDEV-425248 - numerous bug fixes

https://gerrit-git.amd.com/c/compute/ec/clr/+/849086 changed stream sync behaviour and accordingly updated Unit_hipStreamCreateWithFlags_DefaultStreamInteraction.
Update error code to match for hipStreamWaitEvent.
Thread detach makes executions all async and while that is happening, It is unexpected effect for device reset so join thread is better.

Change-Id: I1affa84089626dee478d8bcc5aaa318e320fd6b0


[ROCm/hip-tests commit: be88e1dffc26c378d5eb924bf9b9c9b24eb1082d]
---
 .../catch/unit/stream/hipStreamCreateWithFlags.cc      | 10 +++++-----
 .../hip-tests/catch/unit/stream/hipStreamWaitEvent.cc  |  2 +-
 .../streamperthread/hipStreamPerThread_DeviceReset.cc  |  5 ++++-
 3 files changed, 10 insertions(+), 7 deletions(-)

diff --git a/projects/hip-tests/catch/unit/stream/hipStreamCreateWithFlags.cc b/projects/hip-tests/catch/unit/stream/hipStreamCreateWithFlags.cc
index 5038037452..1ffb9c1523 100644
--- a/projects/hip-tests/catch/unit/stream/hipStreamCreateWithFlags.cc
+++ b/projects/hip-tests/catch/unit/stream/hipStreamCreateWithFlags.cc
@@ -59,10 +59,7 @@ TEST_CASE("Unit_hipStreamCreateWithFlags_Default") {
 TEST_CASE("Unit_hipStreamCreateWithFlags_DefaultStreamInteraction") {
   const hipStream_t defaultStream = GENERATE(static_cast<hipStream_t>(nullptr), hipStreamPerThread);
   const unsigned int flagUnderTest = GENERATE(hipStreamDefault, hipStreamNonBlocking);
-  const hipError_t expectedError = (flagUnderTest == hipStreamDefault) && (defaultStream == nullptr)
-      ? hipErrorNotReady
-      : hipSuccess;
-  CAPTURE(defaultStream, flagUnderTest, expectedError, hipGetErrorString(expectedError));
+  CAPTURE(defaultStream, flagUnderTest);
 
   hipStream_t stream{};
   HIP_CHECK(hipStreamCreateWithFlags(&stream, flagUnderTest));
@@ -70,12 +67,15 @@ TEST_CASE("Unit_hipStreamCreateWithFlags_DefaultStreamInteraction") {
   constexpr auto delay = std::chrono::milliseconds(500);
 
   SECTION("default stream waiting for created stream") {
+    const hipError_t expectedError = (flagUnderTest == hipStreamDefault) && (defaultStream == nullptr)
+      ? hipErrorNotReady
+      : hipSuccess;
     LaunchDelayKernel(delay, stream);
     REQUIRE(hipStreamQuery(defaultStream) == expectedError);
   }
   SECTION("created stream waiting for default stream") {
     LaunchDelayKernel(delay, defaultStream);
-    REQUIRE(hipStreamQuery(stream) == expectedError);
+    REQUIRE(hipStreamQuery(stream) == hipSuccess);
   }
 
   HIP_CHECK(hipDeviceSynchronize());
diff --git a/projects/hip-tests/catch/unit/stream/hipStreamWaitEvent.cc b/projects/hip-tests/catch/unit/stream/hipStreamWaitEvent.cc
index 9877d548c1..32dbe9983b 100644
--- a/projects/hip-tests/catch/unit/stream/hipStreamWaitEvent.cc
+++ b/projects/hip-tests/catch/unit/stream/hipStreamWaitEvent.cc
@@ -73,7 +73,7 @@ TEST_CASE("Unit_hipStreamWaitEvent_UninitializedStream_Negative") {
 
   HIP_CHECK(hipEventCreate(&event));
 
-  HIP_CHECK_ERROR(hipStreamWaitEvent(stream, event, 0), hipErrorContextIsDestroyed);
+  HIP_CHECK_ERROR(hipStreamWaitEvent(stream, event, 0), hipErrorInvalidHandle);
 
   HIP_CHECK(hipEventDestroy(event));
 }
diff --git a/projects/hip-tests/catch/unit/streamperthread/hipStreamPerThread_DeviceReset.cc b/projects/hip-tests/catch/unit/streamperthread/hipStreamPerThread_DeviceReset.cc
index 31f4079f61..1366390a51 100644
--- a/projects/hip-tests/catch/unit/streamperthread/hipStreamPerThread_DeviceReset.cc
+++ b/projects/hip-tests/catch/unit/streamperthread/hipStreamPerThread_DeviceReset.cc
@@ -51,8 +51,11 @@ TEST_CASE("Unit_hipStreamPerThread_DeviceReset_1") {
 
   for (auto &th : threads) {
     th = std::thread(Copy_to_device);
-    th.detach();
   }
+  for (auto &th : threads) {
+    th.join();
+  }
+
   HIP_CHECK(hipDeviceReset());
 }