merge release code into development

catch/external/Catch2/cmake/Catch2/Catch.cmake

@@ -217,6 +217,8 @@ function(hip_add_exe_to_target)
   if(UNIX)
     set(_LINKER_LIBS ${_LINKER_LIBS} stdc++fs)
     set(_LINKER_LIBS ${_LINKER_LIBS} -ldl)
+    set(_LINKER_LIBS ${_LINKER_LIBS} pthread)
+    set(_LINKER_LIBS ${_LINKER_LIBS} rt)
   else()
     # res files are built resource files using rc files.
     # use llvm-rc exe to build the res files

catch/unit/device/hipDeviceSetGetCacheConfig.cc

@@ -38,16 +38,12 @@ TEST_CASE("Unit_hipDeviceSetCacheConfig_Positive_Basic") {
 
   const auto cache_config =
       GENERATE(from_range(std::begin(kCacheConfigs), std::end(kCacheConfigs)));
-#if HT_AMD
-  HIP_CHECK_ERROR(hipDeviceSetCacheConfig(cache_config), hipErrorNotSupported);
-#elif HT_NVIDIA
   HIP_CHECK(hipDeviceSetCacheConfig(cache_config));
-#endif
 }
 
 TEST_CASE("Unit_hipDeviceSetCacheConfig_Negative_Parameters") {
 #if HT_AMD
-  HIP_CHECK_ERROR(hipDeviceSetCacheConfig(static_cast<hipFuncCache_t>(-1)), hipErrorNotSupported);
+  HIP_CHECK_ERROR(hipDeviceSetCacheConfig(static_cast<hipFuncCache_t>(-1)), hipSuccess);
 #elif HT_NVIDIA
   HIP_CHECK_ERROR(hipDeviceSetCacheConfig(static_cast<hipFuncCache_t>(-1)), hipErrorInvalidValue);
 #endif
@@ -105,4 +101,4 @@ TEST_CASE("Unit_hipDeviceGetCacheConfig_Positive_Threaded") {
 
 TEST_CASE("Unit_HipDeviceGetCacheConfig_Negative_Parameters") {
   HIP_CHECK_ERROR(hipDeviceGetCacheConfig(nullptr), hipErrorInvalidValue);
-}
+}

catch/unit/device/hipDeviceSetGetSharedMemConfig.cc

@@ -35,17 +35,13 @@ TEST_CASE("Unit_hipDeviceSetSharedMemConfig_Positive_Basic") {
   HIP_CHECK(hipSetDevice(device));
   INFO("Current device is " << device);
 
-#if HT_AMD
-  HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(mem_config), hipErrorNotSupported);
-#elif HT_NVIDIA
   HIP_CHECK(hipDeviceSetSharedMemConfig(mem_config));
-#endif
 }
 
 TEST_CASE("Unit_hipDeviceSetSharedMemConfig_Negative_Parameters") {
 #if HT_AMD
   HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(static_cast<hipSharedMemConfig>(-1)),
-                  hipErrorNotSupported);
+                  hipSuccess);
 #elif HT_NVIDIA
   HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(static_cast<hipSharedMemConfig>(-1)),
                   hipErrorInvalidValue);
@@ -118,4 +114,4 @@ TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Positive_Threaded") {
 
 TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Negative_Parameters") {
   HIP_CHECK_ERROR(hipDeviceGetSharedMemConfig(nullptr), hipErrorInvalidValue);
-}
+}

catch/unit/event/hipEventSynchronize.cc

