SWDEV-523137 - Enable failing negative tests

catch/hipTestMain/config/config_amd_linux

@@ -187,9 +187,6 @@
         "Unit_hipModuleGetTexRef_Negative_Name_Is_Empty_String",
         "SWDEV-441785: Below tests failing in stress test on 05/01/24 ===",
         "Unit_hipMemcpyParam2DAsync_Positive_Basic",
-        "SWDEV-442583: Below tests failing in stress test on 12/01/24 ===",
-        "Unit_hipLaunchCooperativeKernelMultiDevice_Negative_Parameters",
-        "Unit_hipLaunchCooperativeKernelMultiDevice_Negative_MultiKernelSameDevice",
         "Unit_hipExtLaunchMultiKernelMultiDevice_Negative_MultiKernelSameDevice",
         "=== Below tests are failing PSDB ===",
         "Unit_hipMemcpy3D_Positive_Synchronization_Behavior",
@@ -225,8 +222,6 @@
         "Unit_atomicXor_Negative_Parameters_RTC",
         "Unit_atomicMin_Negative_Parameters_RTC",
         "Unit_atomicMax_Negative_Parameters_RTC",
-        "Unit_Kernel_Launch_bounds_Negative_OutOfBounds",
-        "Unit_Kernel_Launch_bounds_Negative_Parameters_RTC",
         "Unit_Device_sin_Accuracy_Positive - float",
         "Unit_Device_sin_Accuracy_Positive - double",
         "Unit_Device_cos_Accuracy_Positive - float",
@@ -694,8 +689,6 @@
         "SWDEV-447384, SWDEV-447932: These tests fail in gfx1100, gfx1101 & gfx1102",
         "Unit_hipFreeAsync_Negative_Parameters",
         "SWDEV-445928: These tests fail in PSDB stress test on 09/02/2024",
-        "Unit_hipCreateSurfaceObject_Negative_Parameters",
-        "Unit_hipDestroySurfaceObject_Negative_Parameters",
         "Unit_Device___float2half_rd_Accuracy_Limited_Positive",
         "Unit_Device___float2half_ru_Accuracy_Limited_Positive",
         "Unit_Device___float2half_rz_Accuracy_Limited_Positive",

catch/hipTestMain/config/config_nvidia_linux.json

@@ -13,8 +13,6 @@
         "Unit_ChannelDescriptor_Positive_Basic_3D - long3",
         "Unit_ChannelDescriptor_Positive_Basic_4D - ulong4",
         "Unit_ChannelDescriptor_Positive_Basic_4D - long4",
-        "=== Below test fails in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/38 ===",
-        "Unit_hipFreeAsync_Negative_Parameters",
         "=== Below test fails in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/92 ===",
         "Unit_hipGetTexObjectResourceDesc_positive",
         "Unit_hipGetTexObjectResourceDesc_Negative_Parameters",
@@ -64,8 +62,6 @@
         "Unit_hipFreeMipmappedArray_Negative_Nullptr",
         "Unit_hipFreeMipmappedArrayMultiTArray - int",
         "Unit_hipFreeMipmappedArray_Negative_Parameters",
-        "Unit_hipCreateSurfaceObject_Negative_Parameters",
-        "Unit_hipDestroySurfaceObject_Negative_Parameters",
         "Unit_hipMemcpy3D_Positive_Synchronization_Behavior",
         "Unit_hipMemcpy2D_Positive_Synchronization_Behavior",
         "Unit_hipDrvMemcpy3D_Positive_Synchronization_Behavior",

catch/unit/executionControl/hipLaunchCooperativeKernelMultiDevice.cc

@@ -34,6 +34,10 @@ TEST_CASE("Unit_hipLaunchCooperativeKernelMultiDevice_Positive_Basic") {
   }
 
   const auto device_count = HipTest::getDeviceCount();
+  if (device_count < 2) {
+    SUCCEED("Test requires at least 2 devices");
+    return;
+  }
 
   std::vector<hipLaunchParams> params_list(device_count);
 
@@ -66,6 +70,10 @@ TEST_CASE("Unit_hipLaunchCooperativeKernelMultiDevice_Negative_Parameters") {
   }
 
   const auto device_count = HipTest::getDeviceCount();
+  if (device_count < 2) {
+    SUCCEED("Test requires at least 2 devices");
+    return;
+  }
 
   std::vector<hipLaunchParams> params_list(device_count);
 
@@ -137,6 +145,12 @@ TEST_CASE("Unit_hipLaunchCooperativeKernelMultiDevice_Negative_MultiKernelSameDe
     return;
   }
 
+  const auto device_count = HipTest::getDeviceCount();
+  if (device_count < 2) {
+    SUCCEED("Test requires at least 2 devices");
+    return;
+  }
+
   HIP_CHECK(hipSetDevice(0));
 
   std::vector<hipLaunchParams> params_list(2);
@@ -156,4 +170,4 @@ TEST_CASE("Unit_hipLaunchCooperativeKernelMultiDevice_Negative_MultiKernelSameDe
   for (const auto params : params_list) {
     HIP_CHECK(hipStreamDestroy(params.stream));
   }
-}
+}

catch/unit/launchBounds/launch_bounds.cc

