From cfb6a08bc01f986bee1e490e0430e19cd846d78e Mon Sep 17 00:00:00 2001 From: Ioannis Assiouras Date: Wed, 19 Feb 2025 16:35:31 +0000 Subject: [PATCH] SWDEV-516461 - Skip hipmemcpy call to device dst that is freed, for xnack Change-Id: Icd08675de881e70054e2ba7be17c62d095d33d8c --- .../hipGetProcAddressVmmApis.cc | 15 ++++++++++++++- 1 file changed, 14 insertions(+), 1 deletion(-) diff --git a/catch/unit/virtualMemoryManagement/hipGetProcAddressVmmApis.cc b/catch/unit/virtualMemoryManagement/hipGetProcAddressVmmApis.cc index 04d6011c0e..fe664bc3e7 100644 --- a/catch/unit/virtualMemoryManagement/hipGetProcAddressVmmApis.cc +++ b/catch/unit/virtualMemoryManagement/hipGetProcAddressVmmApis.cc @@ -46,6 +46,17 @@ TEST_CASE("Unit_hipGetProcAddress_VMM") { return; } + hipDeviceProp_t devProp; + int device; + bool xnackEnabled = false; + HIP_CHECK(hipGetDevice(&device)); + HIP_CHECK(hipGetDeviceProperties(&devProp, device)); + std::string gfxName(devProp.gcnArchName); + + if (gfxName.find("xnack+") != std::string::npos) { + xnackEnabled = true; + } + void* hipMemGetAllocationGranularity_ptr = nullptr; void* hipMemCreate_ptr = nullptr; void* hipMemAddressReserve_ptr = nullptr; @@ -242,7 +253,9 @@ TEST_CASE("Unit_hipGetProcAddress_VMM") { HIP_CHECK(dyn_hipMemRelease_ptr(handle)); // Performing operation on ptr, to check it is invalidated or not - REQUIRE(hipMemcpy(ptr, hostMem, Nbytes, hipMemcpyHostToDevice) + if (!xnackEnabled) { + REQUIRE(hipMemcpy(ptr, hostMem, Nbytes, hipMemcpyHostToDevice) == hipErrorInvalidValue); + } free(hostMem); }