Add AMD GPU support via ROCm/HIP#94
Open
jeffdaily wants to merge 2 commits into
Open
Conversation
This adds an optional AMD GPU build to cuPDLPx through ROCm/HIP, alongside the existing CUDA path. The CUDA build is unchanged when USE_HIP is off. To review: start with internal/cuda_to_hip.h, which routes the CUDA runtime, cuBLAS, cuSPARSE, and CUB symbols used by the solver to their hipRT, hipBLAS, hipSPARSE, and hipCUB equivalents on a HIP build, and includes the standard CUDA headers otherwise. The device sources keep their CUDA spelling and are compiled as HIP. internal/cusparse_compat.h selects the standard hipsparseSpMV path on ROCm, since hipSPARSE does not provide the cusparseSpMVOp variant. CMakeLists.txt gains a USE_HIP option (off by default). When enabled the project is configured with the HIP language, the .cu sources are compiled as HIP, and the targets link hipBLAS, hipSPARSE, and hipCUB instead of the CUDA libraries. GPU architectures are chosen with CMAKE_HIP_ARCHITECTURES, defaulting to gfx90a. On Windows the CLI-only mps_parser.c is excluded from the core library because it relies on strtok_r. The interface test gains a case that runs the GPU solver path with presolve disabled, exercising the hipBLAS and hipSPARSE execution path end to end. Test Plan: Built and ran on an AMD Instinct MI200 (gfx90a) with ROCm 7.2.1: ``` cmake -B build -DUSE_HIP=ON -DCMAKE_HIP_ARCHITECTURES=gfx90a -DCMAKE_PREFIX_PATH=/opt/rocm \ -DCUPDLPX_BUILD_CLI=ON -DCUPDLPX_BUILD_TESTS=ON -DCMAKE_BUILD_TYPE=Release cmake --build build -j$(nproc) ./build/tests/test_interface ``` The interface suite passes, including the GPU solver case (Status: OPTIMAL). The same configuration builds cleanly for gfx1100 (RDNA3) and gfx1201 (RDNA4); the device code objects are identical across the documentation and formatting commits that followed validation. The CUDA build path is unaffected by these changes. This work was authored with the assistance of Claude, an AI assistant by Anthropic.
The ROCm support commit routed all CUDA/HIP includes through internal/cuda_to_hip.h and pulled it into utils.h and internal_types.h, which are included by the C translation units (cli.c, cupdlpx.c, mps_parser.c, presolve.c). On the CUDA path that header included <cub/device/device_reduce.cuh> unconditionally; cub is C++ only, so the C compiler failed with "unknown type name 'namespace'", breaking every CUDA build job (all Linux and Windows toolchains, CUDA 12.4 through 13.1). The HIP path was unaffected because its hipcub include was already guarded with #ifdef __cplusplus. The fix mirrors that guard on the CUDA branch: the cub header is only included for C++ translation units (the .cu device sources that actually use cub::DeviceReduce). The change is entirely within the #else CUDA branch, so the HIP/ROCm device code is unchanged. Authored with assistance from Claude. Test Plan: reproduced and verified the CUDA path locally with the CUDA 12.8 toolkit (gcc 13, ninja), matching the upstream CI configure: ``` cmake -B build -G Ninja -DCMAKE_BUILD_TYPE=Release \ -DCUPDLPX_BUILD_TESTS=OFF -DCMAKE_CUDA_ARCHITECTURES=80 cmake --build build --clean-first ``` Before: cc -std=gnu99 -c src/cupdlpx.c fails on cub/device/device_reduce.cuh. After: clean build, links cupdlpx and libcupdlpx.so with 0 errors.
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
1 participant
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
This adds optional AMD GPU support to cuPDLPx through ROCm/HIP, alongside the existing CUDA path. The CUDA build is unchanged when the new option is off.
What changed
The port keeps every device source in its existing CUDA spelling and routes the CUDA APIs to their HIP equivalents through a single compatibility header,
internal/cuda_to_hip.h. On a HIP build that header maps the CUDA runtime, cuBLAS, cuSPARSE, and CUB symbols to hipRT, hipBLAS, hipSPARSE, and hipCUB; on a CUDA build it includes the standard CUDA headers, so nothing about the NVIDIA path changes.internal/cusparse_compat.hselects the standardhipsparseSpMVpath on ROCm, since hipSPARSE does not provide thecusparseSpMVOpvariant.The build system gains a
USE_HIPoption (off by default). When enabled, the project is configured with the HIP language, the.cusources are compiled as HIP, and the targets link hipBLAS, hipSPARSE, and hipCUB instead of the CUDA libraries. GPU architectures are chosen withCMAKE_HIP_ARCHITECTURES(defaulting togfx90a). On Windows, the CLI-onlymps_parser.cis excluded from the core library because it relies onstrtok_r.The interface test gains a case that runs the GPU solver path with presolve disabled, so the hipBLAS/hipSPARSE execution path is exercised end to end. Enabling the test suite (
-DCUPDLPX_BUILD_TESTS=ON) to run the new case surfaced four stalezero_toleranceassignments intest_interface.creferencing amatrix_desc_tfield that no longer exists; they are removed so the file compiles.Building for AMD GPUs
Set
CMAKE_HIP_ARCHITECTURESto match the target GPU (for examplegfx90afor MI200,gfx1100for RDNA3 desktop, orgfx1201for RDNA4). If the ROCm install is not on CMake's default search path, point-DCMAKE_PREFIX_PATHat it (e.g./opt/rocm) sofind_packagecan locate hip, hipBLAS, hipSPARSE, and hipCUB. The resulting./build/cupdlpxbinary is invoked exactly as in the CUDA build. The README documents this alongside the CUDA build instructions.Validation
Built and run on an AMD Instinct MI200 (gfx90a) with ROCm 7.2.1: the full interface test suite passes, including the GPU solver case that exercises the hipBLAS and hipSPARSE paths (Status: OPTIMAL). The same configuration builds cleanly for gfx1100 (RDNA3) and gfx1201 (RDNA4). The CUDA build path is unaffected by these changes.
This work was prepared with assistance from Claude, an AI assistant by Anthropic.