From 37bb58b556948d62feb5ad863b366dac2f49cf01 Mon Sep 17 00:00:00 2001 From: foreman Date: Mon, 20 Jan 2020 19:05:16 -0500 Subject: [PATCH] P4 to Git Change 2060446 by vsytchen@vsytchen-ocl-win10-2 on 2020/01/20 18:58:37 SWDEV-215533 - [HIP][Windows]Output mismatch with tex3D() When calculating the pitch for hipMallocPitch() align it to the device's image pitch alignment. If we don't, PAL backend will overalign the pitch when creating an image buffer. ReviewBoardURL = http://ocltc.amd.com/reviews/r/18496/diff/ Affected files ... ... //depot/stg/opencl/drivers/opencl/api/hip/hip_memory.cpp#96 edit --- hipamd/api/hip/hip_memory.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/hipamd/api/hip/hip_memory.cpp b/hipamd/api/hip/hip_memory.cpp index 6b23913386..9cd5034eeb 100644 --- a/hipamd/api/hip/hip_memory.cpp +++ b/hipamd/api/hip/hip_memory.cpp @@ -340,7 +340,7 @@ hipError_t ihipMallocPitch(void** ptr, size_t* pitch, size_t width, size_t heigh const amd::Image::Format imageFormat(*image_format); - *pitch = width * imageFormat.getElementSize(); + *pitch = amd::alignUp(width * imageFormat.getElementSize(), device->info().imagePitchAlignment_); size_t sizeBytes = *pitch * height * depth; *ptr = amd::SvmBuffer::malloc(*hip::getCurrentContext(), 0, sizeBytes,