diff --git a/projects/hip/vdi/hip_conversions.hpp b/projects/hip/vdi/hip_conversions.hpp index 6809765eb0..ff88ae4828 100644 --- a/projects/hip/vdi/hip_conversions.hpp +++ b/projects/hip/vdi/hip_conversions.hpp @@ -618,7 +618,7 @@ std::pair getMemoryType(const hipMemcpyKind kind) } inline -_HIP_MEMCPY3D getMemcpy3DParms(const hip_Memcpy2D& desc2D) { +_HIP_MEMCPY3D getDrvMemcpy3DDesc(const hip_Memcpy2D& desc2D) { _HIP_MEMCPY3D desc3D = {}; desc3D.srcXInBytes = desc2D.srcXInBytes; @@ -649,4 +649,64 @@ _HIP_MEMCPY3D getMemcpy3DParms(const hip_Memcpy2D& desc2D) { return desc3D; } + +inline +_HIP_MEMCPY3D getDrvMemcpy3DDesc(const hipMemcpy3DParms& desc) { + _HIP_MEMCPY3D descDrv = {}; + + descDrv.WidthInBytes = desc.extent.width; + descDrv.Height = desc.extent.height; + descDrv.Depth = desc.extent.depth; + + descDrv.srcXInBytes = desc.srcPos.x; + descDrv.srcY = desc.srcPos.y; + descDrv.srcZ = desc.srcPos.z; + descDrv.srcLOD = 0; + + descDrv.dstXInBytes = desc.dstPos.x; + descDrv.dstY = desc.dstPos.y; + descDrv.dstZ = desc.dstPos.z; + descDrv.dstLOD = 0; + + if (desc.srcArray != nullptr) { + descDrv.srcMemoryType = hipMemoryTypeArray; + descDrv.srcArray = desc.srcArray; + // When reffering to array memory, hipPos::x is in elements. + descDrv.srcXInBytes *= getElementSize(desc.srcArray->Format); + } + + if (desc.srcPtr.ptr != nullptr) { + descDrv.srcMemoryType = std::get<0>(hip::getMemoryType(desc.kind)); + descDrv.srcHost = desc.srcPtr.ptr; + descDrv.srcDevice = desc.srcPtr.ptr; + descDrv.srcPitch = desc.srcPtr.pitch; + descDrv.srcHeight = desc.srcPtr.ysize; + } + + if (desc.dstArray != nullptr) { + descDrv.dstMemoryType = hipMemoryTypeArray; + descDrv.dstArray = desc.dstArray; + // When reffering to array memory, hipPos::x is in elements. + descDrv.dstXInBytes *= getElementSize(desc.dstArray->Format); + } + + if (desc.dstPtr.ptr != nullptr) { + descDrv.dstMemoryType = std::get<1>(getMemoryType(desc.kind)); + descDrv.dstHost = desc.dstPtr.ptr; + descDrv.dstDevice = desc.dstPtr.ptr; + descDrv.dstPitch = desc.dstPtr.pitch; + descDrv.dstHeight = desc.dstPtr.ysize; + } + + // If a HIP array is participating in the copy, the extent is defined in terms of that array's elements. + if ((desc.srcArray != nullptr) && (desc.dstArray == nullptr)) { + descDrv.WidthInBytes *= getElementSize(desc.srcArray->Format); + } else if ((desc.srcArray == nullptr) && (desc.dstArray != nullptr)) { + descDrv.WidthInBytes *= getElementSize(desc.dstArray->Format); + } else if ((desc.srcArray != nullptr) && (desc.dstArray != nullptr)) { + descDrv.WidthInBytes *= getElementSize(desc.dstArray->Format); + } + + return descDrv; +} }; diff --git a/projects/hip/vdi/hip_memory.cpp b/projects/hip/vdi/hip_memory.cpp index cc3908ca64..84f8fa9def 100644 --- a/projects/hip/vdi/hip_memory.cpp +++ b/projects/hip/vdi/hip_memory.cpp @@ -1387,7 +1387,7 @@ hipError_t ihipMemcpyParam3D(const _HIP_MEMCPY3D* pCopy, hipError_t ihipMemcpyParam2D(const hip_Memcpy2D* pCopy, hipStream_t stream, bool isAsync = false) { - _HIP_MEMCPY3D desc = hip::getMemcpy3DParms(*pCopy); + _HIP_MEMCPY3D desc = hip::getDrvMemcpy3DDesc(*pCopy); return ihipMemcpyParam3D(&desc, stream, isAsync); } @@ -1558,62 +1558,9 @@ hipError_t ihipMemcpy3D(const hipMemcpy3DParms* p, return hipErrorInvalidValue; } - _HIP_MEMCPY3D pCopy = {}; + const _HIP_MEMCPY3D desc = hip::getDrvMemcpy3DDesc(*p); - pCopy.WidthInBytes = p->extent.width; - pCopy.Height = p->extent.height; - pCopy.Depth = p->extent.depth; - - pCopy.srcXInBytes = p->srcPos.x; - pCopy.srcY = p->srcPos.y; - pCopy.srcZ = p->srcPos.z; - pCopy.srcLOD = 0; - - pCopy.dstXInBytes = p->dstPos.x; - pCopy.dstY = p->dstPos.y; - pCopy.dstZ = p->dstPos.z; - pCopy.dstLOD = 0; - - if (p->srcArray != nullptr) { - pCopy.srcMemoryType = hipMemoryTypeArray; - pCopy.srcArray = p->srcArray; - // When reffering to array memory, hipPos::x is in elements. - pCopy.srcXInBytes *= hip::getElementSize(p->srcArray->Format); - } - - if (p->srcPtr.ptr != nullptr) { - pCopy.srcMemoryType = std::get<0>(hip::getMemoryType(p->kind)); - pCopy.srcHost = p->srcPtr.ptr; - pCopy.srcDevice = p->srcPtr.ptr; - pCopy.srcPitch = p->srcPtr.pitch; - pCopy.srcHeight = p->srcPtr.ysize; - } - - if (p->dstArray != nullptr) { - pCopy.dstMemoryType = hipMemoryTypeArray; - pCopy.dstArray = p->dstArray; - // When reffering to array memory, hipPos::x is in elements. - pCopy.dstXInBytes *= hip::getElementSize(p->dstArray->Format); - } - - if (p->dstPtr.ptr != nullptr) { - pCopy.dstMemoryType = std::get<1>(hip::getMemoryType(p->kind)); - pCopy.dstHost = p->dstPtr.ptr; - pCopy.dstDevice = p->dstPtr.ptr; - pCopy.dstPitch = p->dstPtr.pitch; - pCopy.dstHeight = p->dstPtr.ysize; - } - - // If a HIP array is participating in the copy, the extent is defined in terms of that array's elements. - if ((p->srcArray != nullptr) && (p->dstArray == nullptr)) { - pCopy.WidthInBytes *= hip::getElementSize(p->srcArray->Format); - } else if ((p->srcArray == nullptr) && (p->dstArray != nullptr)) { - pCopy.WidthInBytes *= hip::getElementSize(p->dstArray->Format); - } else if ((p->srcArray != nullptr) && (p->dstArray != nullptr)) { - pCopy.WidthInBytes *= hip::getElementSize(p->dstArray->Format); - } - - return ihipMemcpyParam3D(&pCopy, stream, isAsync); + return ihipMemcpyParam3D(&desc, stream, isAsync); } hipError_t hipMemcpy3D(const hipMemcpy3DParms* p) {