[ROCM] Add rocjpeg support for GPU image decoding

[xytpai]

Adds a rocJPEG-backed GPU JPEG decoding path for ROCm builds, gated by TORCHVISION_USE_ROCJPEG / ROCJPEG_FOUND, alongside the existing nvJPEG CUDA path.

This enables torchvision.io.decode_jpeg(..., device="cuda") on ROCm for supported JPEGs, including RGB, GRAY, and UNCHANGED modes. The implementation uses rocJPEG's hardware backend, handles rocJPEG errors through TORCH_CHECK, avoids process termination on decode failures, and returns contiguous output tensors for parity with nvJPEG.

If you meet any libva related compile errors:

sudo apt install vainfo libva2 libva-drm2 libdrm2 mesa-va-drivers libva-dev

Note

rocJPEG currently uses ROCJPEG_BACKEND_HARDWARE, which has stricter capability limits than the CPU/nvJPEG path: it rejects images smaller than 64x64 and unsupported chroma subsampling such as 4:1:1/unknown. If rocJPEG hardware doesn't support the input, fallback to CPU decode and copy result to CUDA.

cc @jeffdaily @jithunnair-amd

[pytorch-bot]

🔗 Helpful Links

🧪 See artifacts and rendered test results at hud.pytorch.org/pr/pytorch/vision/9342

• 📄 Preview Python docs built from this PR

Note: Links to docs will display an error until the docs builds have been completed.

❗ 1 Active SEVs

There are 1 currently active SEVs. If your PR is affected, please view them below:

• [Housekeeping] Remove pypi-cache on ue1 cluster

This comment was automatically generated by Dr. CI and updates every 15 minutes.

[meta-cla]

Hi @xytpai!

Thank you for your pull request and welcome to our community.

Action Required

In order to merge any pull request (code, docs, etc.), we require contributors to sign our Contributor License Agreement, and we don't seem to have one on file for you.

Process

In order for us to review and merge your suggested changes, please sign at https://code.facebook.com/cla. If you are contributing on behalf of someone else (eg your employer), the individual CLA may not be sufficient and your employer may need to sign the corporate CLA.

Once the CLA is signed, our tooling will perform checks and validations. Afterwards, the pull request will be tagged with CLA signed. The tagging process may take up to 1 hour after signing. Please give it that time before contacting us about it.

If you have received this in error or have any questions, please contact us at cla@meta.com. Thanks!

[xytpai]

@NicolasHug

[NicolasHug]

