diff --git a/catch/hipTestMain/config/config_amd_linux b/catch/hipTestMain/config/config_amd_linux index 59c23e395f..679c418f66 100644 --- a/catch/hipTestMain/config/config_amd_linux +++ b/catch/hipTestMain/config/config_amd_linux @@ -123,7 +123,6 @@ "Unit_hipMemSetAccess_FuncTstOnMultDev", "Unit_hipMemSetAccess_Vmm2PeerDevMemCpy", "Unit_hipMemSetAccess_Vmm2VMMInterDevMemCpy", - "Unit_hipMemSetAccess_MapPhysChksFromMulDev", "Unit_hipMemSetAccess_GrowVMM", "Unit_hipMemMap_PhysicalMemory_Map2MultVMMs", "Unit_hipMemSetAccess_MultiProc", diff --git a/catch/hipTestMain/config/config_amd_windows b/catch/hipTestMain/config/config_amd_windows index 26b2cec602..f965d7cda8 100644 --- a/catch/hipTestMain/config/config_amd_windows +++ b/catch/hipTestMain/config/config_amd_windows @@ -235,7 +235,6 @@ "Unit_hipMemSetAccess_Vmm2UnifiedMemCpy", "Unit_hipMemSetAccess_Vmm2DevMemCpy", "Unit_hipMemSetAccess_Vmm2VMMInterDevMemCpy", - "Unit_hipMemSetAccess_MapPhysChksFromMulDev", "Unit_hipMemSetAccess_GrowVMM", "Unit_hipMemSetAccess_negative", "=== SWDEV-434171: Below tests took long time to complete in stress test on 17/11/23 ===", diff --git a/catch/hipTestMain/config/config_amd_wsl.json b/catch/hipTestMain/config/config_amd_wsl.json index b7aca166ca..857cc32fe0 100644 --- a/catch/hipTestMain/config/config_amd_wsl.json +++ b/catch/hipTestMain/config/config_amd_wsl.json @@ -257,7 +257,6 @@ "Unit_hipMemSetAccess_Vmm2PeerPeerMemCpy", "Unit_hipMemSetAccess_Vmm2VMMMemCpy", "Unit_hipMemSetAccess_Vmm2VMMInterDevMemCpy", - "Unit_hipMemSetAccess_MapPhysChksFromMulDev", "Unit_hipMemSetAccess_GrowVMM", "Unit_hipMemSetAccess_Multithreaded", "Unit_hipMemSetAccess_MultiProc", @@ -491,7 +490,6 @@ "Unit_hipMemSetAccess_Vmm2PeerPeerMemCpy", "Unit_hipMemSetAccess_Vmm2VMMMemCpy", "Unit_hipMemSetAccess_Vmm2VMMInterDevMemCpy", - "Unit_hipMemSetAccess_MapPhysChksFromMulDev", "Unit_hipMemSetAccess_GrowVMM", "Unit_hipMemSetAccess_Multithreaded", "Unit_hipMemSetAccess_MultiProc", diff --git a/catch/unit/virtualMemoryManagement/hipMemSetGetAccess.cc b/catch/unit/virtualMemoryManagement/hipMemSetGetAccess.cc index 4c8beadc77..250ee8911f 100644 --- a/catch/unit/virtualMemoryManagement/hipMemSetGetAccess.cc +++ b/catch/unit/virtualMemoryManagement/hipMemSetGetAccess.cc @@ -937,89 +937,6 @@ TEST_CASE("Unit_hipMemSetAccess_Vmm2VMMInterDevMemCpy") { HIP_CHECK(hipMemAddressFree(ptrA, size_mem)); } -/** - * Test Description - * ------------------------ - * - Allocate a chunk of memory and map it to device0. Allocate another - * chunk of memory and map it to device1. Check if these 2 distinct memory - * chunks can be mapped to a single address space. - * ------------------------ - * - unit/virtualMemoryManagement/hipMemSetGetAccess.cc - * Test requirements - * ------------------------ - * - HIP_VERSION >= 6.1 - */ -TEST_CASE("Unit_hipMemSetAccess_MapPhysChksFromMulDev") { - int devicecount = 0; - HIP_CHECK(hipGetDeviceCount(&devicecount)); - int numOfBuffers = devicecount; - constexpr int N = DATA_SIZE; - size_t buffer_size = N * sizeof(int), granularity = 0; - int deviceId = 0; - // Allocate resources for all gpus - hipMemGenericAllocationHandle_t* handle = static_cast( - malloc(sizeof(hipMemGenericAllocationHandle_t) * numOfBuffers)); - REQUIRE(handle != nullptr); - size_t* size_mem = static_cast(malloc(sizeof(size_t) * numOfBuffers)); - REQUIRE(size_mem != nullptr); - size_t total_mem = 0; - // Create memory chunks - for (deviceId = 0; deviceId < numOfBuffers; deviceId++) { - hipDevice_t device; - HIP_CHECK(hipDeviceGet(&device, deviceId)); - checkVMMSupported(device); - hipMemAllocationProp prop_loc{}; - prop_loc.type = hipMemAllocationTypePinned; - prop_loc.location.type = hipMemLocationTypeDevice; - prop_loc.location.id = device; // Current Devices - HIP_CHECK(hipMemGetAllocationGranularity(&granularity, &prop_loc, - hipMemAllocationGranularityMinimum)); - REQUIRE(granularity > 0); - size_mem[deviceId] = ((granularity + buffer_size - 1) / granularity) * granularity; - total_mem = total_mem + size_mem[deviceId]; - // Allocate physical memory chunks - HIP_CHECK(hipMemCreate(&handle[deviceId], size_mem[deviceId], &prop_loc, 0)); - } - // Allocate virtual address range for all the memory chunks - hipDeviceptr_t ptrA; - HIP_CHECK(hipMemAddressReserve(&ptrA, total_mem, 0, 0, 0)); - // Map the allocated chunks - for (deviceId = 0; deviceId < numOfBuffers; deviceId++) { - hipDevice_t device; - HIP_CHECK(hipDeviceGet(&device, deviceId)); - uint64_t uiptr = reinterpret_cast(ptrA); - uiptr = uiptr + deviceId * size_mem[deviceId]; - HIP_CHECK( - hipMemMap(reinterpret_cast(uiptr), size_mem[deviceId], 0, handle[deviceId], 0)); - HIP_CHECK(hipMemRelease(handle[deviceId])); - // Set access - hipMemAccessDesc accessDesc_loc = {}; - accessDesc_loc.location.type = hipMemLocationTypeDevice; - accessDesc_loc.location.id = device; - accessDesc_loc.flags = hipMemAccessFlagsProtReadWrite; - // Make the address accessible to deviceId - HIP_CHECK( - hipMemSetAccess(reinterpret_cast(uiptr), size_mem[deviceId], &accessDesc_loc, 1)); - } - std::vector A_h(numOfBuffers * N), B_h(numOfBuffers * N); - // Fill Data - for (int idx = 0; idx < (numOfBuffers * N); idx++) { - A_h[idx] = idx * idx; - } - HIP_CHECK(hipMemcpyHtoD(ptrA, A_h.data(), numOfBuffers * buffer_size)); - HIP_CHECK(hipMemcpyDtoH(B_h.data(), ptrA, numOfBuffers * buffer_size)); - // Validate Results - REQUIRE(true == std::equal(B_h.begin(), B_h.end(), A_h.data())); - for (deviceId = 0; deviceId < numOfBuffers; deviceId++) { - uint64_t uiptr = reinterpret_cast(ptrA); - uiptr = uiptr + deviceId * size_mem[deviceId]; - HIP_CHECK(hipMemUnmap(reinterpret_cast(uiptr), size_mem[deviceId])); - } - HIP_CHECK(hipMemAddressFree(ptrA, total_mem)); - free(handle); - free(size_mem); -} - class vmm_resize_class { size_t current_size_tot; size_t current_size_rounded_tot;