(SWDEV-228782)

The only requirment from hipMallocPitch() is that the returned pitch is aligned to the HW image pitch alignment. There is no restriction on the size of the allocation, since the memory might not be used for images.

Change-Id: I97438e5fe4012ca4721b14b85f514dbac803c17c


[ROCm/clr commit: 111897dbea]

projects/clr/hipamd/vdi/hip_memory.cpp

@@ -358,17 +358,13 @@ hipError_t ihipMallocPitch(void** ptr, size_t* pitch, size_t width, size_t heigh
 
   amd::Device* device = hip::getCurrentDevice()->devices()[0];
 
-  if ((width == 0) || (height == 0)) {
-    *ptr = nullptr;
-    return hipSuccess;
-  }
-  else if (!(device->info().image2DMaxWidth_ >= width &&
-           device->info().image2DMaxHeight_ >= height ) || (ptr == nullptr)) {
+  if (ptr == nullptr) {
     return hipErrorInvalidValue;
   }
 
-  if (device->info().maxMemAllocSize_ < (width * height)) {
-    return hipErrorOutOfMemory;
+  if ((width == 0) || (height == 0)) {
+    *ptr = nullptr;
+    return hipSuccess;
   }
 
   const amd::Image::Format imageFormat(*image_format);
@@ -376,6 +372,11 @@ hipError_t ihipMallocPitch(void** ptr, size_t* pitch, size_t width, size_t heigh
   *pitch = amd::alignUp(width * imageFormat.getElementSize(), device->info().imagePitchAlignment_);
 
   size_t sizeBytes = *pitch * height * depth;
+
+  if (device->info().maxMemAllocSize_ < sizeBytes) {
+    return hipErrorOutOfMemory;
+  }
+
   *ptr = amd::SvmBuffer::malloc(*hip::getCurrentDevice()->asContext(), 0, sizeBytes,
                                 device->info().memBaseAddrAlign_);