@@ -127,11 +127,8 @@ TEST_CASE("Unit_Kernel_Launch_bounds_Negative_OutOfBounds") {
  * ------------------------
  *  - Validates handling of invalid arguments:
  *    -# Compiles kernels that are not created appropriately:
- *      - Maximum number of threads is 0
- *      - Maximum number of threads is negative
- *      - Minimum number of warps is negative
  *      - Maximum number of threads is not integer value
- *      - Mimimum number of warps is not integer value
+ *      - Minimum number of warps is not integer value
  *    -# Expected output: compiler error
  *  - Uses RTC for compilation.
  * Test source
@@ -144,13 +141,7 @@ TEST_CASE("Unit_Kernel_Launch_bounds_Negative_OutOfBounds") {
 TEST_CASE("Unit_Kernel_Launch_bounds_Negative_Parameters_RTC") {
   hiprtcProgram program{};
 
-#if HT_AMD
-  const auto program_source = GENERATE(kMaxThreadsZero, kMaxThreadsNegative, kMinWarpsNegative,
-                                       kMaxThreadsNotInt, kMinWarpsNotInt);
-#else
-  // Aligned with CUDA behavior and expected behavior on NVIDIA
   const auto program_source = GENERATE(kMaxThreadsNotInt, kMinWarpsNotInt);
-#endif
 
   HIPRTC_CHECK(hiprtcCreateProgram(&program, program_source, "launch_bounds_negative.cc", 0,
                                    nullptr, nullptr));
@@ -176,6 +167,6 @@ TEST_CASE("Unit_Kernel_Launch_bounds_Negative_Parameters_RTC") {
 }
 
 /**
-* End doxygen group DeviceLanguageTest.
-* @}
-*/
+ * End doxygen group DeviceLanguageTest.
+ * @}
+ */

catch/unit/memory/hipFreeAsync.cc

@@ -57,7 +57,11 @@ TEST_CASE("Unit_hipFreeAsync_Negative_Parameters") {
   StreamGuard stream(Streams::created);
 
   SECTION("dev_ptr is nullptr") {
+#if HT_AMD
     HIP_CHECK_ERROR(hipFreeAsync(nullptr, stream.stream()), hipErrorInvalidValue);
+#else
+    HIP_CHECK(hipFreeAsync(nullptr, stream.stream()));
+#endif
   }
 
   SECTION("Invalid stream handle") {
@@ -79,9 +83,9 @@ TEST_CASE("Unit_hipFreeAsync_Negative_Parameters") {
 }
 
 /**
-* End doxygen group StreamOTest.
-* @}
-*/
+ * End doxygen group StreamOTest.
+ * @}
+ */
 
 /**
  * Test Description
@@ -109,9 +113,8 @@ TEST_CASE("Unit_hipFreeAsync_capturehipFreeAsync") {
 
   // Start Capturing
   HIP_CHECK(hipStreamBeginCapture(stream, hipStreamCaptureModeGlobal));
-  HIP_CHECK(hipMallocFromPoolAsync(reinterpret_cast<void**>(&devMem),
-                                   sizeof(int) * rows * cols, memPool,
-                                   stream));
+  HIP_CHECK(hipMallocFromPoolAsync(reinterpret_cast<void**>(&devMem), sizeof(int) * rows * cols,
+                                   memPool, stream));
   HIP_CHECK(hipFreeAsync(devMem, stream));
   // End Capture
   HIP_CHECK(hipStreamEndCapture(stream, &graph));

catch/unit/surface/hipCreateSurfaceObject.cc

@@ -63,30 +63,26 @@ TEST_CASE("Unit_hipCreateSurfaceObject_Negative_Parameters") {
 
   SECTION("invalid resource type") {
     resc.resType = hipResourceTypeLinear;
+#if HT_AMD
     HIP_CHECK_ERROR(hipCreateSurfaceObject(&surf, &resc), hipErrorInvalidValue);
+#else
+    HIP_CHECK_ERROR(hipCreateSurfaceObject(&surf, &resc), hipErrorInvalidChannelDescriptor);
+#endif
   }
 
-#if HT_NVIDIA  // DIsalbed due to defect EXSWHTEC-366
   SECTION("array handle is nullptr") {
     resc.res.array.array = nullptr;
+#if HT_AMD
+    HIP_CHECK_ERROR(hipCreateSurfaceObject(&surf, &resc), hipErrorInvalidValue);
+#else
     HIP_CHECK_ERROR(hipCreateSurfaceObject(&surf, &resc), hipErrorInvalidHandle);
-  }
 #endif
-
-#if HT_NVIDIA  // Disalbed due to defect EXSWHTEC-367
-  SECTION("freed array handle") {
-    hipArray_t invalid_array;
-    HIP_CHECK(hipMallocArray(&invalid_array, &desc, 64, 0, hipArraySurfaceLoadStore));
-    HIP_CHECK(hipFreeArray(invalid_array));
-    resc.res.array.array = invalid_array;
-    HIP_CHECK_ERROR(hipCreateSurfaceObject(&surf, &resc), hipErrorContextIsDestroyed);
   }
-#endif
 
   HIP_CHECK(hipFreeArray(array));
 }
 
 /**
-* End doxygen group SurfaceTest.
-* @}
-*/
+ * End doxygen group SurfaceTest.
+ * @}
+ */

catch/unit/surface/hipDestroySurfaceObject.cc

@@ -61,13 +61,17 @@ TEST_CASE("Unit_hipDestroySurfaceObject_Negative_Parameters") {
     HIP_CHECK(hipCreateSurfaceObject(&surf, &resc));
 
     HIP_CHECK(hipDestroySurfaceObject(surf));
+#if HT_AMD
     HIP_CHECK_ERROR(hipDestroySurfaceObject(surf), hipErrorInvalidValue);
+#else
+    HIP_CHECK(hipDestroySurfaceObject(surf));
+#endif
 
     HIP_CHECK(hipFreeArray(array));
   }
 }
 
 /**
-* End doxygen group SurfaceTest.
-* @}
-*/
+ * End doxygen group SurfaceTest.
+ * @}
+ */