diff --git a/projects/clr/hipamd/src/hip_memory.cpp b/projects/clr/hipamd/src/hip_memory.cpp index 772ee9934d..5cebd56a8f 100644 --- a/projects/clr/hipamd/src/hip_memory.cpp +++ b/projects/clr/hipamd/src/hip_memory.cpp @@ -1801,6 +1801,7 @@ hipError_t ihipMemcpyDtoHCommand(amd::Command*& command, void* dstHost, amd::Coo waitList.push_back(cmd); } } + amd::ReadMemoryCommand* readCommand = new amd::ReadMemoryCommand(*stream, CL_COMMAND_READ_BUFFER_RECT, waitList, *srcMemory, srcStart, copyRegion, dstHost, srcRect, dstRect, copyMetadata); @@ -1813,6 +1814,10 @@ hipError_t ihipMemcpyDtoHCommand(amd::Command*& command, void* dstHost, amd::Coo return hipErrorInvalidValue; } command = readCommand; + + if (!waitList.empty()) { + waitList[0]->release(); + } } return hipSuccess; @@ -1949,6 +1954,10 @@ hipError_t ihipMemcpyHtoACommand(amd::Command*& command, amd::Image* dstImage, return hipErrorInvalidValue; } command = writeMemCmd; + + if (!waitList.empty()) { + waitList[0]->release(); + } } return hipSuccess; @@ -1997,6 +2006,10 @@ hipError_t ihipMemcpyAtoHCommand(amd::Command*& command, void* dstHost, amd::Coo return hipErrorInvalidValue; } command = readMemCmd; + + if (!waitList.empty()) { + waitList[0]->release(); + } } return hipSuccess; diff --git a/projects/clr/hipamd/src/hip_mempool.cpp b/projects/clr/hipamd/src/hip_mempool.cpp index 6fcb3c869c..f80c564389 100644 --- a/projects/clr/hipamd/src/hip_mempool.cpp +++ b/projects/clr/hipamd/src/hip_mempool.cpp @@ -425,7 +425,7 @@ hipError_t hipMemPoolImportFromShareableHandle(hipMemPool_t* mem_pool, void* sha } auto device = g_devices[0]; - auto pool = new hip::MemoryPool(device); + auto pool = new hip::MemoryPool(device, nullptr, true); if (pool == nullptr) { HIP_RETURN(hipErrorOutOfMemory); }