SWDEV-514841 - Remove invalid stream tests which leads to segfault

Change-Id: I230c4ccc0d2f591cbf2cd0e019f910094f4f3f26

catch/hipTestMain/config/config_nvidia_linux.json

@@ -225,8 +225,21 @@
         "=== Below tests fail in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/210 ===",
         "Unit_hipMemImportFromShareableHandle_Positive_MultiProc",
         "Unit_hipMemMapArrayAsync_Positive_Basic",
-        "=== Disabling failing nvidia tests ====",
+        "=== Disabling failing nvidia tests ===",
         "Unit_hipDeviceSetLimit_Negative_PrintfFifoSize",
-        "Unit_hipDeviceSetLimit_Negative_MallocHeapSize"
+        "Unit_hipDeviceSetLimit_Negative_MallocHeapSize",
+        "=== Disabling tests which no longer behave the same on nvidia platform ===",
+        "Unit_hipGraphInstantiateWithParams_Negative",
+        "Unit_hipGraphAddChildGraphNode_OrgGraphAsChildGraph",
+        "Unit_hipStreamBeginCapture_Negative_Parameters",
+        "Unit_hipStreamDestroy_Negative_DoubleDestroy",
+        "Unit_hipDeviceSynchronize_Positive_Nullstream",
+        "Unit_hipDeviceSynchronize_Functional",
+        "Unit_hipGetSetDeviceFlags_InvalidFlag",
+        "Unit_hipDeviceReset_Positive_Basic",
+        "Unit_hipDeviceReset_Positive_Threaded",
+        "Unit_hipModuleLaunchKernel_Negative_Parameters",
+        "Unit_hipModuleGetTexRef_Positive_Basic",
+        "Unit_hipLaunchCooperativeKernel_Negative_Parameters"
     ]
 }

catch/hipTestMain/config/config_nvidia_windows.json

@@ -58,6 +58,19 @@
         "Unit_hipMallocArray_Negative_BadNumberOfBits",
         "Unit_hipMallocArray_Negative_3ChannelElement",
         "Unit_hipMallocArray_Negative_ChannelAfterZeroChannel",
-        "Unit_hipMallocArray_Negative_InvalidChannelFormat"
+        "Unit_hipMallocArray_Negative_InvalidChannelFormat",
+        "=== Disabling tests which no longer behave the same on nvidia platform ===",
+        "Unit_hipGraphInstantiateWithParams_Negative",
+        "Unit_hipGraphAddChildGraphNode_OrgGraphAsChildGraph",
+        "Unit_hipStreamBeginCapture_Negative_Parameters",
+        "Unit_hipStreamDestroy_Negative_DoubleDestroy",
+        "Unit_hipDeviceSynchronize_Positive_Nullstream",
+        "Unit_hipDeviceSynchronize_Functional",
+        "Unit_hipGetSetDeviceFlags_InvalidFlag",
+        "Unit_hipDeviceReset_Positive_Basic",
+        "Unit_hipDeviceReset_Positive_Threaded",
+        "Unit_hipModuleLaunchKernel_Negative_Parameters",
+        "Unit_hipModuleGetTexRef_Positive_Basic",
+        "Unit_hipLaunchCooperativeKernel_Negative_Parameters"
     ]
 }

catch/multiproc/hipSetGetDeviceMproc.cc

