diff --git a/projects/clr/hipamd/src/hip_internal.hpp b/projects/clr/hipamd/src/hip_internal.hpp index 67733e15aa..12b61cb3f1 100644 --- a/projects/clr/hipamd/src/hip_internal.hpp +++ b/projects/clr/hipamd/src/hip_internal.hpp @@ -477,7 +477,7 @@ extern hipError_t ihipDeviceGetCount(int* count); extern int ihipGetDevice(); extern hipError_t ihipMalloc(void** ptr, size_t sizeBytes, unsigned int flags); -extern amd::Memory* getMemoryObject(const void* ptr, size_t& offset); +extern amd::Memory* getMemoryObject(const void* ptr, size_t& offset, size_t size = 0); extern amd::Memory* getMemoryObjectWithOffset(const void* ptr, const size_t size); extern void getStreamPerThread(hipStream_t& stream); extern hipError_t ihipUnbindTexture(textureReference* texRef); diff --git a/projects/clr/hipamd/src/hip_memory.cpp b/projects/clr/hipamd/src/hip_memory.cpp index 12da05183e..b8b1b2f778 100644 --- a/projects/clr/hipamd/src/hip_memory.cpp +++ b/projects/clr/hipamd/src/hip_memory.cpp @@ -28,7 +28,7 @@ #include "amdocl/cl_vk_amd.hpp" // ================================================================================================ -amd::Memory* getMemoryObject(const void* ptr, size_t& offset) { +amd::Memory* getMemoryObject(const void* ptr, size_t& offset, size_t size) { amd::Memory *memObj = amd::MemObjMap::FindMemObj(ptr); if (memObj != nullptr) { const char* hostPtr = reinterpret_cast(ptr); @@ -51,7 +51,8 @@ amd::Memory* getMemoryObject(const void* ptr, size_t& offset) { } } else { // If memObj not found, use arena_mem_obj. arena_mem_obj is null, if HMM and Xnack is disabled. - memObj = (hip::getCurrentDevice()->asContext()->svmDevices()[0])->GetArenaMemObj(ptr, offset); + memObj = (hip::getCurrentDevice()->asContext()->svmDevices()[0])->GetArenaMemObj( + ptr, offset, size); } return memObj; } @@ -2413,7 +2414,7 @@ hipError_t hipMemsetD32Async(hipDeviceptr_t dst, int value, size_t count, hipError_t ihipMemset3D_validate(hipPitchedPtr pitchedDevPtr, int value, hipExtent extent, size_t sizeBytes) { size_t offset = 0; - amd::Memory* memory = getMemoryObject(pitchedDevPtr.ptr, offset); + amd::Memory* memory = getMemoryObject(pitchedDevPtr.ptr, offset, sizeBytes); if (memory == nullptr) { return hipErrorInvalidValue;