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

SWDEV-275317 - Correctly pass HIP image width to ROCclr APIs in AtoD/DtoA copies

Browse Source 
HIP assumes that image width is in bytes, but OCL/ROCclr assumes that
it's in pixels. AtoD/DtoA need to account for this.

Change-Id: I275bd41d8b03e141caaf951bc6b714e51ca72dfc


[ROCm/hip commit: e967adbdaf]
This commit is contained in:
Vladislav Sytchenko
2021-03-15 21:09:58 -04:00
parent 77d1508b7d
commit bde082af6e
1 changed files with 8 additions and 0 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/hip/rocclr/hip_memory.cpp
+8
View File
@@ -971,6 +971,10 @@ hipError_t ihipMemcpyAtoD(hipArray* srcArray,
}
amd::Image* srcImage = as_amd(srcMemObj)->asImage();
// HIP assumes the width is in bytes, but OCL assumes it's in pixels.
const size_t elementSize = srcImage->getImageFormat().getElementSize();
static_cast<size_t*>(srcOrigin)[0] /= elementSize;
static_cast<size_t*>(copyRegion)[0] /= elementSize;
amd::BufferRect srcRect;
if (!srcRect.create(static_cast<size_t*>(srcOrigin), static_cast<size_t*>(copyRegion), srcImage->getRowPitch(), srcImage->getSlicePitch())) {
@@ -1031,6 +1035,10 @@ hipError_t ihipMemcpyDtoA(void* srcDevice,
}
amd::Image* dstImage = as_amd(dstMemObj)->asImage();
// HIP assumes the width is in bytes, but OCL assumes it's in pixels.
const size_t elementSize = dstImage->getImageFormat().getElementSize();
static_cast<size_t*>(dstOrigin)[0] /= elementSize;
static_cast<size_t*>(copyRegion)[0] /= elementSize;
amd::BufferRect srcRect;
if (!srcRect.create(static_cast<size_t*>(srcOrigin), static_cast<size_t*>(copyRegion), srcRowPitch, srcSlicePitch)) {