AI/rocm-systems
2
0
Vork 0
Code Kwesties Pull-aanvragen Actions Paketten Projecten Publicaties Wiki Activiteit

Merge "(SWDEV-228488)" into amd-master-next

Bladeren bron
This commit is contained in:
German Andryeyev
2020-04-06 16:30:37 -04:00
gecommit door Gerrit Code Review
bovenliggende da1444bfc8 1bd640b659
commit 382d5ce77f
1 gewijzigde bestanden met toevoegingen van 12 en 19 verwijderingen
Download Patch-bestand Download Diff-bestand Expand all files Collapse all files
vdi/hip_memory.cpp
+12 -19
Bestand weergeven
@@ -884,19 +884,6 @@ hipError_t hipMemcpyDtoHAsync(void* dstHost,
HIP_RETURN(ihipMemcpy(dstHost, srcDevice, ByteCount, hipMemcpyDeviceToHost, *hip::getQueue(stream), true));
}
inline void adjustOrigin(amd::Coord3D &origin,
size_t offset,
size_t rowPitch,
size_t slicePitch) {
size_t zOffset = offset / (slicePitch ? slicePitch : 1);
size_t yOffset = (offset - slicePitch * zOffset) / (rowPitch ? rowPitch : 1);
size_t xOffset = (offset - slicePitch * zOffset - rowPitch * yOffset);
static_cast<size_t*>(origin)[0] += xOffset;
static_cast<size_t*>(origin)[1] += yOffset;
static_cast<size_t*>(origin)[2] += zOffset;
}
hipError_t ihipMemcpyAtoD(hipArray* srcArray,
void* dstDevice,
amd::Coord3D srcOrigin,
@@ -914,7 +901,6 @@ hipError_t ihipMemcpyAtoD(hipArray* srcArray,
amd::Image* srcImage = as_amd(srcMemObj)->asImage();
size_t dstOffset = 0;
amd::Memory* dstMemory = getMemoryObject(dstDevice, dstOffset);
adjustOrigin(dstOrigin, dstOffset, dstRowPitch, dstSlicePitch);
amd::BufferRect srcRect;
if (!srcRect.create(static_cast<size_t*>(srcOrigin), static_cast<size_t*>(copyRegion), srcImage->getRowPitch(), srcImage->getSlicePitch())) {
@@ -925,6 +911,8 @@ hipError_t ihipMemcpyAtoD(hipArray* srcArray,
if (!dstRect.create(static_cast<size_t*>(dstOrigin), static_cast<size_t*>(copyRegion), dstRowPitch, dstSlicePitch)) {
return hipErrorInvalidValue;
}
dstRect.start_ += dstOffset;
dstRect.end_ += dstOffset;
const size_t copySizeInBytes = copyRegion[0] * copyRegion[1] * copyRegion[2] * srcImage->getImageFormat().getElementSize();
if (!srcImage->validateRegion(srcOrigin, copyRegion) ||
@@ -972,13 +960,14 @@ hipError_t ihipMemcpyDtoA(void* srcDevice,
size_t srcOffset = 0;
amd::Memory* srcMemory = getMemoryObject(srcDevice, srcOffset);
adjustOrigin(srcOrigin, srcOffset, srcRowPitch, srcSlicePitch);
amd::Image* dstImage = as_amd(dstMemObj)->asImage();
amd::BufferRect srcRect;
if (!srcRect.create(static_cast<size_t*>(srcOrigin), static_cast<size_t*>(copyRegion), srcRowPitch, srcSlicePitch)) {
return hipErrorInvalidValue;
}
srcRect.start_ += srcOffset;
srcRect.end_ += srcOffset;
amd::BufferRect dstRect;
if (!dstRect.create(static_cast<size_t*>(dstOrigin), static_cast<size_t*>(copyRegion), dstImage->getRowPitch(), dstImage->getSlicePitch())) {
@@ -1028,15 +1017,15 @@ hipError_t ihipMemcpyDtoD(void* srcDevice,
bool isAsync = false) {
size_t srcOffset = 0;
amd::Memory *srcMemory = getMemoryObject(srcDevice, srcOffset);
adjustOrigin(srcOrigin, srcOffset, srcRowPitch, srcSlicePitch);
size_t dstOffset = 0;
amd::Memory *dstMemory = getMemoryObject(dstDevice, dstOffset);
adjustOrigin(dstOrigin, dstOffset, dstRowPitch, dstSlicePitch);
amd::BufferRect srcRect;
if (!srcRect.create(static_cast<size_t*>(srcOrigin), static_cast<size_t*>(copyRegion), srcRowPitch, srcSlicePitch)) {
return hipErrorInvalidValue;
}
srcRect.start_ += srcOffset;
srcRect.end_ += srcOffset;
amd::Coord3D srcStart(srcRect.start_, 0, 0);
amd::Coord3D srcEnd(srcRect.end_, 1, 1);
@@ -1048,6 +1037,8 @@ hipError_t ihipMemcpyDtoD(void* srcDevice,
if (!dstRect.create(static_cast<size_t*>(dstOrigin), static_cast<size_t*>(copyRegion), dstRowPitch, dstSlicePitch)) {
return hipErrorInvalidValue;
}
dstRect.start_ += dstOffset;
dstRect.end_ += dstOffset;
amd::Coord3D dstStart(dstRect.start_, 0, 0);
amd::Coord3D dstEnd(dstRect.end_, 1, 1);
@@ -1092,12 +1083,13 @@ hipError_t ihipMemcpyDtoH(void* srcDevice,
bool isAsync = false) {
size_t srcOffset = 0;
amd::Memory *srcMemory = getMemoryObject(srcDevice, srcOffset);
adjustOrigin(srcOrigin, srcOffset, srcRowPitch, srcSlicePitch);
amd::BufferRect srcRect;
if (!srcRect.create(static_cast<size_t*>(srcOrigin), static_cast<size_t*>(copyRegion), srcRowPitch, srcSlicePitch)) {
return hipErrorInvalidValue;
}
srcRect.start_ += srcOffset;
srcRect.end_ += srcOffset;
amd::Coord3D srcStart(srcRect.start_, 0, 0);
amd::Coord3D srcEnd(srcRect.end_, 1, 1);
@@ -1146,7 +1138,6 @@ hipError_t ihipMemcpyHtoD(const void* srcHost,
bool isAsync = false) {
size_t dstOffset = 0;
amd::Memory *dstMemory = getMemoryObject(dstDevice, dstOffset);
adjustOrigin(dstOrigin, dstOffset, dstRowPitch, dstSlicePitch);
amd::BufferRect srcRect;
if (!srcRect.create(static_cast<size_t*>(srcOrigin), static_cast<size_t*>(copyRegion), srcRowPitch, srcSlicePitch)) {
@@ -1157,6 +1148,8 @@ hipError_t ihipMemcpyHtoD(const void* srcHost,
if (!dstRect.create(static_cast<size_t*>(dstOrigin), static_cast<size_t*>(copyRegion), dstRowPitch, dstSlicePitch)) {
return hipErrorInvalidValue;
}
dstRect.start_ += dstOffset;
dstRect.end_ += dstOffset;
amd::Coord3D dstStart(dstRect.start_, 0, 0);
amd::Coord3D dstEnd(dstRect.end_, 1, 1);