SWDEV-412993 - As origin is passed with start, region should be end.

Change-Id: I7077484d4d8c302fb2f2714cb68c15ba5173eae3


[ROCm/clr commit: a28a5a9a38]

projects/clr/hipamd/src/hip_memory.cpp

@@ -1645,7 +1645,7 @@ hipError_t ihipMemcpyDtoDValidate(void* srcDevice, void* dstDevice, amd::Coord3D
   srcRect.end_ += srcOffset;
 
   amd::Coord3D srcStart(srcRect.start_, 0, 0);
-  amd::Coord3D srcSize(srcRect.end_ - srcRect.start_, 1, 1);
+  amd::Coord3D srcSize(srcRect.end_, 1, 1);
   if (!srcMemory->validateRegion(srcStart, srcSize)) {
     return hipErrorInvalidValue;
   }
@@ -1658,7 +1658,7 @@ hipError_t ihipMemcpyDtoDValidate(void* srcDevice, void* dstDevice, amd::Coord3D
   dstRect.end_ += dstOffset;
 
   amd::Coord3D dstStart(dstRect.start_, 0, 0);
-  amd::Coord3D dstSize(dstRect.end_ - dstRect.start_, 1, 1);
+  amd::Coord3D dstSize(dstRect.end_, 1, 1);
   if (!dstMemory->validateRegion(dstStart, dstSize)) {
     return hipErrorInvalidValue;
   }
@@ -1718,7 +1718,7 @@ hipError_t ihipMemcpyDtoHValidate(void* srcDevice, void* dstHost, amd::Coord3D&
   srcRect.end_ += srcOffset;
 
   amd::Coord3D srcStart(srcRect.start_, 0, 0);
-  amd::Coord3D srcSize(srcRect.end_ - srcRect.start_, 1, 1);
+  amd::Coord3D srcSize(srcRect.end_, 1, 1);
   if (!srcMemory->validateRegion(srcStart, srcSize)) {
     return hipErrorInvalidValue;
   }
@@ -1788,7 +1788,7 @@ hipError_t ihipMemcpyHtoDValidate(const void* srcHost, void* dstDevice, amd::Coo
   dstRect.end_ += dstOffset;
 
   amd::Coord3D dstStart(dstRect.start_, 0, 0);
-  amd::Coord3D dstSize(dstRect.end_ - dstRect.start_, 1, 1);
+  amd::Coord3D dstSize(dstRect.end_, 1, 1);
   if (!dstMemory->validateRegion(dstStart, dstSize)) {
     return hipErrorInvalidValue;
   }