@@ -110,8 +110,6 @@ TEST_CASE("Unit_hipEventSynchronize_NoEventRecord_Positive") {
 
   // Record the end_event
   HIP_CHECK(hipEventRecord(end_event, NULL));
-  // End event has not been completed
-  HIP_CHECK_ERROR(hipEventQuery(end_event), hipErrorNotReady);
 
   // When hipEventSynchronized is called on event that has not been recorded,
   // the function returns immediately

catch/unit/memory/hipMemAdvise.cc

@@ -224,8 +224,6 @@ TEST_CASE("Unit_hipMemAdvise_TstFlags") {
 }
 
 TEST_CASE("Unit_hipMemAdvise_NegtveTsts") {
-  HipTest::HIP_SKIP_TEST("Fixed few issues to match with Nvidia, Skip now to avoid CI failures");
-  return;
   int MangdMem = HmmAttrPrint();
   if (MangdMem == 1) {
     bool IfTestPassed = true;
@@ -234,15 +232,6 @@ TEST_CASE("Unit_hipMemAdvise_NegtveTsts") {
     std::string str;
     HIP_CHECK(hipGetDeviceCount(&NumDevs));
     HIP_CHECK(hipMallocManaged(&Hmm, MEM_SIZE * 2, hipMemAttachGlobal));
-#if HT_AMD
-    // Passing invalid value(99) device param
-    IfTestPassed &= CheckError(hipMemAdvise(Hmm, MEM_SIZE * 2,
-                               hipMemAdviseSetReadMostly, 99), __LINE__);
-
-    // Passing invalid value(-12) device param
-    IfTestPassed &= CheckError(hipMemAdvise(Hmm, MEM_SIZE * 2,
-                               hipMemAdviseSetReadMostly, -12), __LINE__);
-#endif
     // Passing NULL as first parameter instead of valid pointer to a memory
     IfTestPassed &= CheckError(hipMemAdvise(NULL, MEM_SIZE * 2,
                                hipMemAdviseSetReadMostly, 0), __LINE__);
@@ -380,6 +369,12 @@ TEST_CASE("Unit_hipMemAdvise_ReadMostly") {
       WARN("out value: " << out);
       IfTestPassed = false;
     }
+    // hipMemAdvise should succeed for SetReadMostly and UnsetReadMostly
+    // irrespective of the device
+    HIP_CHECK(hipMemAdvise(Hmm, MEM_SIZE, hipMemAdviseSetReadMostly, 99));
+
+    HIP_CHECK(hipMemAdvise(Hmm, MEM_SIZE, hipMemAdviseUnsetReadMostly, -12));
+
     HIP_CHECK(hipFree(Hmm));
     REQUIRE(IfTestPassed);
   } else {

catch/unit/texture/hipBindTex2DPitch.cc

@@ -20,6 +20,8 @@ THE SOFTWARE.
 #include <hip_test_common.hh>
 #include <hip_test_checkers.hh>
 
+#if CUDA_VERSION < CUDA_12000
+
 #define SIZE_H 8
 #define SIZE_W 12
 #define TYPE_t float
@@ -77,3 +79,5 @@ TEST_CASE("Unit_hipBindTexture2D_Pitch") {
   HIP_CHECK(hipFree(devPtrA));
   HIP_CHECK(hipFree(devPtrB));
 }
+
+#endif // CUDA_VERSION < CUDA_12000

catch/unit/texture/hipBindTexRef1DFetch.cc

@@ -19,6 +19,8 @@ THE SOFTWARE.
 
 #include <hip_test_common.hh>
 
+#if CUDA_VERSION < CUDA_12000
+
 #define N 512
 
 texture<float, 1, hipReadModeElementType> tex;
@@ -79,3 +81,5 @@ TEST_CASE("Unit_hipBindTexture_tex1DfetchVerification") {
   HIP_CHECK(hipFree(texBuf));
   HIP_CHECK(hipFree(devBuf));
 }
+
+#endif // CUDA_VERSION < CUDA_12000

catch/unit/texture/hipNormalizedFloatValueTex.cc

@@ -19,6 +19,8 @@ THE SOFTWARE.
 
 #include <hip_test_common.hh>
 
+#if CUDA_VERSION < CUDA_12000
+
 #define SIZE          10
 #define EPSILON       0.00001
 #define THRESH_HOLD   0.01  // For filter mode
@@ -159,3 +161,5 @@ TEST_CASE("Unit_hipNormalizedFloatValueTex_CheckModes") {
     runTest_hipTextureFilterMode<hipFilterModeLinear>();
   }
 }
+
+#endif  // CUDA_VERSION < CUDA_12000

catch/unit/texture/hipSimpleTexture2DLayered.cc

@@ -20,6 +20,8 @@ THE SOFTWARE.
 #include <hip_test_common.hh>
 #include <hip_test_checkers.hh>
 
+#if CUDA_VERSION < CUDA_12000
+
 typedef float T;
 
 // Texture reference for 2D Layered texture
@@ -107,3 +109,5 @@ TEST_CASE("Unit_hipSimpleTexture2DLayered_Check") {
   free(hData);
   free(hOutputData);
 }
+
+#endif // CUDA_VERSION < CUDA_12000

catch/unit/texture/hipSimpleTexture3D.cc

@@ -20,6 +20,8 @@ THE SOFTWARE.
 #include <hip_test_common.hh>
 #include <hip_test_checkers.hh>
 
+#if CUDA_VERSION < CUDA_12000
+
 // Texture reference for 3D texture
 texture<float, hipTextureType3D, hipReadModeElementType> texf;
 texture<int, hipTextureType3D, hipReadModeElementType>   texi;
@@ -119,3 +121,5 @@ TEST_CASE("Unit_hipSimpleTexture3D_Check_DataTypes") {
     runSimpleTexture3D_Check<char>(i, i+1, i, &texc);
   }
 }
+
+#endif // CUDA_VERSION < CUDA_12000

catch/unit/texture/hipTextureRef2D.cc

@@ -19,6 +19,8 @@ THE SOFTWARE.
 
 #include <hip_test_common.hh>
 
+#if CUDA_VERSION < CUDA_12000
+
 texture<float, 2, hipReadModeElementType> tex;
 
 __global__ void tex2DKernel(float* outputData, int width) {
@@ -90,3 +92,5 @@ TEST_CASE("Unit_hipTextureRef2D_Check") {
   HIP_CHECK(hipFree(dData));
   HIP_CHECK(hipFreeArray(hipArray));
 }
+
+#endif // CUDA_VERSION < CUDA_12000