@@ -189,7 +189,7 @@ static void testValidDevices(int numDevices, bool useRocrEnv, int *deviceList,
   if (cPid == 0) {
 #ifdef __HIP_PLATFORM_NVIDIA__
     unsetenv("CUDA_VISIBLE_DEVICES");
-    setenv("CUDA_VISIBLE_DEVICES", visibleDeviceString, 1);
+    setenv("CUDA_VISIBLE_DEVICES", visibleDeviceString.c_str(), 1);
 #else
     unsetenv("ROCR_VISIBLE_DEVICES");
     unsetenv("HIP_VISIBLE_DEVICES");

catch/unit/graph/hipGraphExecUpdate.cc

@@ -413,11 +413,7 @@ TEST_CASE("Unit_hipGraphExecUpdate_Negative_Functional_CountDiffer_1") {
   ret = hipGraphExecUpdate(graphExec1, graph2, &hErrorNode_out, &updateResult_out);
 
   REQUIRE(hipErrorGraphExecUpdateFailure == ret);
-#if HT_NVIDIA
-  REQUIRE(hipGraphExecUpdateErrorNotSupported == updateResult_out);
-#else
   REQUIRE(hipGraphExecUpdateErrorTopologyChanged == updateResult_out);
-#endif
   REQUIRE(NULL == hErrorNode_out);
 
   HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false);
@@ -480,13 +476,7 @@ TEST_CASE("Unit_hipGraphExecUpdate_Negative_Functional_CountDiffer_2") {
                                       hipMemcpyHostToDevice));
     HIP_CHECK(hipGraphInstantiate(&graphExec2, graph2, nullptr, nullptr, 0));
     ret = hipGraphExecUpdate(graphExec2, graph1, &hErrorNode_out, &updateResult_out);
-#if HT_NVIDIA
-    REQUIRE(hipErrorGraphExecUpdateFailure == ret);
-    REQUIRE(hipGraphExecUpdateErrorNotSupported == updateResult_out);
-    REQUIRE(NULL == hErrorNode_out);
-#else
     REQUIRE(hipSuccess == ret);
-#endif
     HIP_CHECK(hipGraphExecDestroy(graphExec2));
     HIP_CHECK(hipGraphDestroy(graph2));
   }
@@ -499,11 +489,7 @@ TEST_CASE("Unit_hipGraphExecUpdate_Negative_Functional_CountDiffer_2") {
                                       hipMemcpyHostToDevice));
     ret = hipGraphExecUpdate(graphExec3, graph3, &hErrorNode_out, &updateResult_out);
     REQUIRE(hipErrorGraphExecUpdateFailure == ret);
-#if HT_NVIDIA
-    REQUIRE(hipGraphExecUpdateErrorNotSupported == updateResult_out);
-#else
     REQUIRE(hipGraphExecUpdateErrorTopologyChanged == updateResult_out);
-#endif
     REQUIRE(NULL == hErrorNode_out);
 
     HIP_CHECK(hipGraphExecDestroy(graphExec3));
@@ -626,11 +612,7 @@ TEST_CASE("Unit_hipGraphExecUpdate_Negative_NodeType_Changed") {
   HIP_CHECK(hipGraphAddDependencies(graph2, &memcpy_A, &memsetNode, 1));
   ret = hipGraphExecUpdate(graphExec, graph2, &hErrorNode_out, &updateResult_out);
   REQUIRE(hipErrorGraphExecUpdateFailure == ret);
-#if HT_NVIDIA
-  REQUIRE(hipGraphExecUpdateErrorTopologyChanged == updateResult_out);
-#else
   REQUIRE(hipGraphExecUpdateErrorNodeTypeChanged == updateResult_out);
-#endif
   REQUIRE(memsetNode == hErrorNode_out);
 
   HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false);

catch/unit/graph/hipGraphLaunch_old.cc

@@ -95,8 +95,9 @@ TEST_CASE("Unit_hipGraphLaunch_Negative") {
     HIP_CHECK(hipGraphDestroy(graph));
     HIP_CHECK(hipStreamDestroy(stream));
   }
