From be88e1dffc26c378d5eb924bf9b9c9b24eb1082d 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
---
 catch/unit/stream/hipStreamCreateWithFlags.cc          | 10 +++++-----
 catch/unit/stream/hipStreamWaitEvent.cc                |  2 +-
 .../streamperthread/hipStreamPerThread_DeviceReset.cc  |  5 ++++-
 3 files changed, 10 insertions(+), 7 deletions(-)

diff --git a/catch/unit/stream/hipStreamCreateWithFlags.cc b/catch/unit/stream/hipStreamCreateWithFlags.cc
index 5038037452..1ffb9c1523 100644
--- a/catch/unit/stream/hipStreamCreateWithFlags.cc
+++ b/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/catch/unit/stream/hipStreamWaitEvent.cc b/catch/unit/stream/hipStreamWaitEvent.cc
index 9877d548c1..32dbe9983b 100644
--- a/catch/unit/stream/hipStreamWaitEvent.cc
+++ b/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/catch/unit/streamperthread/hipStreamPerThread_DeviceReset.cc b/catch/unit/streamperthread/hipStreamPerThread_DeviceReset.cc
index 31f4079f61..1366390a51 100644
--- a/catch/unit/streamperthread/hipStreamPerThread_DeviceReset.cc
+++ b/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());
 }