Hi @xytpai , thanks for the PR.
I don't think this is something we'll have the bandwidth to prioritize at this moment, although this might eventually become in scope for TorchCodec, e.g. via the "third-party TC extensions" that we plan to support (where decoder extensions could be added in third-parties repos, out of TC's core).

May I ask, is there already a Python library that exposes the Rocm jpeg decoder?

[xytpai]

Hi, @NicolasHug thanks for the feedback!

I understand that this may not be a priority at the moment. I'd like to ask if we could consider merging this in its current form to reach a functionally ready state. This would allow the ROCm JPEG decoding path to be available for users and contributors to build upon. We're happy to address any remaining review comments and ensure the code meets the project's quality standards. Please let us know if there's anything we can do to move this forward. Thanks!

May I ask, is there already a Python library that exposes the Rocm jpeg decoder?

Yes, there is already a Python library that exposes the ROCm JPEG decoder. https://github.com/ROCm/rocPyDecode

[xytpai]

@NicolasHug

[jeffdaily]

Review — [ROCM] Add rocjpeg support for GPU image decoding (RGB output)

Summary

Adds a rocJPEG-backed decode_jpegs_cuda (RGB-only) alongside the existing nvJPEG path, gated by a new ROCJPEG_FOUND macro and TORCHVISION_USE_ROCJPEG build flag. The implementation is largely an AMD rocJPEG sample grafted onto the nvJPEG wrapper, and several pieces of the nvJPEG scaffolding were copied without being wired up. There are correctness and robustness issues that block approval, plus no test coverage for the reduced-capability backend.

Code Quality

• CHECK_ROCJPEG / CHECK_HIP call exit(1) on failure (decode_jpegs_cuda.h, new macros). A library must never terminate the host process. Any malformed/unsupported JPEG, OOM, or transient HIP error would kill the user's entire Python interpreter with no traceback and no chance to try/except. The nvJPEG path uses TORCH_CHECK, which throws a catchable c10::Error. These macros must use TORCH_CHECK(status == ROCJPEG_STATUS_SUCCESS, ...) / TORCH_CHECK(hip_status == hipSuccess, ...) instead of std::cerr/std::cout + exit(1). This is the most serious issue.

• Stray std::cout diagnostics remain. getChannelPitchAndSizes prints "Unknown chroma subsampling!" and "Unknown output format!" to stdout before returning EXIT_FAILURE. A "rm cout" commit is in the history but missed these. Replace with TORCH_CHECK(false, ...) so the message surfaces as an exception rather than polluting stdout.

• Dead code / unused locals in decode_images:

  • getChromaSubsamplingStr(...) fills chroma_sub_sampling, which is never read.
  • num_components (from rocJpegGetImageInfo) is never read.
  • std::vector<int> channels(num_images) and channels[j] = num_channels are never used afterward.
  • prior_channel_sizes is declared and sized but never used.
  • The commented-out // if (current_batch_size == 2) { should be removed, not left in.

  Given only ROCJPEG_OUTPUT_RGB_PLANAR is reachable (the dispatcher rejects every other mode), the entire multi-format switch in getChannelPitchAndSizes and the channel_sizes machinery is dead for this PR. Either reduce it to the RGB-planar case or add a comment that it is staged for future formats — as written it is a large amount of untested, unreachable code.

• Copy-pasted nvJPEG error string: "The input tensor must be on CPU when decoding with nvjpeg" should say rocJPEG. Same for the decode_images doc comment, which lists output_format ... ROCJPEG_OUTPUT_RGB and a device argument the function does not take.

Correctness / Thread Safety

• The stream/event synchronization is copy-pasted from nvJPEG but never actually connected to rocJPEG. In the nvJPEG path, nvjpegDecodeBatched(..., stream) enqueues work on decoder->stream, so event.record(stream); event.block(current_stream) and the cudaStreamSynchronize(stream) calls are meaningful. Here, rocJpegDecodeBatched takes no stream argument and runs on rocJPEG's own internal stream; the decoder's stream member is never passed to any rocJPEG call. Consequently:

  • Both cudaStreamSynchronize(stream) calls in decode_images synchronize an unrelated, idle pool stream — no-ops.
  • event.record(rocJpegDecoder->stream) records on that idle stream, so event.block(current_stream) blocks on nothing.

  This is only safe if rocJpegDecodeBatched is fully host-synchronous. If it is, all of this scaffolding is dead and misleading and should be removed (or replaced with a single hipStreamSynchronize on rocJPEG's actual stream). If it is not host-synchronous, then the decoded tensors are returned to the caller's current stream with no real dependency edge, which is a data race / silent corruption. Please confirm rocJPEG's completion semantics and either remove the ineffective scaffolding or add the correct synchronization.

Testing

• No tests added, and the existing GPU test will fail on ROCm. test_decode_jpegs_cuda (test/test_image.py) is gated only by @needs_cuda, which is true under ROCm (torch.cuda.is_available()), and it parametrizes over mode in {UNCHANGED, GRAY, RGB}. This backend only supports RGB; GRAY/UNCHANGED hit TORCH_CHECK(false, "mode is not supported for ROCJPEG decoding on GPU") and the test errors. The test also decodes every .jpg asset and compares against CPU — any asset smaller than 64x64 or using 4:1:1 subsampling will trip the new TORCH_CHECKs ("not supported by VCN Hardware"). The PR must either restrict/parametrize these tests for the rocJPEG capability set or add a dedicated rocJPEG test path. Shipping a backend that breaks the existing CUDA test suite on ROCm is a blocking gap.

API Design / Backward Compatibility

• Silent capability regression vs. the nvJPEG/CPU path. On CUDA, decode_jpeg(..., device='cuda') supports GRAY/UNCHANGED and arbitrary image sizes (falling back to a non-hardware backend when needed). The rocJPEG path uses ROCJPEG_BACKEND_HARDWARE only, with no software fallback, and hard-errors on images <64px in either dimension and on 4:1:1/unknown subsampling. The same user code that works on CUDA/CPU will raise on ROCm for valid JPEGs. At minimum document this; ideally fall back to a non-hardware backend or to CPU decode for unsupported inputs.

• Returned tensors are non-contiguous strided views, unlike nvJPEG. The output tensor is allocated at 16-aligned height/width and then narrow-ed back to the true dimensions, so for widths not divisible by 16 the returned tensor has a row stride > width. The nvJPEG path returns exact-size contiguous tensors. Downstream code or tests that assume .is_contiguous() will behave differently across backends. Consider returning .contiguous() for parity.

setup.py

• USE_ROCJPEG defaults to "1", so it is enabled on every build including NVIDIA/CPU. The nvjpeg_found and USE_NVJPEG branch is checked first so an NVIDIA box still builds nvJPEG, and rocjpeg_found requires ROCM_HOME + the header, so this is functionally fine — but the combined warning "Building torchvision without NVJPEG or ROCJPEG support" will now fire on plain CUDA builds where only nvJPEG was ever expected. Minor, but worth confirming the messaging is intentional.

Recommendation

Request Changes. Blocking: (1) exit(1) in the CHECK_* macros must become TORCH_CHECK; (2) the stream/event synchronization is ineffective as wired — confirm rocJPEG sync semantics and either remove it or synchronize correctly; (3) the existing CUDA test suite breaks on ROCm and no rocJPEG-appropriate tests were added. The dead code, stray std::cout, copy-pasted error strings, and the documented capability/contiguity differences should also be addressed before merge.

🤖 Generated with Claude Code

[rrawther]

@xytpai: I have added some review comments for this PR. Also, please address the comments from @jeffdaily too.

[pytorch-bot]

The following ciflow label(s) have been added but CI has not been triggered yet because the workflows are awaiting approval:

• ciflow/rocm

Once a maintainer approves the workflows (scroll to the bottom of the PR page), the corresponding CI jobs will be triggered automatically. Please ping one of the reviewers if you do not have access to approve and run workflows.

[NicolasHug]

Also what's the testing situation here? We don't have Rocm test runners in torchvision I believe. Is AMD testing these out-of-core somehow?

[xytpai]

Also what's the testing situation here? We don't have Rocm test runners in torchvision I believe. Is AMD testing these out-of-core somehow?

Yes, torchvision CI does not currently have ROCm image decode runners, so this path is not covered by upstream CI yet.

On the AMD side, we are validating this out-of-core on ROCm hardware with the existing test_decode_jpegs_cuda coverage, including UNCHANGED, GRAY, and RGB modes, and comparing against the CPU decode outputs.

[NicolasHug]

yes, I think I'd prefer having the ROCm code in a ROCm specific file. The common decode_jpegs_cuda() can be moved to a common file too.

[NicolasHug]

Can we just leave the previous if USE_NVJPEG and (torch.cuda.is_available() or FORCE_CUDA): block exactly like it was, and just have a separate (indepentent) ROCm-specific block below it? They should be mutually exclusive?

[xytpai]

I don't think we can rely on USE_NVJPEG and USE_ROCJPEG being mutually exclusive because both default to true. On ROCm builds, the nvJPEG block would still run and warn unless it is guarded by not IS_ROCM. I think the mutual exclusion should be based on the backend (not IS_ROCM for nvJPEG, IS_ROCM for rocJPEG), not on the two USE flags.

[NicolasHug]

Thanks for the PR @xytpai , this LGTM so far. Just giving you a heads up, before moving forward I'd like to:

• get a small ROCm unit-test CI job within the torchvision repo so that this can be tested here. I opened Add unittest rocm job #9527 for that, I'm currently waiting on getting permission.
• port the CUDA jpeg decoder to the stable ABI. We started some effort to port the whole torchvision codebase to rely on the torch stable ABI (Stand up dual stable-ABI build harness and port nms as the first op. #9524) - I'd like to port the CUDA jpeg decoder next so that we can the update this ROCm decoder PR to directly rely on the stable ABI (rather than having to port it later) - CC @adabeyta

[adabeyta]

Thanks for the PR @xytpai , this LGTM so far. Just giving you a heads up, before moving forward I'd like to:

• get a small ROCm unit-test CI job within the torchvision repo so that this can be tested here. I opened Add unittest rocm job #9527 for that, I'm currently waiting on getting permission.
• port the CUDA jpeg decoder to the stable ABI. We started some effort to port the whole torchvision codebase to rely on the torch stable ABI (Stand up dual stable-ABI build harness and port nms as the first op. #9524) - I'd like to port the CUDA jpeg decoder next so that we can the update this ROCm decoder PR to directly rely on the stable ABI (rather than having to port it later) - CC @adabeyta

The CUDA jpeg decoder port to stable-ABI is now in #9533

[NicolasHug]

@xytpai do you mind updating this PR so that it also compiles into the image_stable extension, and relies on the stable ABI?

Regarding testing, I added a ROCm CI job but unfortunately it won't be triggered on PRs that come from a fork like yours. So I created #9529 which should allow us to see the ROCm test results. I'll have to update it with your latest changes each time (on top of trigger the CI here in this PR).

[xytpai]

@xytpai do you mind updating this PR so that it also compiles into the image_stable extension, and relies on the stable ABI?

Regarding testing, I added a ROCm CI job but unfortunately it won't be triggered on PRs that come from a fork like yours. So I created #9529 which should allow us to see the ROCm test results. I'll have to update it with your latest changes each time (on top of trigger the CI here in this PR).

Thanks! I’ve updated the PR so the ROCm JPEG decoder now builds into the image_stable extension and uses the stable ABI path. Please take another look when you get a chance.