-/* In this case in CUDA setup this api call is giving - unknown error (999)
-   So enabling this test for both AMD and CUDA by checking with hipSuccess */
+/* In this case in CUDA setup this api call leads to undefined behavior
+   So enabling this test only for AMD and checking with hipSuccess */
+#if HT_AMD
   SECTION("Destroy stream and try to launch respective executable graph") {
     constexpr size_t Nbytes = 1024;
     hipGraph_t graph;
@@ -133,6 +134,7 @@ TEST_CASE("Unit_hipGraphLaunch_Negative") {
     HIP_CHECK(hipGraphExecDestroy(graphExec));
     HIP_CHECK(hipGraphDestroy(graph));
   }
+#endif
   SECTION("Destroy graph and try to launch respective executable graph") {
     constexpr size_t Nbytes = 1024;
     hipGraph_t graph;

catch/unit/graph/hipGraphNodeFindInClone.cc

@@ -94,12 +94,6 @@ TEST_CASE("Unit_hipGraphNodeFindInClone_Negative") {
     REQUIRE(hipGraphNodeFindInClone(&clonedgraphnode, graphnode, graph) ==
             hipErrorInvalidValue);
   }
-
-  SECTION("Destroy the graph node and find in cloned graph") {
-    HIP_CHECK(hipGraphDestroyNode(graphnode));
-    REQUIRE(hipGraphNodeFindInClone(&clonedgraphnode, graphnode, clonedgraph) ==
-            hipErrorInvalidValue);
-  }
 #endif
 
   SECTION("Pass invalid original graphnode") {

catch/unit/graph/hipLaunchHostFunc.cc

@@ -67,7 +67,7 @@ TEST_CASE("Unit_hipLaunchHostFunc_Negative_Parameters") {
   SECTION("Pass functions as nullptr") {
     HIP_CHECK_ERROR(hipLaunchHostFunc(stream, nullptr, nullptr), hipErrorInvalidValue);
   }
-
+#if HT_AMD
   SECTION("Pass uninitialized stream") {
     hipHostFn_t fn = hostNodeCallbackDummy;
     constexpr auto InvalidStream = [] {
@@ -76,6 +76,7 @@ TEST_CASE("Unit_hipLaunchHostFunc_Negative_Parameters") {
     };
     HIP_CHECK_ERROR(hipLaunchHostFunc(InvalidStream(), fn, nullptr), hipErrorContextIsDestroyed);
   }
+#endif
 }
 
 /**

catch/unit/graph/hipStreamEndCapture.cc

@@ -56,12 +56,13 @@ TEST_CASE("Unit_hipStreamEndCapture_Negative_Parameters") {
   }
 #if HT_NVIDIA
   SECTION("Pass graph as nullptr") {
-    HIP_CHECK_ERROR(hipStreamEndCapture(stream, nullptr), hipErrorInvalidValue);
+    HIP_CHECK_ERROR(hipStreamEndCapture(stream, nullptr), hipErrorIllegalState);
   }
 #endif
   SECTION("End capture on stream where capture has not yet started") {
     HIP_CHECK_ERROR(hipStreamEndCapture(stream, &graph), hipErrorIllegalState);
   }
+#if HT_AMD
   SECTION("Destroy stream and try to end capture") {
     hipStream_t destroyed_stream;
     HIP_CHECK(hipStreamCreate(&destroyed_stream));
@@ -69,6 +70,7 @@ TEST_CASE("Unit_hipStreamEndCapture_Negative_Parameters") {
     HIP_CHECK(hipStreamDestroy(destroyed_stream));
     HIP_CHECK_ERROR(hipStreamEndCapture(destroyed_stream, &graph), hipErrorContextIsDestroyed);
   }
+#endif
 }
 
 /**

catch/unit/graph/hipStreamEndCapture_old.cc

@@ -64,12 +64,12 @@ TEST_CASE("Unit_hipStreamEndCapture_Negative") {
     hipStream_t stream;
     HIP_CHECK(hipStreamCreate(&stream));
     ret = hipStreamEndCapture(stream, nullptr);
-    REQUIRE(hipErrorInvalidValue == ret);
+    REQUIRE(hipErrorIllegalState == ret);
     HIP_CHECK(hipStreamDestroy(stream));
   }
   SECTION("Pass graph as nullptr and stream as hipStreamPerThread") {
     ret = hipStreamEndCapture(hipStreamPerThread, nullptr);
-    REQUIRE(hipErrorInvalidValue == ret);
+    REQUIRE(hipErrorIllegalState == ret);
   }
 #endif
   SECTION("End capture on stream where capture has not yet started") {
@@ -80,6 +80,7 @@ TEST_CASE("Unit_hipStreamEndCapture_Negative") {
     REQUIRE(hipErrorIllegalState == ret);
     HIP_CHECK(hipStreamDestroy(stream));
   }
+#if HT_AMD
   SECTION("Destroy stream and try to end capture") {
     hipStream_t stream;
     hipGraph_t graph;
@@ -89,6 +90,7 @@ TEST_CASE("Unit_hipStreamEndCapture_Negative") {
     ret = hipStreamEndCapture(stream, &graph);
     REQUIRE(hipErrorContextIsDestroyed == ret);
   }
+#endif
   SECTION("Destroy graph and try to end capture in between") {
     hipStream_t stream{nullptr};
     hipGraph_t graph{nullptr};

catch/unit/graph/hipStreamGetCaptureInfo.cc

@@ -185,15 +185,6 @@ TEST_CASE("Unit_hipStreamGetCaptureInfo_Negative_Parameters") {
     HIP_CHECK(hipStreamEndCapture(stream, &graph));
     HIP_CHECK(hipGraphDestroy(graph));
   }
-  SECTION("Capture status when stream is uninitialized") {
-    constexpr auto InvalidStream = [] {
-      StreamGuard sg(Streams::created);
-      return sg.stream();
-    };
-
-    HIP_CHECK_ERROR(hipStreamGetCaptureInfo(InvalidStream(), &cStatus, &capSequenceID),
-                    hipErrorContextIsDestroyed);
-  }
 #endif
 }
 

catch/unit/graph/hipStreamGetCaptureInfo_v2.cc

@@ -246,16 +246,6 @@ TEST_CASE("Unit_hipStreamGetCaptureInfo_v2_Negative_Parameters") {
     HIP_CHECK(hipStreamEndCapture(stream, &graph));
     HIP_CHECK(hipGraphDestroy(graph));
   }
-  SECTION("Capture status when stream is uninitialized") {
-    constexpr auto InvalidStream = [] {
-      StreamGuard sg(Streams::created);
-      return sg.stream();
-    };
-
-    HIP_CHECK_ERROR(hipStreamGetCaptureInfo_v2(InvalidStream(), &captureStatus, &capSequenceID,
-                                               &capInfoGraph, &nodelist, &numDependencies),
-                    hipErrorContextIsDestroyed);
-  }
 #endif
 }
 

catch/unit/graph/hipStreamIsCapturing.cc

@@ -62,18 +62,6 @@ TEST_CASE("Unit_hipStreamIsCapturing_Negative_Parameters") {
     HIP_CHECK(hipStreamEndCapture(stream, &graph));
     HIP_CHECK(hipGraphDestroy(graph));
   }
-#if HT_NVIDIA  // EXSWHTEC-216
-  SECTION("Check capture status when stream is uninitialized") {
-    hipStreamCaptureStatus cStatus;
-
-    constexpr auto InvalidStream = [] {
-      StreamGuard sg(Streams::created);
-      return sg.stream();
-    };
-
-    HIP_CHECK_ERROR(hipStreamIsCapturing(InvalidStream(), &cStatus), hipErrorContextIsDestroyed);
-  }
-#endif
 }
 
 /**

catch/unit/memory/hipDrvMemcpy3DAsync.cc

@@ -210,16 +210,6 @@ TEST_CASE("Unit_hipDrvMemcpy3DAsync_Negative_Parameters") {
           hipErrorInvalidValue);
     }
 #endif
-
-#if HT_NVIDIA // Disabled on AMD due to defect - EXSWHTEC-235
-    SECTION("Invalid stream") {
-      StreamGuard stream_guard(Streams::created);
-      HIP_CHECK(hipStreamDestroy(stream_guard.stream()));
-      HIP_CHECK_ERROR(DrvMemcpy3DWrapper<async>(dst_ptr, dst_pos, src_ptr, src_pos, extent, kind,
-                                                stream_guard.stream()),
-                      hipErrorContextIsDestroyed);
-    }
-#endif
   };
 
   SECTION("Host to Device") {
@@ -261,4 +251,4 @@ TEST_CASE("Unit_hipDrvMemcpy3DAsync_Negative_Parameters") {
     NegativeTests(dst_alloc.pitched_ptr(), make_hipPos(0, 0, 0), src_alloc.pitched_ptr(),
                   make_hipPos(0, 0, 0), extent, hipMemcpyDeviceToDevice);
   }
-}
+}

catch/unit/memory/hipMemPrefetchAsync.cc

@@ -155,7 +155,7 @@ TEST_CASE("Unit_hipMemPrefetchAsync_Negative_Parameters") {
     HIP_CHECK_ERROR(hipMemPrefetchAsync(alloc.ptr(), kPageSize, hipInvalidDeviceId),
                     hipErrorInvalidDevice);
   }
-
+#if HT_AMD
   SECTION("Invalid stream") {
     hipStream_t stream;
     HIP_CHECK(hipStreamCreate(&stream));
@@ -163,4 +163,5 @@ TEST_CASE("Unit_hipMemPrefetchAsync_Negative_Parameters") {
     HIP_CHECK_ERROR(hipMemPrefetchAsync(alloc.ptr(), kPageSize, device, stream),
                     hipErrorContextIsDestroyed);
   }
+#endif
 }

catch/unit/memory/hipMemcpy2DAsync.cc

@@ -149,15 +149,6 @@ TEST_CASE("Unit_hipMemcpy2DAsync_Negative_Parameters") {
                                        static_cast<hipMemcpyKind>(-1), nullptr),
                       hipErrorInvalidMemcpyDirection);
     }
-#endif
-#if HT_NVIDIA // Disabled on AMD due to defect - EXSWHTEC-235
-    SECTION("Invalid stream") {
-      StreamGuard stream_guard(Streams::created);
-      HIP_CHECK(hipStreamDestroy(stream_guard.stream()));
-      HIP_CHECK_ERROR(
-          hipMemcpy2DAsync(dst, dpitch, src, spitch, width, height, kind, stream_guard.stream()),
-          hipErrorContextIsDestroyed);
-    }
 #endif
   };
 

catch/unit/memory/hipMemcpy2DFromArrayAsync.cc

@@ -230,12 +230,6 @@ TEST_CASE("Unit_hipMemcpy2DFromArrayAsync_Negative_Parameters") {
                                                 height, static_cast<hipMemcpyKind>(-1), nullptr),
                       hipErrorInvalidMemcpyDirection);
     }
-    SECTION("Invalid stream") {
-      HIP_CHECK_ERROR(hipMemcpy2DFromArrayAsync(host_alloc.ptr(), 2 * width * sizeof(int),
-                                                array_alloc.ptr(), 0, 0, width * sizeof(int),
-                                                height, hipMemcpyDeviceToHost, InvalidStream()),
-                      hipErrorContextIsDestroyed);
-    }
 #endif
   }
   SECTION("Array to device") {
@@ -284,12 +278,6 @@ TEST_CASE("Unit_hipMemcpy2DFromArrayAsync_Negative_Parameters") {
                                                 height, static_cast<hipMemcpyKind>(-1), nullptr),
                       hipErrorInvalidMemcpyDirection);
     }
-    SECTION("Invalid stream") {
-      HIP_CHECK_ERROR(hipMemcpy2DFromArrayAsync(device_alloc.ptr(), device_alloc.pitch(),
-                                                array_alloc.ptr(), 0, 0, width * sizeof(int),
-                                                height, hipMemcpyDeviceToDevice, InvalidStream()),
-                      hipErrorContextIsDestroyed);
-    }
 #endif
   }
 }

catch/unit/memory/hipMemcpy2DToArrayAsync.cc

@@ -225,12 +225,6 @@ TEST_CASE("Unit_hipMemcpy2DToArrayAsync_Negative_Parameters") {
                                               static_cast<hipMemcpyKind>(-1), nullptr),
                       hipErrorInvalidMemcpyDirection);
     }
-    SECTION("Invalid stream") {
-      HIP_CHECK_ERROR(hipMemcpy2DToArrayAsync(array_alloc.ptr(), 0, 0, host_alloc.ptr(),
-                                              2 * width * sizeof(int), width * sizeof(int), height,
-                                              hipMemcpyHostToDevice, InvalidStream()),
-                      hipErrorContextIsDestroyed);
-    }
 #endif
   }
   SECTION("Device to Array") {
@@ -279,12 +273,6 @@ TEST_CASE("Unit_hipMemcpy2DToArrayAsync_Negative_Parameters") {
                                               static_cast<hipMemcpyKind>(-1), nullptr),
                       hipErrorInvalidMemcpyDirection);
     }
-    SECTION("Invalid stream") {
-      HIP_CHECK_ERROR(hipMemcpy2DToArrayAsync(array_alloc.ptr(), 0, 0, device_alloc.ptr(),
-                                              device_alloc.pitch(), width * sizeof(int), height,
-                                              hipMemcpyDeviceToDevice, InvalidStream()),
-                      hipErrorContextIsDestroyed);
-    }
 #endif
   }
 }

catch/unit/memory/hipMemcpy3DAsync.cc

@@ -205,16 +205,6 @@ TEST_CASE("Unit_hipMemcpy3DAsync_Negative_Parameters") {
                       hipErrorInvalidMemcpyDirection);
     }
 #endif
-
-#if HT_NVIDIA // Disabled on AMD due to defect - EXSWHTEC-235
-    SECTION("Invalid stream") {
-      StreamGuard stream_guard(Streams::created);
-      HIP_CHECK(hipStreamDestroy(stream_guard.stream()));
-      HIP_CHECK_ERROR(Memcpy3DWrapper<async>(dst_ptr, dst_pos, src_ptr, src_pos, extent, kind,
-                                             stream_guard.stream()),
-                      hipErrorContextIsDestroyed);
-    }
-#endif
   };
 
   SECTION("Host to Device") {

catch/unit/memory/hipMemcpyAsync.cc

@@ -97,12 +97,13 @@ TEST_CASE("Unit_hipMemcpyAsync_Negative_Parameters") {
                                      static_cast<hipMemcpyKind>(-1), nullptr),
                       hipErrorInvalidMemcpyDirection);
     }
-
+#if HT_AMD
     SECTION("Invalid stream") {
       HIP_CHECK_ERROR(hipMemcpyAsync(device_alloc.ptr(), host_alloc.ptr(), kPageSize,
                                      hipMemcpyHostToDevice, InvalidStream()),
                       hipErrorContextIsDestroyed);
     }
+#endif
   }
 
   SECTION("Device to host") {
@@ -118,11 +119,13 @@ TEST_CASE("Unit_hipMemcpyAsync_Negative_Parameters") {
                       hipErrorInvalidMemcpyDirection);
     }
 
+#if HT_AMD
     SECTION("Invalid stream") {
       HIP_CHECK_ERROR(hipMemcpyAsync(host_alloc.ptr(), device_alloc.ptr(), kPageSize,
                                      hipMemcpyDeviceToHost, InvalidStream()),
                       hipErrorContextIsDestroyed);
     }
+#endif
   }
 
   SECTION("Host to host") {
@@ -137,12 +140,13 @@ TEST_CASE("Unit_hipMemcpyAsync_Negative_Parameters") {
                                      static_cast<hipMemcpyKind>(-1), nullptr),
                       hipErrorInvalidMemcpyDirection);
     }
-
+#if HT_AMD
     SECTION("Invalid stream") {
       HIP_CHECK_ERROR(hipMemcpyAsync(dst_alloc.ptr(), src_alloc.ptr(), kPageSize,
                                      hipMemcpyHostToHost, InvalidStream()),
                       hipErrorContextIsDestroyed);
     }
+#endif
   }
 
   SECTION("Device to device") {
@@ -158,11 +162,12 @@ TEST_CASE("Unit_hipMemcpyAsync_Negative_Parameters") {
                                      static_cast<hipMemcpyKind>(-1), nullptr),
                       hipErrorInvalidMemcpyDirection);
     }
-
+#if HT_AMD
     SECTION("Invalid stream") {
       HIP_CHECK_ERROR(hipMemcpyAsync(dst_alloc.ptr(), src_alloc.ptr(), kPageSize,
                                      hipMemcpyDeviceToDevice, InvalidStream()),
                       hipErrorContextIsDestroyed);
     }
+#endif
   }
 }

catch/unit/memory/hipMemcpyAsync_derivatives.cc

@@ -70,13 +70,14 @@ TEST_CASE("Unit_hipMemcpyDtoHAsync_Negative_Parameters") {
         return hipMemcpyDtoHAsync(dst, reinterpret_cast<hipDeviceptr_t>(src), count, nullptr);
       },
       host_alloc.ptr(), device_alloc.ptr(), kPageSize);
-
+#if HT_AMD
   SECTION("Invalid stream") {
     HIP_CHECK_ERROR(
         hipMemcpyDtoHAsync(host_alloc.ptr(), reinterpret_cast<hipDeviceptr_t>(device_alloc.ptr()),
                            kPageSize, InvalidStream()),
         hipErrorContextIsDestroyed);
   }
+#endif
 }
 
 TEST_CASE("Unit_hipMemcpyHtoDAsync_Positive_Basic") {
@@ -115,12 +116,13 @@ TEST_CASE("Unit_hipMemcpyHtoDAsync_Negative_Parameters") {
         return hipMemcpyHtoDAsync(reinterpret_cast<hipDeviceptr_t>(dst), src, count, nullptr);
       },
       device_alloc.ptr(), host_alloc.ptr(), kPageSize);
-
+#if HT_AMD
   SECTION("Invalid stream") {
     HIP_CHECK_ERROR(hipMemcpyHtoDAsync(reinterpret_cast<hipDeviceptr_t>(device_alloc.ptr()),
                                        host_alloc.ptr(), kPageSize, InvalidStream()),
                     hipErrorContextIsDestroyed);
   }
+#endif
 }
 
 TEST_CASE("Unit_hipMemcpyDtoDAsync_Positive_Basic") {
@@ -165,13 +167,14 @@ TEST_CASE("Unit_hipMemcpyDtoDAsync_Negative_Parameters") {
                                   reinterpret_cast<hipDeviceptr_t>(src), count, nullptr);
       },
       dst_alloc.ptr(), src_alloc.ptr(), kPageSize);
-
+#if HT_AMD
   SECTION("Invalid stream") {
     HIP_CHECK_ERROR(hipMemcpyDtoDAsync(reinterpret_cast<hipDeviceptr_t>(dst_alloc.ptr()),
                                        reinterpret_cast<hipDeviceptr_t>(src_alloc.ptr()), kPageSize,
                                        InvalidStream()),
                     hipErrorContextIsDestroyed);
   }
+#endif
 }
 
 /**

catch/unit/memory/hipMemcpyParam2DAsync.cc

@@ -182,15 +182,6 @@ TEST_CASE("Unit_hipMemcpyParam2DAsync_Negative_Parameters") {
                           dst, dpitch, src, spitch, width, height, kind),
                       hipErrorInvalidValue);
     }
-#endif
-#if HT_NVIDIA // Disabled on AMD due to defect - EXSWHTEC-235
-    SECTION("Invalid stream") {
-      StreamGuard stream_guard(Streams::created);
-      HIP_CHECK(hipStreamDestroy(stream_guard.stream()));
-      HIP_CHECK_ERROR(MemcpyParam2DAdapter<async>()(dst, dpitch, src, spitch, width, height, kind,
-                                                    stream_guard.stream()),
-                      hipErrorContextIsDestroyed);
-    }
 #endif
   };
 
@@ -221,4 +212,4 @@ TEST_CASE("Unit_hipMemcpyParam2DAsync_Negative_Parameters") {
     NegativeTests(dst_alloc.ptr(), dst_alloc.pitch(), src_alloc.ptr(), src_alloc.pitch(),
                   dst_alloc.width(), dst_alloc.height(), hipMemcpyDeviceToDevice);
   }
-}
+}