From e074cb2f9b86d048d8423112410223ae2be6aef7 Mon Sep 17 00:00:00 2001 From: Aryan Salmanpour Date: Thu, 20 Feb 2025 20:33:54 -0500 Subject: [PATCH] Fix a bug that prevented copying the decoded image into the output buffer when the output buffer is larger than the input image (#122) * Fix a bug that prevented copying the decoded image into the output buffer when the output buffer is larger than the input image * add check for a valid roi_width --- CHANGELOG.md | 6 ++++ api/rocjpeg_version.h | 2 +- src/rocjpeg_decoder.cpp | 67 ++++++++++++++++++++++++++++++++++++++++- 3 files changed, 73 insertions(+), 2 deletions(-) diff --git a/CHANGELOG.md b/CHANGELOG.md index 7cef113eef..e9f295815f 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -3,6 +3,12 @@ Documentation for rocJPEG is available at [https://rocm.docs.amd.com/projects/rocJPEG/en/latest/](https://rocm.docs.amd.com/projects/rocJPEG/en/latest/) +## rocJPEG 0.9.0 (Unreleased) + +### Resolved issues + +* Fixed a bug that prevented copying the decoded image into the output buffer when the output buffer is larger than the input image. + ## rocJPEG 0.8.0 for ROCm 6.4 ### Changed diff --git a/api/rocjpeg_version.h b/api/rocjpeg_version.h index 781cdc69e0..bb627246ae 100644 --- a/api/rocjpeg_version.h +++ b/api/rocjpeg_version.h @@ -34,7 +34,7 @@ THE SOFTWARE. extern "C" { #endif #define ROCJPEG_MAJOR_VERSION 0 -#define ROCJPEG_MINOR_VERSION 8 +#define ROCJPEG_MINOR_VERSION 9 #define ROCJPEG_MICRO_VERSION 0 diff --git a/src/rocjpeg_decoder.cpp b/src/rocjpeg_decoder.cpp index f764261402..6f6b7c62bd 100644 --- a/src/rocjpeg_decoder.cpp +++ b/src/rocjpeg_decoder.cpp @@ -388,12 +388,77 @@ RocJpegStatus RocJpegDecoder::CopyChannel(HipInteropDeviceMem& hip_interop_dev_m } roi_offset = top * hip_interop_dev_mem.pitch[channel_index] + left; } + + uint32_t channel_widths[ROCJPEG_MAX_COMPONENT] = {}; + uint32_t roi_width = decode_params->crop_rectangle.right - decode_params->crop_rectangle.left; + bool is_roi_width_valid = roi_width > 0 && roi_width <= hip_interop_dev_mem.width; + switch (decode_params->output_format) { + case ROCJPEG_OUTPUT_NATIVE: + switch (hip_interop_dev_mem.surface_format) { + case VA_FOURCC_444P: + channel_widths[2] = channel_widths[1] = channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + break; + case VA_FOURCC_422V: + channel_widths[2] = channel_widths[1] = channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + break; + case VA_FOURCC_YUY2: + channel_widths[0] = (is_roi_width_valid ? roi_width : hip_interop_dev_mem.width) * 2; + break; + case VA_FOURCC_NV12: + channel_widths[1] = channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + break; + case VA_FOURCC_Y800: + channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + break; + default: + ERR("Unknown output format!"); + return ROCJPEG_STATUS_INVALID_PARAMETER; + } + break; + case ROCJPEG_OUTPUT_YUV_PLANAR: + switch (hip_interop_dev_mem.surface_format) { + case VA_FOURCC_444P: + channel_widths[2] = channel_widths[1] = channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + break; + case VA_FOURCC_422V: + channel_widths[2] = channel_widths[1] = channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + break; + case VA_FOURCC_YUY2: + channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + channel_widths[2] = channel_widths[1] = channel_widths[0] >> 1; + break; + case VA_FOURCC_NV12: + channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + channel_widths[2] = channel_widths[1] = channel_widths[0] >> 1; + break; + case VA_FOURCC_Y800: + channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + break; + default: + ERR("Unknown output format!"); + return ROCJPEG_STATUS_INVALID_PARAMETER; + } + break; + case ROCJPEG_OUTPUT_Y: + channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + break; + case ROCJPEG_OUTPUT_RGB: + channel_widths[0] = (is_roi_width_valid ? roi_width : hip_interop_dev_mem.width) * 3; + break; + case ROCJPEG_OUTPUT_RGB_PLANAR: + channel_widths[2] = channel_widths[1] = channel_widths[0] = is_roi_width_valid ? roi_width : hip_interop_dev_mem.width; + break; + default: + ERR("Unknown output format!"); + return ROCJPEG_STATUS_INVALID_PARAMETER; + } + if (destination->pitch[channel_index] == hip_interop_dev_mem.pitch[channel_index]) { uint32_t channel_size = destination->pitch[channel_index] * channel_height; CHECK_HIP(hipMemcpyDtoDAsync(destination->channel[channel_index], hip_interop_dev_mem.hip_mapped_device_mem + hip_interop_dev_mem.offset[channel_index] + roi_offset, channel_size, hip_stream_)); } else { CHECK_HIP(hipMemcpy2DAsync(destination->channel[channel_index], destination->pitch[channel_index], hip_interop_dev_mem.hip_mapped_device_mem + hip_interop_dev_mem.offset[channel_index] + roi_offset, hip_interop_dev_mem.pitch[channel_index], - destination->pitch[channel_index], channel_height, hipMemcpyDeviceToDevice, hip_stream_)); + channel_widths[channel_index], channel_height, hipMemcpyDeviceToDevice, hip_stream_)); } } return ROCJPEG_STATUS_SUCCESS;