From 05e52df05c698b24fe22d01bcf1863135751ac8e Mon Sep 17 00:00:00 2001
From: Rahul Manocha <rmanocha@amd.com>
Date: Mon, 21 Oct 2024 10:16:27 -0700
Subject: [PATCH] SWDEV-483107,SWDEV-492076,SWDEV-491371,SWDEV-484940  -
 hipStreamBeginCaptureToGraph catch exception fix

1) Catch Section should not be used inside host thread call.

Change-Id: I52c1bb80d6f92034dfbe2484d928e66eacf67d4c


[ROCm/hip-tests commit: decf563391860d9ae377929702621367cc04fdd5]
---
 .../graph/hipStreamBeginCaptureToGraph.cc     | 28 +++++++++++--------
 1 file changed, 16 insertions(+), 12 deletions(-)

diff --git a/projects/hip-tests/catch/unit/graph/hipStreamBeginCaptureToGraph.cc b/projects/hip-tests/catch/unit/graph/hipStreamBeginCaptureToGraph.cc
index 91cc39ad95..20b1d5b13c 100644
--- a/projects/hip-tests/catch/unit/graph/hipStreamBeginCaptureToGraph.cc
+++ b/projects/hip-tests/catch/unit/graph/hipStreamBeginCaptureToGraph.cc
@@ -1279,17 +1279,10 @@ static std::atomic<int> retValG(1);
 void threadCaptureExec(int *A_d, int *B_d, int *C_d,
                        int *A_h, int *B_h, int *C_h,
                        hipStream_t *stream1, hipStream_t *stream2,
-                       hipGraph_t *graph) {
+                       hipGraph_t *graph, bool verifyStreamSync) {
   bool ret = false;
-  SECTION("Verify after hipGraphExecDestroy()") {
-    ret = CaptureStreamAndLaunchGraph(A_d, B_d, C_d, A_h, B_h, C_h, hipStreamCaptureModeRelaxed,
-          *stream1, *stream2, *graph, false);
-  }
-  SECTION("Verify after hipStreamSynchronize()") {
-    ret = CaptureStreamAndLaunchGraph(A_d, B_d, C_d, A_h, B_h, C_h, hipStreamCaptureModeRelaxed,
-          *stream1, *stream2, *graph, true);
-  }
-
+  ret = CaptureStreamAndLaunchGraph(A_d, B_d, C_d, A_h, B_h, C_h, hipStreamCaptureModeRelaxed,
+          *stream1, *stream2, *graph, verifyStreamSync);
   int val = 0;
   if (ret) {
     val = 1;
@@ -1329,15 +1322,26 @@ TEST_CASE("Unit_hipStreamBeginCaptureToGraph_IndepGraphsThreads") {
   HIP_CHECK(hipMalloc(&C2_d, Nbytes));
 
   // Capture an independent graph from stream
+  bool verifyStreamSync = false;
+
+  SECTION("Verify after hipGraphExecDestroy()") {
+    verifyStreamSync = false;
+  }
+
+  SECTION("Verify after hipStreamSynchronize()") {
+    verifyStreamSync = true;
+  }
+
   std::thread thread1(threadCaptureExec, A1_d, B1_d, C1_d,
                       A1_h.data(), B1_h.data(), C1_h.data(),
-                      &stream1, &stream2, &graph1);
+                      &stream1, &stream2, &graph1, verifyStreamSync);
 
   std::thread thread2(threadCaptureExec, A2_d, B2_d, C2_d,
                       A2_h.data(), B2_h.data(), C2_h.data(),
-                      &stream3, &stream4, &graph2);
+                      &stream3, &stream4, &graph2, verifyStreamSync);
   thread1.join();
   thread2.join();
+
   REQUIRE(retValG.load() == 1);
   HIP_CHECK(hipStreamDestroy(stream1));
   HIP_CHECK(hipStreamDestroy(stream2));