From d8f39ab8d23a3528be5a945187ebf6aff6580780 Mon Sep 17 00:00:00 2001
From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com>
Date: Tue, 24 Jan 2023 04:28:47 +0530
Subject: [PATCH 1/3] SWDEV-364604 - Add hipEventDisableReleaseToSystem (#3132)

Add the above flag to be used with hipEventCreateWithFlags API

Change-Id: I4dcd259e579546178865879f69b78410f5ad1b57
---
 include/hip/hip_runtime_api.h | 7 +++++++
 1 file changed, 7 insertions(+)

diff --git a/include/hip/hip_runtime_api.h b/include/hip/hip_runtime_api.h
index e8f82ed937..8a466e75d7 100644
--- a/include/hip/hip_runtime_api.h
+++ b/include/hip/hip_runtime_api.h
@@ -585,6 +585,10 @@ enum hipLimit_t {
 /** Event can support IPC. hipEventDisableTiming also must be set.*/
 #define hipEventInterprocess 0x4
 
+/** Disable any system scope releases for events. May improve performance.
+ * The flag is a no-op on CUDA platforms.*/
+#define hipEventDisableReleaseToSystem 0x20000000
+
 /** Use a device-scope release when recording this event. This flag is useful to obtain more
  * precise timings of commands between events.  The flag is a no-op on CUDA platforms.*/
 #define hipEventReleaseToDevice  0x40000000
@@ -2416,6 +2420,9 @@ hipError_t hipStreamWriteValue64(hipStream_t stream, void* ptr, uint64_t value,
  would not record profiling data and provide best performance if used for synchronization.
  * #hipEventInterprocess : The event can be used as an interprocess event. hipEventDisableTiming
  flag also must be set when hipEventInterprocess flag is set.
+ * #hipEventDisableReleaseToSystem : Disable releasing any cached memory to system scope. This may
+ improve performance but device memory may not be visible to the host and other devices
+ if this flag is set.
  *
  * @returns #hipSuccess, #hipErrorNotInitialized, #hipErrorInvalidValue,
  #hipErrorLaunchFailure, #hipErrorOutOfMemory

From a388eb1e6ffa19184a10078353bd8e9dbd739a60 Mon Sep 17 00:00:00 2001
From: Siu Chi Chan <siuchi.chan@amd.com>
Date: Tue, 24 Jan 2023 06:42:52 +0000
Subject: [PATCH 2/3] SWDEV-355608 - deprecate/cleanup hipcc link flags (#3128)

- deprecate -use-staticlib, -use-sharedlib which no longer provide any
functional values
- use --hip-link instead of specifying the HIP runtime by name when
linking
- fix linker option bug in HIT test's cmake
- update build options for unit tests requiring pthread or rt

Change-Id: Ib49978773c80fb40c71dc52b050ce921943ee3e4
---
 bin/hipcc.pl                                  | 43 ++++++-------------
 samples/0_Intro/square/README.md              |  2 +-
 tests/catch/multiproc/CMakeLists.txt          |  6 ++-
 tests/catch/unit/device/CMakeLists.txt        |  7 ++-
 tests/catch/unit/deviceLib/CMakeLists.txt     |  3 +-
 tests/catch/unit/errorHandling/CMakeLists.txt |  7 ++-
 tests/catch/unit/event/CMakeLists.txt         |  4 ++
 tests/catch/unit/graph/CMakeLists.txt         |  7 ++-
 tests/catch/unit/memory/CMakeLists.txt        |  5 ++-
 tests/catch/unit/multiThread/CMakeLists.txt   |  7 ++-
 tests/catch/unit/stream/CMakeLists.txt        |  7 ++-
 .../catch/unit/streamperthread/CMakeLists.txt |  7 ++-
 tests/hit/HIT.cmake                           |  2 +-
 .../src/deviceLib/hip_threadfence_system.cpp  |  2 +-
 tests/src/hipEnvVarDriver.cpp                 |  2 +-
 tests/src/ipc/hipMultiProcIpcEvent.cpp        |  2 +-
 tests/src/ipc/hipMultiProcIpcMem.cpp          |  2 +-
 .../event/hipEventMultiThreaded.cpp           |  2 +-
 .../runtimeApi/memory/hipIpcMemAccessTest.cpp |  2 +-
 .../memory/hipMallocConcurrency.cpp           |  2 +-
 tests/src/runtimeApi/memory/hipMemcpy.cpp     |  2 +-
 .../memory/hipMemcpyNegativeMThrdMSize.cpp    |  2 +-
 .../memory/hipMemcpyWithStreamMultiThread.cpp |  2 +-
 .../hipMemset2DAsyncMultiThreadAndKernel.cpp  |  2 +-
 .../memory/hipMemset3DRegressMultiThread.cpp  |  2 +-
 .../memory/hipMemsetAsyncMultiThread.cpp      |  2 +-
 .../hipMultiMemcpyMultiThrdMultiStrm.cpp      |  2 +-
 .../memory/hipMultiMemcpyMultiThread.cpp      |  2 +-
 .../memory/hipPointerAttributes.cpp           |  2 +-
 .../hipModuleLoadDataMultThreadOnMultGPU.cpp  |  2 +-
 .../module/hipModuleLoadDataMultThreaded.cpp  |  2 +-
 .../module/hipModuleLoadMultiThreaded.cpp     |  2 +-
 .../module/hipModuleTexture2dDrv.cpp          |  2 +-
 .../multiThread/hipMultiThreadDevice.cpp      |  2 +-
 .../multiThread/hipMultiThreadStreams1.cpp    |  2 +-
 .../multiThread/hipMultiThreadStreams2.cpp    |  2 +-
 .../p2p/hipP2pLinkTypeAndHopFunc.cpp          |  2 +-
 .../stream/hipStreamACb_MStrm_Mgpu.cpp        |  2 +-
 .../stream/hipStreamACb_MultiThread.cpp       |  2 +-
 .../stream/hipStreamAddCallbackCatch.cpp      |  2 +-
 .../streamOperations/hipstream_operations.cpp |  2 +-
 .../cache_coherency_cpu_gpu.cpp               |  2 +-
 42 files changed, 93 insertions(+), 72 deletions(-)

diff --git a/bin/hipcc.pl b/bin/hipcc.pl
index 431701fb50..6898cd0bba 100644
--- a/bin/hipcc.pl
+++ b/bin/hipcc.pl
@@ -199,10 +199,6 @@ if ($HIP_PLATFORM eq "amd") {
         print ("HIP_CLANG_RT_LIB=$HIP_CLANG_RT_LIB\n");
     }
 
-    $HIPLDFLAGS .= " -L\"$HIP_LIB_PATH\"";
-    if ($isWindows) {
-      $HIPLDFLAGS .= " -lamdhip64";
-    }
     if ($HIP_CLANG_HCC_COMPAT_MODE) {
         ## Allow __fp16 as function parameter and return type.
         $HIPCXXFLAGS .= " -Xclang -fallow-half-arguments-and-returns -D__HIP_HCC_COMPAT_MODE__=1";
@@ -245,8 +241,6 @@ my $printCXXFlags = 0;      # print HIPCXXFLAGS
 my $printLDFlags = 0;       # print HIPLDFLAGS
 my $runCmd = 1;
 my $buildDeps = 0;
-my $linkType = 1;
-my $setLinkType = 0;
 my $hsacoVersion = 0;
 my $funcSupp = 0;      # enable function support
 my $rdc = 0;           # whether -fgpu-rdc is on
@@ -361,16 +355,11 @@ foreach $arg (@ARGV)
         $compileOnly = 1;
         $buildDeps = 1;
     }
-    if(($trimarg eq '-use-staticlib') and ($setLinkType eq 0))
-    {
-        $linkType = 0;
-        $setLinkType = 1;
-        $swallowArg = 1;
+    if($trimarg eq '-use-staticlib') {
+        print "Warning: The -use-staticlib option has been deprecated and is no longer needed.\n"
     }
-    if(($trimarg eq '-use-sharedlib') and ($setLinkType eq 0))
-    {
-        $linkType = 1;
-        $setLinkType = 1;
+    if($trimarg eq '-use-sharedlib') {
+        print "Warning: The -use-sharedlib option has been deprecated and is no longer needed.\n"
     }
     if($arg =~ m/^-O/)
     {
@@ -558,12 +547,6 @@ if ($buildDeps and $HIP_PLATFORM eq 'amd') {
     $HIPCXXFLAGS .= " --cuda-host-only";
 }
 
-# Add --hip-link only if it is compile only and -fgpu-rdc is on.
-if ($rdc and !$compileOnly and $HIP_PLATFORM eq 'amd') {
-    $HIPLDFLAGS .= " --hip-link";
-    $HIPLDFLAGS .= $HIPLDARCHFLAGS;
-}
-
 # hipcc currrently requires separate compilation of source files, ie it is not possible to pass
 # CPP files combined with .O files
 # Reason is that NVCC uses the file extension to determine whether to compile in CUDA mode or
@@ -588,18 +571,16 @@ if ($HIP_PLATFORM eq "amd") {
             $HIPCXXFLAGS .= " --hip-device-lib-path=\"$DEVICE_LIB_PATH\"";
         }
     }
-    if (not $isWindows) {
-        $HIPLDFLAGS .= " -lgcc_s -lgcc -lpthread -lm -lrt";
-    }
 
-    if (not $isWindows  and not $compileOnly) {
-      if ($linkType eq 0) {
-        $toolArgs = " -L$HIP_LIB_PATH -lamdhip64 -L$ROCM_PATH/lib -lhsa-runtime64 -ldl -lnuma " . ${toolArgs};
-      } else {
-        $toolArgs = ${toolArgs} . " -Wl,-rpath=$HIP_LIB_PATH:$ROCM_PATH/lib -lamdhip64 ";
-      }
+    if (!$compileOnly) {
+        $HIPLDFLAGS .= " --hip-link";
+        if ($rdc) {
+            $HIPLDFLAGS .= $HIPLDARCHFLAGS;
+        }
+        if (not $isWindows) {
+            $HIPLDFLAGS .= " --rtlib=compiler-rt -unwindlib=libgcc";
 
-      $toolArgs .= " -L$HIP_CLANG_RT_LIB -lclang_rt.builtins-x86_64 "
+        }
     }
 }
 
diff --git a/samples/0_Intro/square/README.md b/samples/0_Intro/square/README.md
index 807f08754e..e06fbceda1 100644
--- a/samples/0_Intro/square/README.md
+++ b/samples/0_Intro/square/README.md
@@ -21,7 +21,7 @@ $ cd ~/hip/samples/0_Intro/square
 $ make
 /opt/rocm/hip/bin/hipify-perl square.cu > square.cpp
 /opt/rocm/hip/bin/hipcc  square.cpp -o square.out
-/opt/rocm/hip/bin/hipcc -use-staticlib  square.cpp -o square.out.static
+/opt/rocm/hip/bin/hipcc  square.cpp -o square.out
 ```
 - Execute file
 ```
diff --git a/tests/catch/multiproc/CMakeLists.txt b/tests/catch/multiproc/CMakeLists.txt
index 5485ee9ca5..58bec6f70f 100644
--- a/tests/catch/multiproc/CMakeLists.txt
+++ b/tests/catch/multiproc/CMakeLists.txt
@@ -19,6 +19,9 @@ set(LINUX_TEST_SRC
 )
 
 add_custom_target(dummy_kernel.code COMMAND ${CMAKE_CXX_COMPILER} --genco ${CMAKE_CURRENT_SOURCE_DIR}/dummy_kernel.cpp -o ${CMAKE_CURRENT_BINARY_DIR}/../multiproc/dummy_kernel.code -I${CMAKE_CURRENT_SOURCE_DIR}/../../../../include/ -I${CMAKE_CURRENT_SOURCE_DIR}/../../include)
+if (UNIX)
+  set(__LINKER_LIBS__ pthread)
+endif()
 
 # the last argument linker libraries is required for this test but optional to the function
 if(HIP_PLATFORM MATCHES "nvidia")
@@ -30,7 +33,6 @@ elseif(HIP_PLATFORM MATCHES "amd")
 hip_add_exe_to_target(NAME MultiProc
                       TEST_SRC ${LINUX_TEST_SRC}
                       TEST_TARGET_NAME build_tests
-                      LINKER_LIBS ${CMAKE_DL_LIBS})
+                      LINKER_LIBS ${CMAKE_DL_LIBS} ${__LINKER_LIBS__})
 endif()
 add_dependencies(build_tests dummy_kernel.code)
-
diff --git a/tests/catch/unit/device/CMakeLists.txt b/tests/catch/unit/device/CMakeLists.txt
index 1e26944d0f..b9fe3726f2 100644
--- a/tests/catch/unit/device/CMakeLists.txt
+++ b/tests/catch/unit/device/CMakeLists.txt
@@ -43,10 +43,15 @@ set_source_files_properties(hipDeviceGetP2PAttribute.cc PROPERTIES COMPILE_FLAGS
 add_executable(getDeviceCount EXCLUDE_FROM_ALL getDeviceCount_exe.cc)
 add_executable(hipDeviceGetP2PAttribute EXCLUDE_FROM_ALL hipDeviceGetP2PAttribute_exe.cc)
 
+if (UNIX)
+  set(__LINKER_LIBS__ pthread)
+endif()
+
 hip_add_exe_to_target(NAME DeviceTest
                       TEST_SRC ${TEST_SRC}
                       TEST_TARGET_NAME build_tests
-                      COMPILE_OPTIONS -std=c++14)
+                      COMPILE_OPTIONS -std=c++14
+                      LINKER_LIBS ${__LINKER_LIBS__})
 
 add_dependencies(DeviceTest getDeviceCount)
 add_dependencies(DeviceTest hipDeviceGetP2PAttribute)
diff --git a/tests/catch/unit/deviceLib/CMakeLists.txt b/tests/catch/unit/deviceLib/CMakeLists.txt
index c54cd68c08..3d882a6644 100644
--- a/tests/catch/unit/deviceLib/CMakeLists.txt
+++ b/tests/catch/unit/deviceLib/CMakeLists.txt
@@ -18,6 +18,7 @@ set(TEST_SRC
 if(UNIX)
     set(TEST_SRC ${TEST_SRC}
         deviceAllocation.cc)
+    set(__LINKER_LIBS__ pthread)
 endif()
 
 # AMD only tests
@@ -86,7 +87,7 @@ endif()
     hip_add_exe_to_target(NAME UnitDeviceTests
                       TEST_SRC ${TEST_SRC}
                       TEST_TARGET_NAME build_tests
-                      LINKER_LIBS hiprtc)
+                      LINKER_LIBS hiprtc ${__LINKER_LIBS__})
 elseif(HIP_PLATFORM MATCHES "nvidia")
     hip_add_exe_to_target(NAME UnitDeviceTests
                       TEST_SRC ${TEST_SRC}
diff --git a/tests/catch/unit/errorHandling/CMakeLists.txt b/tests/catch/unit/errorHandling/CMakeLists.txt
index 78933fd3f8..1b89eb066e 100644
--- a/tests/catch/unit/errorHandling/CMakeLists.txt
+++ b/tests/catch/unit/errorHandling/CMakeLists.txt
@@ -6,7 +6,12 @@ set(TEST_SRC
     hipPeekAtLastError.cc
 )
 
+if (UNIX)
+  set(__LINKER_LIBS__ pthread)
+endif()
+
 hip_add_exe_to_target(NAME ErrorHandlingTest
                       TEST_SRC ${TEST_SRC}
                       TEST_TARGET_NAME build_tests
-                      COMPILE_OPTIONS -std=c++14)
\ No newline at end of file
+		      LINKER_LIBS ${__LINKER_LIBS__}
+		      COMPILE_OPTIONS -std=c++14)
diff --git a/tests/catch/unit/event/CMakeLists.txt b/tests/catch/unit/event/CMakeLists.txt
index b4d5f2c3b6..42b8699ba3 100644
--- a/tests/catch/unit/event/CMakeLists.txt
+++ b/tests/catch/unit/event/CMakeLists.txt
@@ -20,6 +20,10 @@ if(HIP_PLATFORM MATCHES "amd")
   set(TEST_SRC ${TEST_SRC} ${AMD_SRC})
 endif()
 
+if (UNIX)
+  set(__LINKER_LIBS__ pthread)
+endif()
+
 hip_add_exe_to_target(NAME EventTest
                       TEST_SRC ${TEST_SRC}
                       TEST_TARGET_NAME build_tests)
diff --git a/tests/catch/unit/graph/CMakeLists.txt b/tests/catch/unit/graph/CMakeLists.txt
index e2b66872da..a28810a7e8 100644
--- a/tests/catch/unit/graph/CMakeLists.txt
+++ b/tests/catch/unit/graph/CMakeLists.txt
@@ -83,6 +83,11 @@ set(TEST_SRC
   hipUserObjectCreate.cc
 )
 
+if (UNIX)
+  set(__LINKER_LIBS__ pthread)
+endif()
+
 hip_add_exe_to_target(NAME GraphsTest
                       TEST_SRC ${TEST_SRC}
-                      TEST_TARGET_NAME build_tests)
+                      TEST_TARGET_NAME build_tests
+                      LINKER_LIBS ${__LINKER_LIBS__})
diff --git a/tests/catch/unit/memory/CMakeLists.txt b/tests/catch/unit/memory/CMakeLists.txt
index 1b4cef3698..74165c5db0 100644
--- a/tests/catch/unit/memory/CMakeLists.txt
+++ b/tests/catch/unit/memory/CMakeLists.txt
@@ -182,8 +182,11 @@ if(UNIX)
     set(TEST_SRC ${TEST_SRC}
                  hipHmmOvrSubscriptionTst.cc
                  hipMemoryAllocateCoherent.cc)
+                 
+    set(__LINKER_LIBS__ pthread)
 endif()
 
 hip_add_exe_to_target(NAME MemoryTest
                       TEST_SRC ${TEST_SRC}
-                      TEST_TARGET_NAME build_tests)
+                      TEST_TARGET_NAME build_tests
+                      LINKER_LIBS ${__LINKER_LIBS__})
diff --git a/tests/catch/unit/multiThread/CMakeLists.txt b/tests/catch/unit/multiThread/CMakeLists.txt
index 32abf0f5f3..515832ef95 100644
--- a/tests/catch/unit/multiThread/CMakeLists.txt
+++ b/tests/catch/unit/multiThread/CMakeLists.txt
@@ -5,6 +5,11 @@ set(TEST_SRC
   hipMultiThreadStreams2.cc
 )
 
+if (UNIX)
+  set(__LINKER_LIBS__ pthread)
+endif()
+
 hip_add_exe_to_target(NAME MultiThreadTest
                       TEST_SRC ${TEST_SRC}
-                      TEST_TARGET_NAME build_tests)
+                      TEST_TARGET_NAME build_tests
+                      LINKER_LIBS ${__LINKER_LIBS__})
diff --git a/tests/catch/unit/stream/CMakeLists.txt b/tests/catch/unit/stream/CMakeLists.txt
index 548bb2e52c..0a8d0b1a3b 100644
--- a/tests/catch/unit/stream/CMakeLists.txt
+++ b/tests/catch/unit/stream/CMakeLists.txt
@@ -45,7 +45,12 @@ set(TEST_SRC
     # set_source_files_properties(hipStreamAttachMemAsync.cc PROPERTIES COMPILE_FLAGS -std=c++17)
 endif()
 
+if (UNIX)
+  set(__LINKER_LIBS__ pthread)
+endif()
+
 hip_add_exe_to_target(NAME StreamTest
                       TEST_SRC ${TEST_SRC}
                       TEST_TARGET_NAME build_tests
-                      COMPILE_OPTIONS -std=c++17)
+                      COMPILE_OPTIONS -std=c++17
+                      LINKER_LIBS ${__LINKER_LIBS__})
diff --git a/tests/catch/unit/streamperthread/CMakeLists.txt b/tests/catch/unit/streamperthread/CMakeLists.txt
index dd5b54dabc..51e2255fa3 100644
--- a/tests/catch/unit/streamperthread/CMakeLists.txt
+++ b/tests/catch/unit/streamperthread/CMakeLists.txt
@@ -7,6 +7,11 @@ set(TEST_SRC
   hipStreamPerThrdTsts.cc
 )
 
+if (UNIX)
+  set(__LINKER_LIBS__ pthread)
+endif()
+
 hip_add_exe_to_target(NAME StreamPerThreadTest
                       TEST_SRC ${TEST_SRC}
-                      TEST_TARGET_NAME build_tests)
+                      TEST_TARGET_NAME build_tests
+                      LINKER_LIBS ${__LINKER_LIBS__})
diff --git a/tests/hit/HIT.cmake b/tests/hit/HIT.cmake
index 68c3df3c4f..e1c8fcbc4b 100755
--- a/tests/hit/HIT.cmake
+++ b/tests/hit/HIT.cmake
@@ -354,7 +354,7 @@ macro(HIT_ADD_FILES _config _dir _label _parent)
                 set_source_files_properties(${_sources} PROPERTIES HIP_SOURCE_PROPERTY_FORMAT 1)
                 hip_reset_flags()
                 hip_add_executable(${target} ${_sources} HIPCC_OPTIONS ${_hipcc_options} CLANG_OPTIONS ${_clang_options} NVCC_OPTIONS ${_nvcc_options} EXCLUDE_FROM_ALL)
-                target_link_libraries(${target} PRIVATE ${_link_options})
+                target_link_options(${target} PRIVATE ${_link_options})
                 set_target_properties(${target} PROPERTIES OUTPUT_NAME ${_target} RUNTIME_OUTPUT_DIRECTORY ${_label} LINK_DEPENDS "${HIP_LIB_FILES}")
                 add_dependencies(${_parent} ${target})
                 foreach(_dependency ${_depends})
diff --git a/tests/src/deviceLib/hip_threadfence_system.cpp b/tests/src/deviceLib/hip_threadfence_system.cpp
index 097dae2273..11e84a3b87 100644
--- a/tests/src/deviceLib/hip_threadfence_system.cpp
+++ b/tests/src/deviceLib/hip_threadfence_system.cpp
@@ -21,7 +21,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/hipEnvVarDriver.cpp b/tests/src/hipEnvVarDriver.cpp
index 2d303e08fe..bd5de426d6 100644
--- a/tests/src/hipEnvVarDriver.cpp
+++ b/tests/src/hipEnvVarDriver.cpp
@@ -16,7 +16,7 @@ DAMAGES OR OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE,
 OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. */
 
 /* HIT_START
- * BUILD: %t %s test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s test_common.cpp LINK_OPTIONS -lpthread NVCC_OPTIONS -std=c++11
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/ipc/hipMultiProcIpcEvent.cpp b/tests/src/ipc/hipMultiProcIpcEvent.cpp
index 418b68a5ed..b06c010a01 100644
--- a/tests/src/ipc/hipMultiProcIpcEvent.cpp
+++ b/tests/src/ipc/hipMultiProcIpcEvent.cpp
@@ -18,7 +18,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../test_common.cpp
+ * BUILD: %t %s ../test_common.cpp LINK_OPTIONS -lrt
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/ipc/hipMultiProcIpcMem.cpp b/tests/src/ipc/hipMultiProcIpcMem.cpp
index 3829853774..ec579dfd7b 100644
--- a/tests/src/ipc/hipMultiProcIpcMem.cpp
+++ b/tests/src/ipc/hipMultiProcIpcMem.cpp
@@ -18,7 +18,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../test_common.cpp
+ * BUILD: %t %s ../test_common.cpp LINK_OPTIONS -lrt
  * TEST: %t --N 4
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/event/hipEventMultiThreaded.cpp b/tests/src/runtimeApi/event/hipEventMultiThreaded.cpp
index 65f2ff1851..2bf5a78969 100644
--- a/tests/src/runtimeApi/event/hipEventMultiThreaded.cpp
+++ b/tests/src/runtimeApi/event/hipEventMultiThreaded.cpp
@@ -21,7 +21,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp
+ * BUILD: %t %s ../../test_common.cpp LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/memory/hipIpcMemAccessTest.cpp b/tests/src/runtimeApi/memory/hipIpcMemAccessTest.cpp
index cd9809ded5..94142a76bd 100644
--- a/tests/src/runtimeApi/memory/hipIpcMemAccessTest.cpp
+++ b/tests/src/runtimeApi/memory/hipIpcMemAccessTest.cpp
@@ -18,7 +18,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp
+ * BUILD: %t %s ../../test_common.cpp LINK_OPTIONS -lrt -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/memory/hipMallocConcurrency.cpp b/tests/src/runtimeApi/memory/hipMallocConcurrency.cpp
index 4bfd65a7f9..44a4c5d086 100644
--- a/tests/src/runtimeApi/memory/hipMallocConcurrency.cpp
+++ b/tests/src/runtimeApi/memory/hipMallocConcurrency.cpp
@@ -58,7 +58,7 @@ Testcase Scenarios :
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11 LINK_OPTIONS -lpthread
  * TEST_NAMED: %t hipMalloc_ArgValidation  --tests 1
  * TEST_NAMED: %t hipMalloc_LoopRegression_AllocFreeCycle --tests 2
  * TEST_NAMED: %t hipMalloc_LoopRegression_AllocPool --tests 3
diff --git a/tests/src/runtimeApi/memory/hipMemcpy.cpp b/tests/src/runtimeApi/memory/hipMemcpy.cpp
index c6bb98be0c..6735616b0b 100644
--- a/tests/src/runtimeApi/memory/hipMemcpy.cpp
+++ b/tests/src/runtimeApi/memory/hipMemcpy.cpp
@@ -21,7 +21,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST_NAMED: %t hipMemcpy-modes --tests 0x1
  * TEST_NAMED: %t hipMemcpy-size --tests 0x6
  * TEST_NAMED: %t hipMemcpy-dev-offsets --tests 0x10
diff --git a/tests/src/runtimeApi/memory/hipMemcpyNegativeMThrdMSize.cpp b/tests/src/runtimeApi/memory/hipMemcpyNegativeMThrdMSize.cpp
index 3f103915ee..53dae66758 100644
--- a/tests/src/runtimeApi/memory/hipMemcpyNegativeMThrdMSize.cpp
+++ b/tests/src/runtimeApi/memory/hipMemcpyNegativeMThrdMSize.cpp
@@ -24,7 +24,7 @@ THE SOFTWARE.
 // of 8 hipmemcpy apis
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11 LINK_OPTIONS -lpthread
  * TEST_NAMED: %t hipMemcpyNegativeMThrdMSize_Negative_tests  --tests 1
  * TEST_NAMED: %t hipMemcpyNegativeMThrdMSize_MultiThread_tests --tests 2
  * TEST_NAMED: %t hipMemcpyNegativeMThrdMSize_MultiSize_singleType --tests 3 --memcpyPeersOnly 0 --testAllTypes 0
diff --git a/tests/src/runtimeApi/memory/hipMemcpyWithStreamMultiThread.cpp b/tests/src/runtimeApi/memory/hipMemcpyWithStreamMultiThread.cpp
index d5580a4f6d..ddd9c853ea 100644
--- a/tests/src/runtimeApi/memory/hipMemcpyWithStreamMultiThread.cpp
+++ b/tests/src/runtimeApi/memory/hipMemcpyWithStreamMultiThread.cpp
@@ -24,7 +24,7 @@ THE SOFTWARE.
  */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/memory/hipMemset2DAsyncMultiThreadAndKernel.cpp b/tests/src/runtimeApi/memory/hipMemset2DAsyncMultiThreadAndKernel.cpp
index aed1efbba8..09057677ac 100644
--- a/tests/src/runtimeApi/memory/hipMemset2DAsyncMultiThreadAndKernel.cpp
+++ b/tests/src/runtimeApi/memory/hipMemset2DAsyncMultiThreadAndKernel.cpp
@@ -24,7 +24,7 @@
 //
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp
+ * BUILD: %t %s ../../test_common.cpp LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/memory/hipMemset3DRegressMultiThread.cpp b/tests/src/runtimeApi/memory/hipMemset3DRegressMultiThread.cpp
index 80d92e9355..7ae81ad695 100644
--- a/tests/src/runtimeApi/memory/hipMemset3DRegressMultiThread.cpp
+++ b/tests/src/runtimeApi/memory/hipMemset3DRegressMultiThread.cpp
@@ -34,7 +34,7 @@ Testcase Scenarios :
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t --tests 1
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/memory/hipMemsetAsyncMultiThread.cpp b/tests/src/runtimeApi/memory/hipMemsetAsyncMultiThread.cpp
index a7893d7bd7..7f6d379dc2 100644
--- a/tests/src/runtimeApi/memory/hipMemsetAsyncMultiThread.cpp
+++ b/tests/src/runtimeApi/memory/hipMemsetAsyncMultiThread.cpp
@@ -22,7 +22,7 @@
  */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp
+ * BUILD: %t %s ../../test_common.cpp LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/memory/hipMultiMemcpyMultiThrdMultiStrm.cpp b/tests/src/runtimeApi/memory/hipMultiMemcpyMultiThrdMultiStrm.cpp
index 7388f44b9f..a1eebf20b1 100644
--- a/tests/src/runtimeApi/memory/hipMultiMemcpyMultiThrdMultiStrm.cpp
+++ b/tests/src/runtimeApi/memory/hipMultiMemcpyMultiThrdMultiStrm.cpp
@@ -21,7 +21,7 @@
 // and also launch hipMemcpyAsync() api on the same stream. This test case is simulate the scenario
 // reported in SWDEV-181598.
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/memory/hipMultiMemcpyMultiThread.cpp b/tests/src/runtimeApi/memory/hipMultiMemcpyMultiThread.cpp
index 2a690b9832..ea8f32c965 100644
--- a/tests/src/runtimeApi/memory/hipMultiMemcpyMultiThread.cpp
+++ b/tests/src/runtimeApi/memory/hipMultiMemcpyMultiThread.cpp
@@ -21,7 +21,7 @@
 // and also launch hipMemcpyAsync() api. This test case is simulate the scenario
 // reported in SWDEV-181598.
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/memory/hipPointerAttributes.cpp b/tests/src/runtimeApi/memory/hipPointerAttributes.cpp
index 9666dd4693..a85dc156da 100644
--- a/tests/src/runtimeApi/memory/hipPointerAttributes.cpp
+++ b/tests/src/runtimeApi/memory/hipPointerAttributes.cpp
@@ -21,7 +21,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp
+ * BUILD: %t %s ../../test_common.cpp LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/module/hipModuleLoadDataMultThreadOnMultGPU.cpp b/tests/src/runtimeApi/module/hipModuleLoadDataMultThreadOnMultGPU.cpp
index f8b436068b..dadcc2d222 100644
--- a/tests/src/runtimeApi/module/hipModuleLoadDataMultThreadOnMultGPU.cpp
+++ b/tests/src/runtimeApi/module/hipModuleLoadDataMultThreadOnMultGPU.cpp
@@ -18,7 +18,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/module/hipModuleLoadDataMultThreaded.cpp b/tests/src/runtimeApi/module/hipModuleLoadDataMultThreaded.cpp
index 537195c5fb..4b03c3b189 100644
--- a/tests/src/runtimeApi/module/hipModuleLoadDataMultThreaded.cpp
+++ b/tests/src/runtimeApi/module/hipModuleLoadDataMultThreaded.cpp
@@ -18,7 +18,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 EXCLUDE_HIP_PLATFORM nvidia
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 EXCLUDE_HIP_PLATFORM nvidia LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/module/hipModuleLoadMultiThreaded.cpp b/tests/src/runtimeApi/module/hipModuleLoadMultiThreaded.cpp
index dbd41900ce..24f0f782cf 100644
--- a/tests/src/runtimeApi/module/hipModuleLoadMultiThreaded.cpp
+++ b/tests/src/runtimeApi/module/hipModuleLoadMultiThreaded.cpp
@@ -19,7 +19,7 @@ THE SOFTWARE.
 
 /* HIT_START
  * BUILD_CMD: empty_kernel.code %hc --genco %S/empty_kernel.cpp -o empty_kernel.code
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/module/hipModuleTexture2dDrv.cpp b/tests/src/runtimeApi/module/hipModuleTexture2dDrv.cpp
index 67c62f412e..94e753a585 100644
--- a/tests/src/runtimeApi/module/hipModuleTexture2dDrv.cpp
+++ b/tests/src/runtimeApi/module/hipModuleTexture2dDrv.cpp
@@ -23,7 +23,7 @@ THE SOFTWARE.
 /* Tests 6 and 7 are skipped for CUDA 11.2 due to cuda runtime issues */
 /* HIT_START
  * BUILD_CMD: tex2d_kernel.code %hc --genco %S/tex2d_kernel.cpp -o tex2d_kernel.code
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t --tests 0x01
  * TEST: %t --tests 0x02
  * TEST: %t --tests 0x03
diff --git a/tests/src/runtimeApi/multiThread/hipMultiThreadDevice.cpp b/tests/src/runtimeApi/multiThread/hipMultiThreadDevice.cpp
index af48d5a367..18bf9c0bd2 100644
--- a/tests/src/runtimeApi/multiThread/hipMultiThreadDevice.cpp
+++ b/tests/src/runtimeApi/multiThread/hipMultiThreadDevice.cpp
@@ -1,5 +1,5 @@
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST_NAMED: %t hipMultiThreadDevice-serial --tests 0x1
  * TEST_NAMED: %t hipMultiThreadDevice-pyramid --tests 0x4
  * TEST_NAMED: %t hipMultiThreadDevice-nearzero --tests 0x10
diff --git a/tests/src/runtimeApi/multiThread/hipMultiThreadStreams1.cpp b/tests/src/runtimeApi/multiThread/hipMultiThreadStreams1.cpp
index f1c683927b..b8fca4d02f 100644
--- a/tests/src/runtimeApi/multiThread/hipMultiThreadStreams1.cpp
+++ b/tests/src/runtimeApi/multiThread/hipMultiThreadStreams1.cpp
@@ -21,7 +21,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/multiThread/hipMultiThreadStreams2.cpp b/tests/src/runtimeApi/multiThread/hipMultiThreadStreams2.cpp
index 2f2367775f..752ffd870d 100644
--- a/tests/src/runtimeApi/multiThread/hipMultiThreadStreams2.cpp
+++ b/tests/src/runtimeApi/multiThread/hipMultiThreadStreams2.cpp
@@ -21,7 +21,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/p2p/hipP2pLinkTypeAndHopFunc.cpp b/tests/src/runtimeApi/p2p/hipP2pLinkTypeAndHopFunc.cpp
index 3cd5d915b5..6786124095 100644
--- a/tests/src/runtimeApi/p2p/hipP2pLinkTypeAndHopFunc.cpp
+++ b/tests/src/runtimeApi/p2p/hipP2pLinkTypeAndHopFunc.cpp
@@ -21,7 +21,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD_CMD: %t %hc %S/%s -o %T/%t %S/../../test_common.cpp -I %S/../../ -L%rocm-path/rocm_smi/lib -lrocm_smi64 -ldl EXCLUDE_HIP_PLATFORM nvidia
+ * BUILD_CMD: %t %hc %S/%s -o %T/%t %S/../../test_common.cpp -I %S/../../ -L%rocm-path/rocm_smi/lib -lrocm_smi64 -lpthread -ldl EXCLUDE_HIP_PLATFORM nvidia
  * TEST: %t --tests 0x1
  * TEST: %t --tests 0x2
  * TEST: %t --tests 0x3
diff --git a/tests/src/runtimeApi/stream/hipStreamACb_MStrm_Mgpu.cpp b/tests/src/runtimeApi/stream/hipStreamACb_MStrm_Mgpu.cpp
index 35f0e15b11..d56ce2b3c0 100644
--- a/tests/src/runtimeApi/stream/hipStreamACb_MStrm_Mgpu.cpp
+++ b/tests/src/runtimeApi/stream/hipStreamACb_MStrm_Mgpu.cpp
@@ -21,7 +21,7 @@
 // kernel. Verify that all the kernels queued are executed before the callback.
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/stream/hipStreamACb_MultiThread.cpp b/tests/src/runtimeApi/stream/hipStreamACb_MultiThread.cpp
index d26e41019c..704621761e 100644
--- a/tests/src/runtimeApi/stream/hipStreamACb_MultiThread.cpp
+++ b/tests/src/runtimeApi/stream/hipStreamACb_MultiThread.cpp
@@ -22,7 +22,7 @@ THE SOFTWARE.
 // This test case is disabled currently.
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/stream/hipStreamAddCallbackCatch.cpp b/tests/src/runtimeApi/stream/hipStreamAddCallbackCatch.cpp
index c694bf6d2a..59d05204ee 100644
--- a/tests/src/runtimeApi/stream/hipStreamAddCallbackCatch.cpp
+++ b/tests/src/runtimeApi/stream/hipStreamAddCallbackCatch.cpp
@@ -11,7 +11,7 @@
 #include "test_common.h"
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/streamOperations/hipstream_operations.cpp b/tests/src/runtimeApi/streamOperations/hipstream_operations.cpp
index cd4d9fd425..8cd3b7bec2 100644
--- a/tests/src/runtimeApi/streamOperations/hipstream_operations.cpp
+++ b/tests/src/runtimeApi/streamOperations/hipstream_operations.cpp
@@ -39,7 +39,7 @@ THE SOFTWARE.
 */
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11 EXCLUDE_HIP_PLATFORM nvidia
+ * BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS --std=c++11 EXCLUDE_HIP_PLATFORM nvidia LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */
diff --git a/tests/src/runtimeApi/synchronization/cache_coherency_cpu_gpu.cpp b/tests/src/runtimeApi/synchronization/cache_coherency_cpu_gpu.cpp
index 3e4aec033c..dae95f1264 100644
--- a/tests/src/runtimeApi/synchronization/cache_coherency_cpu_gpu.cpp
+++ b/tests/src/runtimeApi/synchronization/cache_coherency_cpu_gpu.cpp
@@ -19,7 +19,7 @@ THE SOFTWARE.
 // Simple test for Fine Grained CPU-GPU coherency.
 
 /* HIT_START
- * BUILD: %t %s ../../test_common.cpp HIPCC_OPTIONS -std=c++11 -lpthread EXCLUDE_HIP_PLATFORM nvidia
+ * BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvidia LINK_OPTIONS -lpthread
  * TEST: %t
  * HIT_END
  */

From 436ba9757b965ef967e74f2a9688be5ee0ea863b Mon Sep 17 00:00:00 2001
From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com>
Date: Thu, 26 Jan 2023 21:50:58 +0530
Subject: [PATCH 3/3] SWDEV-377267 - Update hip_build document for ROCm 5.5
 (#3146)

Change-Id: I9fb1527572487c8fa97c7a7c289ec6081f17305d
---
 docs/markdown/hip_build.md | 52 ++++++++++++++++++++------------------
 1 file changed, 28 insertions(+), 24 deletions(-)

diff --git a/docs/markdown/hip_build.md b/docs/markdown/hip_build.md
index 54d85fa540..4bab7ee8e1 100755
--- a/docs/markdown/hip_build.md
+++ b/docs/markdown/hip_build.md
@@ -38,15 +38,15 @@ Install Nvidia driver and pre-build packages (see HIP Installation Guide at http
 
 ## Branch of repository
 
-Before get HIP source code, set the expected branch of repository at the variable HIP_BRANCH.
+Before get HIP source code, set the expected branch of repository at the variable ROCM_BRANCH.
 For example, for ROCm5.0 release branch, set
 ```
-export HIP_BRANCH=rocm-5.0.x
+export ROCM_BRANCH=rocm-5.0.x
 ```
 
-ROCm5.1 release branch, set
+ROCm5.4 release branch, set
 ```
-export HIP_BRANCH=rocm-5.1.x
+export ROCM_BRANCH=rocm-5.4.x
 ```
 Similiar format for future branches.
 
@@ -59,9 +59,10 @@ ROCM_PATH is path where ROCM is installed. BY default ROCM_PATH is at /opt/rocm.
 ## Get HIP source code
 
 ```
-git clone -b $HIP_BRANCH https://github.com/ROCm-Developer-Tools/hipamd.git
-git clone -b $HIP_BRANCH https://github.com/ROCm-Developer-Tools/hip.git
-git clone -b $HIP_BRANCH https://github.com/ROCm-Developer-Tools/ROCclr.git
+git clone -b "$ROCM_BRANCH" https://github.com/ROCm-Developer-Tools/hipamd.git
+git clone -b "$ROCM_BRANCH" https://github.com/ROCm-Developer-Tools/hip.git
+git clone -b "$ROCM_BRANCH" https://github.com/ROCm-Developer-Tools/ROCclr.git
+git clone -b "$ROCM_BRANCH" https://github.com/RadeonOpenCompute/ROCm-OpenCL-Runtime.git
 ```
 
 ## Set the environment variables
@@ -136,51 +137,54 @@ Please note, the integrated HIP directed tests, will be deprecated in future rel
 
 ### Build HIP catch tests
 
-After build and install HIP commands, catch tests can be built via the following instructions,
+HIP catch tests, with new architectured Catch2, are official seperated from HIP project, exist in HIP tests repository, can be built via the following instructions.
+
+#### Get HIP tests source code
 
 ```
-cd "$HIP_DIR"
+git clone -b "$ROCM_BRANCH" https://github.com/ROCm-Developer-Tools/hip-tests.git
+```
+#### Build HIP tests from source
+
+```
+export HIP_TESTS_DIR="$(readlink -f hip-tests)"
+cd "$HIP_TESTS_DIR"
 mkdir -p build; cd build
-export HIP_PATH=$HIPAMD_DIR/build/install
-cmake  ../tests/catch/ -DHIP_PLATFORM=amd
+export HIP_PATH=$HIPAMD_DIR/build/install (or any path where HIP is installed, for example, /opt/rocm)
+cmake ../catch/ -DHIP_PLATFORM=amd
 make -j$(nproc) build_tests
 ctest # run tests
 ```
+HIP catch tests are built under the folder $HIP_TESTS_DIR/build.
 
-HIP catch tests are built under the folder $HIP_DIR/build.
-
-To run a single catch test, the following is an example,
+To run any single catch test, the following is an example,
 
 ```
-cd $HIP_DIR/build/unit/texture
+cd $HIP_TESTS_DIR/build/catch_tests/unit/texture
 ./TextureTest
 ```
 
-### Build HIP Catch2 standalone test
+#### Build HIP Catch2 standalone test
 
 HIP Catch2 supports build a standalone test, for example,
 
 ```
-export PATH=$HIP_DIR/bin:$PATH
-export HIP_PATH=$HIPAMD_DIR/build/install
-
-hipcc $HIP_DIR/tests/catch/unit/memory/hipPointerGetAttributes.cc -I ./tests/catch/include ./tests/catch/hipTestMain/standalone_main.cc -I ./tests/catch/external/Catch2 -g -o hipPointerGetAttributes
+cd "$HIP_TESTS_DIR"
+hipcc $HIP_TESTS_DIR/catch/unit/memory/hipPointerGetAttributes.cc -I ./catch/include ./catch/hipTestMain/standalone_main.cc -I ./catch/external/Catch2 -o hipPointerGetAttributes
 ./hipPointerGetAttributes
 ...
 
 All tests passed
 ```
 
-HIP catch tests, especially new architectured Catch2, will be official HIP tests in the repository and can be built alone as with the instructions shown above.
-
 # Build HIP on NVIDIA platform
 
 
 ## Get HIP source code
 
 ```
-git clone -b $HIP_BRANCH https://github.com/ROCm-Developer-Tools/hip.git
-git clone -b $HIP_BRANCH https://github.com/ROCm-Developer-Tools/hipamd.git
+git clone -b "$ROCM_BRANCH" https://github.com/ROCm-Developer-Tools/hip.git
+git clone -b "$ROCM_BRANCH" https://github.com/ROCm-Developer-Tools/hipamd.git
 ```
 
 ## Set the environment variables