AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

SWDEV-408687 - Prevent updating dstXInBytes/srcXInBytes twice

Browse Source 
Update pCopy->srcMemoryType and pCopy->dstMemoryType so
 that ihipGetMemcpyParam3DCommand() won't redetect memory
 type. This will prevent dstXInBytes and srcXInBytes getting
 updated twice with offset in hipMemoryTypeUnified type.

Change-Id: Ief4ea81941e25d22bcafd213b3b84a65eb23cf94


[ROCm/clr commit: 56daa6c489]
This commit is contained in:
taosang2
2023-06-29 19:01:52 -04:00
committed by Tao Sang
parent 6fa299f3a3
commit 828c7714e0
1 changed files with 4 additions and 0 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/clr/hipamd/src/hip_memory.cpp
+4
View File
@@ -2200,6 +2200,8 @@ hipError_t ihipMemcpyParam3D(const HIP_MEMCPY3D* pCopy, hipStream_t stream, bool
// {src/dst}Host may be unitialized. Copy over {src/dst}Device into it if we detect system memory.
const_cast<HIP_MEMCPY3D*>(pCopy)->srcHost = pCopy->srcDevice;
const_cast<HIP_MEMCPY3D*>(pCopy)->srcXInBytes += offset;
// We don't need detect memory type again for hipMemoryTypeUnified
const_cast<HIP_MEMCPY3D*>(pCopy)->srcMemoryType = srcMemoryType;
}
}
offset = 0;
@@ -2216,6 +2218,8 @@ hipError_t ihipMemcpyParam3D(const HIP_MEMCPY3D* pCopy, hipStream_t stream, bool
if (dstMemoryType == hipMemoryTypeHost) {
const_cast<HIP_MEMCPY3D*>(pCopy)->dstHost = pCopy->dstDevice;
const_cast<HIP_MEMCPY3D*>(pCopy)->dstXInBytes += offset;
// We don't need detect memory type again for hipMemoryTypeUnified
const_cast<HIP_MEMCPY3D*>(pCopy)->dstMemoryType = dstMemoryType;
}
}
// If {src/dst}MemoryType is hipMemoryTypeHost, check if the memory was prepinned.