From 637a82e8484f5b2c4109e046637eb07ccf18d9be Mon Sep 17 00:00:00 2001 From: foreman Date: Wed, 6 Nov 2019 18:00:44 -0500 Subject: [PATCH] P4 to Git Change 2026152 by cpaquot@cpaquot-ocl-lc-lnx on 2019/11/06 17:50:08 SWDEV-206239 - [HIP] RCCL: finegrain VRAM does not work Implemented fine grained VRAM allocation via ATOMICS. Affected files ... ... //depot/stg/opencl/drivers/opencl/api/hip/hip_memory.cpp#84 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/rocdevice.cpp#138 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/rocdevice.hpp#43 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/rocmemory.cpp#45 edit [ROCm/clr commit: 6d0c96993dfdc50c7e8f1aefe53104380df78f46] --- projects/clr/hipamd/api/hip/hip_memory.cpp | 7 +------ 1 file changed, 1 insertion(+), 6 deletions(-) diff --git a/projects/clr/hipamd/api/hip/hip_memory.cpp b/projects/clr/hipamd/api/hip/hip_memory.cpp index 6eed6e2c23..a9dcd5b55c 100644 --- a/projects/clr/hipamd/api/hip/hip_memory.cpp +++ b/projects/clr/hipamd/api/hip/hip_memory.cpp @@ -168,12 +168,7 @@ hipError_t hipExtMallocWithFlags(void** ptr, size_t sizeBytes, unsigned int flag HIP_RETURN(hipErrorInvalidValue); } - // XXX Till we have fine grained VRAM support... - if ((flags & hipDeviceMallocFinegrained) != 0) { - HIP_RETURN(hipErrorMemoryAllocation); - } - - HIP_RETURN(ihipMalloc(ptr, sizeBytes, (flags & hipDeviceMallocFinegrained)? CL_MEM_SVM_FINE_GRAIN_BUFFER | CL_MEM_SVM_ATOMICS: 0)); + HIP_RETURN(ihipMalloc(ptr, sizeBytes, (flags & hipDeviceMallocFinegrained)? CL_MEM_SVM_ATOMICS: 0)); } hipError_t hipMalloc(void** ptr, size_t sizeBytes) {