From 7fa9b1746ce0f111be88ef4582a884e038ee9927 Mon Sep 17 00:00:00 2001
From: nives-vukovic <110852104+nives-vukovic@users.noreply.github.com>
Date: Mon, 10 Oct 2022 09:48:49 +0200
Subject: [PATCH 1/2] EXSWHTEC-40 - Implement additional Stream Management
 negative tests (#2922)

- Implement negative test for hipStream<Write/Wait>Value when stream is uninitialized
- Implement negative test for hipStreamSynchronize when stream is uninitialized
- Invalid flag negative test for hipStreamWriteValue API removed as flag parameter is not used
---
 .../catch/unit/stream/hipStreamSynchronize.cc | 13 +++++++
 tests/catch/unit/stream/hipStreamValue.cc     | 36 +++++++++++++++++--
 2 files changed, 47 insertions(+), 2 deletions(-)

diff --git a/tests/catch/unit/stream/hipStreamSynchronize.cc b/tests/catch/unit/stream/hipStreamSynchronize.cc
index 0555620b1f..a31546a52a 100644
--- a/tests/catch/unit/stream/hipStreamSynchronize.cc
+++ b/tests/catch/unit/stream/hipStreamSynchronize.cc
@@ -33,6 +33,19 @@ TEST_CASE("Unit_hipStreamSynchronize_EmptyStream") {
   HIP_CHECK(hipStreamDestroy(stream));
 }
 
+#if !HT_NVIDIA
+/**
+ * @brief Check that synchronization of uninitialized stream sets its status to
+ * hipErrorContextIsDestroyed
+ *
+ * Test removed for Nvidia devices because it returns unexpected error
+ */
+TEST_CASE("Unit_hipStreamSynchronize_UninitializedStream") {
+  hipStream_t stream{reinterpret_cast<hipStream_t>(0xFFFF)};
+  HIP_CHECK_ERROR(hipStreamSynchronize(stream), hipErrorContextIsDestroyed);
+}
+#endif
+
 #if HT_AMD /* Disabled because frequency based wait is timing out on nvidia platforms */
 
 /**
diff --git a/tests/catch/unit/stream/hipStreamValue.cc b/tests/catch/unit/stream/hipStreamValue.cc
index 164384acee..7a55360eff 100644
--- a/tests/catch/unit/stream/hipStreamValue.cc
+++ b/tests/catch/unit/stream/hipStreamValue.cc
@@ -499,7 +499,40 @@ TEST_CASE("Unit_hipStreamValue_Negative_InvalidMemory") {
   HIP_CHECK(hipStreamDestroy(stream));
 }
 
-TEMPLATE_TEST_CASE("Unit_hipStreamWaitValue_Negative_InvalidFlag", "", uint32_t, uint64_t) {
+TEMPLATE_TEST_CASE("Unit_hipStreamValue_Negative_UninitializedStream", "", uint32_t, uint64_t) {
+  if (!streamWaitValueSupported()) {
+    HipTest::HIP_SKIP_TEST("hipStreamWaitValue not supported on this device.");
+    return;
+  }
+
+  hipStream_t stream{reinterpret_cast<hipStream_t>(0xFFFF)};
+
+  // Allocate Host Memory
+  auto hostPtr = std::make_unique<TestType>();
+
+  // Register Host Memory
+  HIP_CHECK(hipHostRegister(hostPtr.get(), sizeof(TestType), 0));
+
+  // Set dummy data
+  *hostPtr = 0x0;
+
+  const auto compareOp = hipStreamWaitValueGte;
+  const auto expectedError = hipErrorContextIsDestroyed;
+
+  // Stream handle negative tests
+  SECTION("Invalid Stream handle for hipStreamWriteValue") {
+    HIP_CHECK_ERROR(writeFunc<TestType>(stream, hostPtr.get(), 0, writeFlag), expectedError);
+  }
+
+  SECTION("Invalid Stream handle for hipStreamWaitValue") {
+    HIP_CHECK_ERROR(waitFunc<TestType>(stream, hostPtr.get(), 0, compareOp), expectedError);
+  }
+
+  // Cleanup
+  HIP_CHECK(hipHostUnregister(hostPtr.get()));
+}
+
+TEMPLATE_TEST_CASE("Unit_hipStreamValue_Negative_InvalidFlag", "", uint32_t, uint64_t) {
   if (!streamWaitValueSupported()) {
     HipTest::HIP_SKIP_TEST("hipStreamWaitValue not supported on this device.");
     return;
@@ -518,7 +551,6 @@ TEMPLATE_TEST_CASE("Unit_hipStreamWaitValue_Negative_InvalidFlag", "", uint32_t,
   // Set dummy data
   *hostPtr = 0x0;
 
-  /* EXSWCPHIPT-96 */
   HIP_CHECK_ERROR(waitFunc<TestType>(stream, hostPtr.get(), 0, -1), hipErrorInvalidValue);
 
   // Cleanup

From 825225b232c86d2643689e3fa1844aa54b0f5dd9 Mon Sep 17 00:00:00 2001
From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com>
Date: Mon, 10 Oct 2022 15:25:37 +0530
Subject: [PATCH 2/2] SWDEV-343493 - hipcc ignores --rocm-path if env var such
 as ROCM_PATH is set (#2974)

Change-Id: I5723694158f35b73dc0c93e9dd9ee6666df1ca72
---
 bin/hipcc.pl | 21 ++++++++++++++++++++-
 1 file changed, 20 insertions(+), 1 deletion(-)

diff --git a/bin/hipcc.pl b/bin/hipcc.pl
index 76163d97b6..da9559be9e 100755
--- a/bin/hipcc.pl
+++ b/bin/hipcc.pl
@@ -52,6 +52,19 @@ if(scalar @ARGV == 0){
     exit(-1);
 }
 
+# retrieve --rocm-path hipcc option from command line.
+# We need to respect this over the env var ROCM_PATH for this compilation.
+sub get_rocm_path_option {
+  my $rocm_path="";
+  my @CLArgs = @ARGV;
+  foreach $arg (@CLArgs) {
+    if (index($arg,"--rocm-path=") != -1) {
+      ($rocm_path) = $arg=~ /=\s*(.*)\s*$/;
+    }
+  }
+  return $rocm_path;
+}
+
 $verbose = $ENV{'HIPCC_VERBOSE'} // 0;
 # Verbose: 0x1=commands, 0x2=paths, 0x4=hipcc args
 
@@ -88,12 +101,18 @@ sub delete_temp_dirs {
 }
 
 my $base_dir;
+my $rocmPath;
 BEGIN {
     $base_dir = dirname(Cwd::realpath(__FILE__) );
+    $rocmPath = get_rocm_path_option();
+    if ($rocmPath ne '') {
+      # --rocm-path takes precedence over ENV{ROCM_PATH}
+      $ENV{ROCM_PATH}=$rocmPath;
+    }
 }
 use lib "$base_dir/";
-use hipvars;
 
+use hipvars;
 $isWindows      =   $hipvars::isWindows;
 $HIP_RUNTIME    =   $hipvars::HIP_RUNTIME;
 $HIP_PLATFORM   =   $hipvars::HIP_PLATFORM;