SWDEV-515530 - Re-enable passing tests (#592)
* SWDEV-515530 - Re-enable passing tests * SWDEV-515530 - Revert back windows config file * SWDEV-515530 - Fix new line * SWDEV-515530 - Enable a few more tests * SWDEV-515530 - Enable passing VMM tests * SWDEV-515530 - Disable failing tests * SWDEV-515530 - Fix and enable texture tests * SWDEV-515530 - Minor fixes * SWDEV-515530 - Disable one more test --------- Co-authored-by: Marko Arandjelovic <Marko.Arandjelovic@amd.com>
Этот коммит содержится в:
systems-assistant[bot]
коммит произвёл
GitHub
GitHub
родитель
43ac6b2ef5
Коммит
ae874b489d
@@ -9,86 +9,25 @@
|
||||
],
|
||||
"DisabledTests": [
|
||||
#if defined COMMON
|
||||
"Unit_hipMallocFromPoolAsync_MThread_MaxThresh",
|
||||
"Unit_hipMallocFromPoolAsync_MThread_CommonMpool_DefaultMempool",
|
||||
"Unit_hipMemPoolTrimTo_Multithreaded",
|
||||
"Unit_hipMemPoolSetGetAccess_Positive_MultipleGPU",
|
||||
"Unit_hipStreamPerThread_DeviceReset_1",
|
||||
"Unit_hipDeviceGetSharedMemConfig_Positive_Basic",
|
||||
"Unit_hipDeviceGetSharedMemConfig_Positive_Threaded",
|
||||
"Unit_hipGetDeviceFlags_Positive_Context",
|
||||
"Unit_hipInit_Negative",
|
||||
"Unit_hipDeviceReset_Positive_Basic",
|
||||
"Unit_hipDeviceReset_Positive_Threaded",
|
||||
"Unit_hipFuncSetCacheConfig_Positive_Basic",
|
||||
"Unit_hipFuncSetCacheConfig_Negative_Parameters",
|
||||
"Unit_hipFuncSetSharedMemConfig_Positive_Basic",
|
||||
"Unit_hipFuncSetAttribute_Positive_PreferredSharedMemoryCarveout",
|
||||
"NOTE: The following test is disabled due to defect - EXSWHTEC-241",
|
||||
"NOTE: The following test is disabled due to defect - EXSWHTEC-242",
|
||||
"Unit_hipFuncGetAttributes_Positive_Basic",
|
||||
"NOTE: The following test is disabled due to defect - EXSWHTEC-244",
|
||||
"Unit_hipExtLaunchMultiKernelMultiDevice_Negative_Parameters",
|
||||
"Unit_hipOccupancyMaxActiveBlocksPerMultiprocessor_Negative_Parameters",
|
||||
"Unit_hipModuleOccupancyMaxActiveBlocksPerMultiprocessorWithFlags_Negative_Parameters",
|
||||
"Unit_hipModuleOccupancyMaxPotentialBlockSizeWithFlags_Negative_Parameters",
|
||||
"Unit_hipGraphMemcpyNodeSetParamsToSymbol_Positive_Basic",
|
||||
"Unit_hipKernelNameRef_Negative_Parameters",
|
||||
"Unit_hipMemAdvise_No_Flag_Interference",
|
||||
"NOTE: The following 2 tests are disabled due to defect - EXSWHTEC-238",
|
||||
"Unit_hipDrvMemcpy3D_Positive_Array",
|
||||
"Unit_hipDrvMemcpy3DAsync_Positive_Array",
|
||||
"Unit_hipMemRangeGetAttribute_Positive_AccessedBy_Basic",
|
||||
"Unit_hipMemRangeGetAttribute_Positive_AccessedBy_Partial_Range",
|
||||
"Unit_hipMemGetAddressRange_Positive",
|
||||
"Unit_hipGraphAddMemcpyNode1D_Negative_Basic",
|
||||
"Unit_ChannelDescriptor_Positive_16BitFloatingPoint",
|
||||
"intermittent issue: failure expected but sucess returned",
|
||||
"Unit_hipMemAdvise_NegtveTsts",
|
||||
"Note: Following four tests disabled due to defect - EXSWHTEC-203",
|
||||
"Unit_hipStreamSetCaptureDependencies_Positive_Functional",
|
||||
"Note: Test disabled due to defect - EXSWHTEC-207",
|
||||
"Unit_hipStreamCreateWithFlags_DefaultStreamInteraction",
|
||||
"Unit_hipMemset3DSync",
|
||||
"Unit_hipStreamAddCallback_StrmSyncTiming",
|
||||
"Disabling test tracked SWDEV-394199",
|
||||
"Unit_hipStreamCreateWithPriority_MulthreadNonblockingflag",
|
||||
"Disabling test tracked SWDEV-395683",
|
||||
"Unit_hipStreamPerThread_MultiThread",
|
||||
"Disabling tests tracked with SWDEV-389647..",
|
||||
"Unit_hipMemcpy2DToArrayAsync_Positive_Synchronization_Behavior",
|
||||
"Disabling test tracked SWDEV-391555",
|
||||
"Unit_hipMemcpyPeer_Positive_ZeroSize",
|
||||
"Unit_hipMemcpyPeerAsync_Positive_ZeroSize",
|
||||
"Fails in Stress test SWDEV-398971",
|
||||
"SWDEV-398977 fails in stress tests",
|
||||
"Unit_hipMemset2DSync",
|
||||
"SWDEV-398981 fails in stress test",
|
||||
"Unit_hipStreamCreateWithPriority_MulthreadDefaultflag",
|
||||
"SWDEV-402054 fails in external github build",
|
||||
"Unit_hipEventDestroy_WithWaitingStream",
|
||||
"=== Below tests fail in stress test on 30/06/23 ===",
|
||||
"Unit_hipMemcpyParam2DAsync_multiDevice-StreamOnDiffDevice",
|
||||
"=== Below tests fail in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/96 ===",
|
||||
"Unit_hipHostGetDevicePointer_Negative",
|
||||
"Unit_hipExtModuleLaunchKernel_NonUniformWorkGroup",
|
||||
"=== Below tests fail in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/18 ===",
|
||||
"Unit_hipMemcpyAsync_Negative_Parameters",
|
||||
"Unit_hipMemcpyDtoHAsync_Negative_Parameters",
|
||||
"Unit_hipMemcpyHtoDAsync_Negative_Parameters",
|
||||
"Unit_hipMemcpyDtoDAsync_Negative_Parameters",
|
||||
"Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_1",
|
||||
"=== Below tests fail in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/327 ===",
|
||||
"Unit_hiprtcDisabledSlpVectorizeComplrOptnTst",
|
||||
"Unit_hiprtcCombiComplrOptnTst",
|
||||
"=== Below tests fail in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/92 ===",
|
||||
"Unit_hipGetChannelDesc_Negative_Parameters",
|
||||
"Unit_hipGraphAddChildGraphNode_CmplxNstGrph_UpdKerFun_Clone",
|
||||
"=== Below tests fail in stress test on 24/07/23 ===",
|
||||
"Unit_hipStreamCreateWithPriority_ValidateWithEvents",
|
||||
"Unit_hipEventIpc",
|
||||
"=== SWDEV-427101:Below test fails randomly in PSDB ===",
|
||||
"Unit_deviceAllocation_InOneThread_AccessInAllThreads",
|
||||
"=== Below 2 tests are disable due to defect EXSWHTEC-356 ===",
|
||||
"Unit_Device___hisinf2_Accuracy_Positive",
|
||||
"Unit_Device___hisnan2_Accuracy_Positive",
|
||||
@@ -105,20 +44,11 @@
|
||||
"Unit_Device___uhadd_Sanity_Positive",
|
||||
"Unit_Device___rhadd_Sanity_Positive",
|
||||
"Unit_Device___urhadd_Sanity_Positive",
|
||||
"Unit_hipGraphAddMemcpyNode_Negative_Parameters",
|
||||
"=== Below 2 tests are disable due to defect EXSWHTEC-369 ===",
|
||||
"Unit_Device_ilogbf_Accuracy_Positive",
|
||||
"Unit_Device_ilogb_Accuracy_Positive",
|
||||
"NOTE: The following test is disabled due to defect - EXSWHTEC-245",
|
||||
"Unit_hipMemCreate_MapNonContiguousChunks",
|
||||
"Unit_hipMemMap_PhysicalMemoryReuse_MultiDev",
|
||||
"Unit_hipMemMap_VMMMemoryReuse_MultiGPU",
|
||||
"Unit_hipMemSetAccess_FuncTstOnMultDev",
|
||||
"Unit_hipMemSetAccess_Vmm2PeerDevMemCpy",
|
||||
"Unit_hipMemSetAccess_Vmm2VMMInterDevMemCpy",
|
||||
"Unit_hipMemSetAccess_GrowVMM",
|
||||
"Unit_hipMemMap_PhysicalMemory_Map2MultVMMs",
|
||||
"Unit_hipMemSetAccess_MultiProc",
|
||||
"=== SWDEV-434171: Below tests took long time to complete in stress test on 17/11/23 ===",
|
||||
"Unit_Warp_Shfl_Positive_Basic - int",
|
||||
"Unit_Warp_Shfl_Positive_Basic - unsigned int",
|
||||
@@ -141,7 +71,6 @@
|
||||
"Unit_Warp_Shfl_XOR_Positive_Basic - __half",
|
||||
"Unit_Warp_Shfl_XOR_Positive_Basic - __half2",
|
||||
"=== SWDEV-434878: Below tests failed in stress test on 24/11/23 ===",
|
||||
"Unit_hipGraphUpload_Negative_Parameters",
|
||||
"Unit_hipModuleOccupancyMaxPotentialBlockSize_Negative_Parameters",
|
||||
"Unit_hipModuleOccupancyMaxPotentialBlockSize_Positive_RangeValidation",
|
||||
"Unit_hipModuleOccupancyMaxPotentialBlockSizeWithFlags_Positive_RangeValidation",
|
||||
@@ -168,25 +97,8 @@
|
||||
"Unit_hipGraphicsResourceGetMappedPointer_Negative_Parameters",
|
||||
"Unit_hipGraphicsUnmapResources_Negative_Parameters",
|
||||
"Unit_hipGraphicsUnregisterResource_Negative_Parameters",
|
||||
"SWDEV-443760: This test fails when device memory is used for kernel args",
|
||||
"=== Below tests fail in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/356 ===",
|
||||
"Note: Test disabled due to defect - EXSWHTEC-151",
|
||||
"Unit_hipModuleLoad_Negative_Load_From_A_File_That_Is_Not_A_Module",
|
||||
"Note: Following two tests disabled due to defect - EXSWHTEC-153",
|
||||
"Unit_hipModuleLoadData_Negative_Image_Is_An_Empty_String",
|
||||
"Unit_hipModuleLoadDataEx_Negative_Image_Is_An_Empty_String",
|
||||
"Note: Test disabled due to defect - EXSWHTEC-163",
|
||||
"Unit_hipModuleGetGlobal_Negative_Hmod_Is_Nullptr",
|
||||
"Note: Test disabled due to defect - EXSWHTEC-164",
|
||||
"Unit_hipModuleGetGlobal_Negative_Name_Is_Empty_String",
|
||||
"Note: Test disabled due to defect - EXSWHTEC-165",
|
||||
"Unit_hipModuleGetGlobal_Negative_Dptr_And_Bytes_Are_Nullptr",
|
||||
"Note: Test disabled due to defect - EXSWHTEC-166",
|
||||
"Unit_hipModuleGetTexRef_Negative_Hmod_Is_Nullptr",
|
||||
"Note: Test disabled due to defect - EXSWHTEC-167",
|
||||
"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",
|
||||
@@ -196,30 +108,16 @@
|
||||
"Unit_hipMemcpyParam2D_Positive_Synchronization_Behavior",
|
||||
"Unit_hipDrvMemcpy3D_Positive_Synchronization_Behavior",
|
||||
"Unit_hipLaunchCooperativeKernel_Negative_Parameters",
|
||||
"Unit_hipDrvGraphAddMemsetNode_hipMalloc3D_2D",
|
||||
"Unit_hipDrvGraphAddMemsetNode_hipMalloc3D_1D",
|
||||
"Unit_hipDrvGraphAddMemsetNode_hipMalloc_1D",
|
||||
"Unit_hipDrvGraphAddMemsetNode_hipMallocManaged",
|
||||
"Unit_hipExtModuleLaunchKernel_Negative_Parameters",
|
||||
"Unit_hipLaunchKernel_Negative_Parameters",
|
||||
"Unit_hipModuleLaunchCooperativeKernel_Negative_Parameters",
|
||||
"Unit_Device_modf_modff_Negative_RTC",
|
||||
"SWDEV-446588 - Disable graph multi gpu testcases until graph has support for it",
|
||||
"Unit_hipGraphExecUpdate_Negative_MultiDevice_Context_Changed",
|
||||
"Unit_hipGraphMem_Alloc_Free_NodeGetParams_Functional_MultiDevice",
|
||||
"Unit_hipGraphUpload_Functional_multidevice_test",
|
||||
"=== Below tests fail in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/210 ===",
|
||||
"Unit_Assert_Positive_Basic_KernelFail",
|
||||
"Unit_Coalesced_Group_Tiled_Partition_Sync_Positive_Basic - uint8_t",
|
||||
"Unit_Coalesced_Group_Tiled_Partition_Sync_Positive_Basic - uint16_t",
|
||||
"Unit_Coalesced_Group_Tiled_Partition_Sync_Positive_Basic - uint32_t",
|
||||
"=== SWDEV-444987 - Below tests fail in stress testing on 25/01/2023 ===",
|
||||
"Unit_floatTM",
|
||||
"Unit_TestMathFuncComplex",
|
||||
"Unit_AtomicsWithRandomActiveLanesInWavefront_UniformInteger",
|
||||
"Unit_AtomicsWithRandomActiveLanesInWavefront_DivergentInteger",
|
||||
"Unit_hipGraphAddMemcpyNodeToSymbol_Positive_Basic",
|
||||
"Unit_hipStreamBeginCapture_Positive_Functional",
|
||||
"Unit_atomicAnd_Negative_Parameters_RTC",
|
||||
"Unit_atomicOr_Negative_Parameters_RTC",
|
||||
"Unit_atomicXor_Negative_Parameters_RTC",
|
||||
@@ -694,14 +592,10 @@
|
||||
"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",
|
||||
"Unit_hipGraphInstantiateWithFlags_StreamCaptureDeviceContextChg",
|
||||
"=== SWDEV-457316 Below tests are disabled temporarily to avoid combined PSDB ===",
|
||||
"Unit_hipGraphAddMemFreeNode_Negative_NotSupported",
|
||||
"=== These tests fail on linux PSDB 21/11/24 ===",
|
||||
"Unit_atomicMax_Positive_Multi_Kernel_Same_Address - double",
|
||||
"Unit_atomicMax_Positive_Multi_Kernel_Same_Address - float",
|
||||
@@ -715,9 +609,6 @@
|
||||
"Unit_unsafeAtomicMax_Positive_Multi_Kernel_Same_Address - float",
|
||||
"Unit_unsafeAtomicMin_Positive_Multi_Kernel_Same_Address - double",
|
||||
"Unit_unsafeAtomicMin_Positive_Multi_Kernel_Same_Address - float",
|
||||
"=== SWDEV-475482 - Disable tests to merge clr change ===",
|
||||
"Unit_hipCreateTextureObject_LinearResource",
|
||||
"Unit_hipCreateTextureObject_Pitch2DResource",
|
||||
"=== SWDEV-454316 : Below tests fail in stress test ===",
|
||||
"Unit_atomicMin_system_Positive_Peer_GPUs_Same_Address - float",
|
||||
"Unit_atomicMin_system_Positive_Peer_GPUs_Same_Address - double",
|
||||
@@ -727,12 +618,7 @@
|
||||
"Unit_hipIpcOpenMemHandle_Negative_Open_In_Two_Contexts_Same_Device",
|
||||
"Unit_hipIpcCloseMemHandle_Positive_Reference_Counting",
|
||||
"=== SWDEV-517063 Below tests are temporarily disabled due to PSDB failure",
|
||||
"Unit_hipGraphInstantiateWithFlags_FlagAutoFreeOnLaunch_check",
|
||||
"Unit_hipGraphInstantiateWithFlags_AutoFreeOnLaunchInLoop",
|
||||
"Unit_hipGraphInstantiateWithFlags_AutoFreeOnLaunchFillKernel",
|
||||
"Unit_hipGraphInstantiateWithFlags_AutoFreeOnLaunchDoubleKernel",
|
||||
"Unit_hipGraphInstantiateWithFlags_AutoFreeOnLaunchMultiProcess",
|
||||
"Unit_hipGraphInstantiateWithFlags_WithDefaultAndAutoFreeOnLaunch",
|
||||
"=== SWDEV-457316 Below test is skipped due ref count logic (Discussed with German) ===",
|
||||
"Unit_hipGraphAddMemAllocNode_Negative_Free_Alloc_Memory_Again",
|
||||
"=== SWDEV-530762 : This test fails in Linux PSDB ===",
|
||||
|
||||
@@ -37,8 +37,6 @@ THE SOFTWARE.
|
||||
* - @ref Unit_hipEventIpc
|
||||
*/
|
||||
|
||||
#if HT_AMD /* Disabled because frequency based wait is timing out on nvidia platforms */
|
||||
|
||||
static constexpr size_t vectorSize{1024};
|
||||
|
||||
/*
|
||||
@@ -141,15 +139,7 @@ TEST_CASE("Unit_hipEventDestroy_Negative") {
|
||||
hipEvent_t event{nullptr};
|
||||
HIP_CHECK_ERROR(hipEventDestroy(event), hipErrorInvalidResourceHandle);
|
||||
}
|
||||
|
||||
SECTION("Destroy twice") {
|
||||
hipEvent_t event;
|
||||
HIP_CHECK(hipEventCreate(&event));
|
||||
HIP_CHECK(hipEventDestroy(event));
|
||||
HIP_CHECK_ERROR(hipEventDestroy(event), hipErrorContextIsDestroyed);
|
||||
}
|
||||
}
|
||||
#endif
|
||||
|
||||
TEST_CASE("Unit_hipEventDestroy_Verify_Capture") {
|
||||
hipEvent_t event;
|
||||
|
||||
@@ -648,115 +648,6 @@ TEST_CASE("Unit_hipGraphInstantiateWithFlags_AutoFreeOnLaunchDoubleKernel") {
|
||||
delete[] hostMemDst;
|
||||
}
|
||||
|
||||
#if __linux__
|
||||
/**
|
||||
* Test Description
|
||||
* ------------------------
|
||||
* - This test case tests hipGraphInstantiateWithFlags with the flag
|
||||
* - hipGraphInstantiateFlagAutoFreeOnLaunch for multi process scenario :
|
||||
* - 1) Take a shared memory and fill it.
|
||||
* - 2) Create child process.
|
||||
* - 3) In child process, create graph with following nodes,
|
||||
* - a. Node to allocate memory - memAllocNode
|
||||
* - b. Node to copy from shared memory to device - memcpyNodeH2D
|
||||
* - c. Node to perform double operation - kernelNode
|
||||
* - d. Node to copy from device to shared memory - memcpyNodeD2H
|
||||
* - 4) Wait in parent process to complete child process task and
|
||||
* - shared memory should contain the expected value and
|
||||
* - it should not give memory related issues.
|
||||
* Test source
|
||||
* ------------------------
|
||||
* - unit/graph/hipGraphInstantiateWithFlags.cc
|
||||
*/
|
||||
TEST_CASE("Unit_hipGraphInstantiateWithFlags_AutoFreeOnLaunchMultiProcess") {
|
||||
int shmid = shmget(IPC_PRIVATE, NBYTES, 0666);
|
||||
int* shared_mem = reinterpret_cast<int*>(shmat(shmid, NULL, 0));
|
||||
REQUIRE(shared_mem != nullptr);
|
||||
|
||||
std::fill(shared_mem, shared_mem + SIZE, 10);
|
||||
|
||||
auto pid = fork();
|
||||
|
||||
if (pid != 0) { // parent process
|
||||
REQUIRE(wait(NULL) >= 0);
|
||||
|
||||
for (int idx = 0; idx < SIZE; idx++) {
|
||||
INFO("At index : " << idx << ", Got value : " << shared_mem[idx] << ", Expected value : 20"
|
||||
<< "\n");
|
||||
REQUIRE(shared_mem[idx] == 20);
|
||||
}
|
||||
} else { // child process
|
||||
REQUIRE(shared_mem != nullptr);
|
||||
|
||||
hipGraph_t graph;
|
||||
HIP_CHECK(hipGraphCreate(&graph, 0));
|
||||
|
||||
hipStream_t stream;
|
||||
HIP_CHECK(hipStreamCreate(&stream));
|
||||
REQUIRE(stream != nullptr);
|
||||
|
||||
int* devMem = nullptr;
|
||||
|
||||
hipGraphNode_t memAllocNode, memcpyNodeH2D, kernelNode, memcpyNodeD2H;
|
||||
|
||||
hipMemAllocNodeParams memAllocNodeParams{};
|
||||
memAllocNodeParams.poolProps.allocType = hipMemAllocationTypePinned;
|
||||
memAllocNodeParams.poolProps.handleTypes = hipMemHandleTypeNone;
|
||||
memAllocNodeParams.poolProps.location.type = hipMemLocationTypeDevice;
|
||||
memAllocNodeParams.poolProps.location.id = 0;
|
||||
memAllocNodeParams.bytesize = NBYTES;
|
||||
|
||||
HIP_CHECK(hipGraphAddMemAllocNode(&memAllocNode, graph, nullptr, 0, &memAllocNodeParams));
|
||||
devMem = reinterpret_cast<int*>(memAllocNodeParams.dptr);
|
||||
REQUIRE(devMem != nullptr);
|
||||
|
||||
::std::vector<hipGraphNode_t> memcpyNodeH2DDependencies;
|
||||
memcpyNodeH2DDependencies.push_back(memAllocNode);
|
||||
|
||||
HIP_CHECK(hipGraphAddMemcpyNode1D(&memcpyNodeH2D, graph, memcpyNodeH2DDependencies.data(),
|
||||
memcpyNodeH2DDependencies.size(), devMem, shared_mem, NBYTES,
|
||||
hipMemcpyHostToDevice));
|
||||
|
||||
::std::vector<hipGraphNode_t> kernelNodeDependencies;
|
||||
kernelNodeDependencies.push_back(memcpyNodeH2D);
|
||||
|
||||
hipKernelNodeParams kernelNodeParams{};
|
||||
kernelNodeParams.func = reinterpret_cast<void*>(doubleKernel);
|
||||
kernelNodeParams.gridDim = dim3(1, 1, 1);
|
||||
kernelNodeParams.blockDim = dim3(1, 1, 1);
|
||||
kernelNodeParams.sharedMemBytes = 0;
|
||||
int size = SIZE;
|
||||
void* kernelArgs[2] = {reinterpret_cast<void*>(&devMem), reinterpret_cast<void*>(&size)};
|
||||
kernelNodeParams.kernelParams = kernelArgs;
|
||||
kernelNodeParams.extra = nullptr;
|
||||
|
||||
HIP_CHECK(hipGraphAddKernelNode(&kernelNode, graph, kernelNodeDependencies.data(),
|
||||
kernelNodeDependencies.size(), &kernelNodeParams));
|
||||
|
||||
::std::vector<hipGraphNode_t> memcpyNodeD2HDependencies;
|
||||
memcpyNodeD2HDependencies.push_back(kernelNode);
|
||||
|
||||
HIP_CHECK(hipGraphAddMemcpyNode1D(&memcpyNodeD2H, graph, memcpyNodeD2HDependencies.data(),
|
||||
memcpyNodeD2HDependencies.size(), shared_mem, devMem, NBYTES,
|
||||
hipMemcpyDeviceToHost));
|
||||
|
||||
hipGraphExec_t graphExec;
|
||||
HIP_CHECK(
|
||||
hipGraphInstantiateWithFlags(&graphExec, graph, hipGraphInstantiateFlagAutoFreeOnLaunch));
|
||||
|
||||
HIP_CHECK(hipGraphLaunch(graphExec, stream));
|
||||
HIP_CHECK(hipStreamSynchronize(stream));
|
||||
|
||||
HIP_CHECK(hipGraphExecDestroy(graphExec));
|
||||
HIP_CHECK(hipGraphDestroy(graph));
|
||||
HIP_CHECK(hipStreamDestroy(stream));
|
||||
HIP_CHECK(hipFree(devMem));
|
||||
}
|
||||
shmdt(shared_mem);
|
||||
shmctl(shmid, IPC_RMID, 0);
|
||||
}
|
||||
#endif
|
||||
|
||||
/**
|
||||
* Test Description
|
||||
* ------------------------
|
||||
|
||||
@@ -36,12 +36,10 @@ TEST_CASE("Unit_hipMemcpyParam2DAsync_Positive_Basic") {
|
||||
const StreamGuard stream_guard(stream_type);
|
||||
const hipStream_t stream = stream_guard.stream();
|
||||
|
||||
#if HT_NVIDIA // Disabled on AMD due to defect - EXSWHTEC-236
|
||||
SECTION("Device to Host") {
|
||||
Memcpy2DDeviceToHostShell<async>(
|
||||
std::bind(MemcpyParam2DAdapter<async>(), _1, _2, _3, _4, _5, _6, _7, stream), stream);
|
||||
}
|
||||
#endif
|
||||
SECTION("Device to Device") {
|
||||
SECTION("Peer access disabled") {
|
||||
Memcpy2DDeviceToDeviceShell<async, false>(
|
||||
@@ -56,12 +54,10 @@ TEST_CASE("Unit_hipMemcpyParam2DAsync_Positive_Basic") {
|
||||
Memcpy2DHostToDeviceShell<async>(
|
||||
std::bind(MemcpyParam2DAdapter<async>(), _1, _2, _3, _4, _5, _6, _7, stream), stream);
|
||||
}
|
||||
#if HT_NVIDIA // Disabled on AMD due to defect - EXSWHTEC-236
|
||||
SECTION("Host to Host") {
|
||||
Memcpy2DHostToHostShell<async>(
|
||||
std::bind(MemcpyParam2DAdapter<async>(), _1, _2, _3, _4, _5, _6, _7, stream), stream);
|
||||
}
|
||||
#endif
|
||||
}
|
||||
|
||||
TEST_CASE("Unit_hipMemcpyParam2DAsync_Positive_Synchronization_Behavior") {
|
||||
|
||||
@@ -440,10 +440,6 @@ static void doMemsetTest(allocType mallocType, memSetType memset_type, MultiDDat
|
||||
}
|
||||
|
||||
TEST_CASE("Unit_hipMemsetSync") {
|
||||
#if HT_NVIDIA
|
||||
HipTest::HIP_SKIP_TEST("EXSWCPHIPT-86");
|
||||
return;
|
||||
#endif
|
||||
allocType type = GENERATE(allocType::deviceMalloc, allocType::hostMalloc, allocType::hostRegisted,
|
||||
allocType::devRegistered);
|
||||
memSetType memset_type = memSetType::hipMemset;
|
||||
@@ -453,10 +449,6 @@ TEST_CASE("Unit_hipMemsetSync") {
|
||||
}
|
||||
|
||||
TEMPLATE_TEST_CASE("Unit_hipMemsetDSync", "", int8_t, int16_t, uint32_t) {
|
||||
#if HT_NVIDIA
|
||||
HipTest::HIP_SKIP_TEST("EXSWCPHIPT-86");
|
||||
return;
|
||||
#endif
|
||||
allocType mallocType = GENERATE(allocType::hostRegisted, allocType::deviceMalloc,
|
||||
allocType::hostMalloc, allocType::devRegistered);
|
||||
memSetType memset_type;
|
||||
@@ -475,10 +467,6 @@ TEMPLATE_TEST_CASE("Unit_hipMemsetDSync", "", int8_t, int16_t, uint32_t) {
|
||||
}
|
||||
|
||||
TEST_CASE("Unit_hipMemset2DSync") {
|
||||
#if HT_NVIDIA
|
||||
HipTest::HIP_SKIP_TEST("EXSWCPHIPT-86");
|
||||
return;
|
||||
#endif
|
||||
allocType mallocType = GENERATE(allocType::deviceMalloc, allocType::hostMalloc,
|
||||
allocType::hostRegisted, allocType::devRegistered);
|
||||
memSetType memset_type = memSetType::hipMemset2D;
|
||||
@@ -490,10 +478,6 @@ TEST_CASE("Unit_hipMemset2DSync") {
|
||||
}
|
||||
|
||||
TEST_CASE("Unit_hipMemset3DSync") {
|
||||
#if HT_NVIDIA
|
||||
HipTest::HIP_SKIP_TEST("EXSWCPHIPT-86");
|
||||
return;
|
||||
#endif
|
||||
allocType mallocType = GENERATE(allocType::deviceMalloc, allocType::hostMalloc,
|
||||
allocType::hostRegisted, allocType::devRegistered);
|
||||
memSetType memset_type = memSetType::hipMemset3D;
|
||||
|
||||
@@ -84,22 +84,15 @@ TEST_CASE("Unit_hipCreateTextureObject_LinearResource") {
|
||||
}
|
||||
|
||||
SECTION("hipResourceTypeLinear and sizeInBytes(0)") {
|
||||
if ((TestContext::get()).isAmd()) {
|
||||
// Populate resource descriptor
|
||||
resDesc.res.linear.devPtr = texBuf;
|
||||
resDesc.res.linear.desc = hipCreateChannelDesc(xsize, 0, 0, 0, hipChannelFormatKindFloat);
|
||||
resDesc.res.linear.sizeInBytes = 0;
|
||||
// Populate resource descriptor
|
||||
resDesc.res.linear.devPtr = texBuf;
|
||||
resDesc.res.linear.desc = hipCreateChannelDesc(xsize, 0, 0, 0, hipChannelFormatKindFloat);
|
||||
resDesc.res.linear.sizeInBytes = 0;
|
||||
|
||||
// Populate texture descriptor
|
||||
texDesc.readMode = hipReadModeElementType;
|
||||
ret = hipCreateTextureObject(&texObj, &resDesc, &texDesc, nullptr);
|
||||
REQUIRE(ret != hipSuccess);
|
||||
} else {
|
||||
// API expected to return failure. Test skipped
|
||||
// on nvidia as api returns success and would lead
|
||||
// to unexpected behavior with app.
|
||||
WARN("Resource type Linear/sizeInBytes(0) skipped on nvidia");
|
||||
}
|
||||
// Populate texture descriptor
|
||||
texDesc.readMode = hipReadModeElementType;
|
||||
HIP_CHECK(hipCreateTextureObject(&texObj, &resDesc, &texDesc, nullptr));
|
||||
HIP_CHECK(hipDestroyTextureObject(texObj));
|
||||
}
|
||||
|
||||
SECTION("hipResourceTypeLinear and sizeInBytes(max(size_t))") {
|
||||
|
||||
@@ -130,23 +130,18 @@ TEST_CASE("Unit_hipCreateTextureObject_Pitch2DResource") {
|
||||
}
|
||||
|
||||
SECTION("hipResourceTypePitch2D and height(0)") {
|
||||
if ((TestContext::get()).isAmd()) {
|
||||
// Populate resource descriptor
|
||||
resDesc.res.pitch2D.devPtr = devPtrA;
|
||||
resDesc.res.pitch2D.height = 0;
|
||||
resDesc.res.pitch2D.width = SIZE_W;
|
||||
resDesc.res.pitch2D.pitchInBytes = devPitchA;
|
||||
resDesc.res.pitch2D.desc = hipCreateChannelDesc<float>();
|
||||
// Populate resource descriptor
|
||||
resDesc.res.pitch2D.devPtr = devPtrA;
|
||||
resDesc.res.pitch2D.height = 0;
|
||||
resDesc.res.pitch2D.width = SIZE_W;
|
||||
resDesc.res.pitch2D.pitchInBytes = devPitchA;
|
||||
resDesc.res.pitch2D.desc = hipCreateChannelDesc<float>();
|
||||
|
||||
// Populate texture descriptor
|
||||
texDesc.readMode = hipReadModeElementType;
|
||||
// Populate texture descriptor
|
||||
texDesc.readMode = hipReadModeElementType;
|
||||
|
||||
ret = hipCreateTextureObject(&texObj, &resDesc, &texDesc, nullptr);
|
||||
REQUIRE(ret != hipSuccess);
|
||||
} else {
|
||||
// Test expected to return error with height(0).
|
||||
WARN("Resourcetype Pitch2D/height(0) skipped on nvidia");
|
||||
}
|
||||
HIP_CHECK(hipCreateTextureObject(&texObj, &resDesc, &texDesc, nullptr));
|
||||
HIP_CHECK(hipDestroyTextureObject(texObj));
|
||||
}
|
||||
|
||||
SECTION("hipResourceTypePitch2D and height(0)/devptr(nullptr)") {
|
||||
@@ -180,24 +175,19 @@ TEST_CASE("Unit_hipCreateTextureObject_Pitch2DResource") {
|
||||
}
|
||||
|
||||
SECTION("hipResourceTypePitch2D and width(0)") {
|
||||
if ((TestContext::get()).isAmd()) {
|
||||
// Populate resource descriptor
|
||||
resDesc.resType = hipResourceTypePitch2D;
|
||||
resDesc.res.pitch2D.devPtr = devPtrA;
|
||||
resDesc.res.pitch2D.height = SIZE_H;
|
||||
resDesc.res.pitch2D.width = 0;
|
||||
resDesc.res.pitch2D.pitchInBytes = devPitchA;
|
||||
resDesc.res.pitch2D.desc = hipCreateChannelDesc<float>();
|
||||
// Populate resource descriptor
|
||||
resDesc.resType = hipResourceTypePitch2D;
|
||||
resDesc.res.pitch2D.devPtr = devPtrA;
|
||||
resDesc.res.pitch2D.height = SIZE_H;
|
||||
resDesc.res.pitch2D.width = 0;
|
||||
resDesc.res.pitch2D.pitchInBytes = devPitchA;
|
||||
resDesc.res.pitch2D.desc = hipCreateChannelDesc<float>();
|
||||
|
||||
// Populate texture descriptor
|
||||
texDesc.readMode = hipReadModeElementType;
|
||||
// Populate texture descriptor
|
||||
texDesc.readMode = hipReadModeElementType;
|
||||
|
||||
ret = hipCreateTextureObject(&texObj, &resDesc, &texDesc, nullptr);
|
||||
REQUIRE(ret != hipSuccess);
|
||||
} else {
|
||||
// api expected to return failure when width(0) is passed.
|
||||
WARN("ResourceType Pitch2D/width(0) skipped on nvidia");
|
||||
}
|
||||
HIP_CHECK(hipCreateTextureObject(&texObj, &resDesc, &texDesc, nullptr));
|
||||
HIP_CHECK(hipDestroyTextureObject(texObj));
|
||||
}
|
||||
|
||||
SECTION("hipResourceTypePitch2D and width(0)/devPtr(nullptr)") {
|
||||
|
||||
Ссылка в новой задаче
Block a user