HIP build fails: undeclared cudaStreamCreateWithPriority (missing alias in vendors/hip.h)
Summary
Building audio.cpp with the HIP backend fails to compile the ggml CUDA sources.
common.cuh calls cudaStreamCreateWithPriority, but that symbol has no
CUDA→HIP alias in external/ggml/src/ggml-cuda/vendors/hip.h, so every .cu
file that includes common.cuh fails with use of undeclared identifier
(the compiler itself suggests the correct HIP equivalent).
Environment
- audio.cpp: 0.2.1
- ggml (vendored): 0.12.0, commit
8ceb983
- ROCm: 7.2.4 (
/opt/rocm-7.2.4)
- HIP compiler: Clang 22.0.0 (
/opt/rocm/lib/llvm/bin/clang)
- GPU target: gfx1150
- OS: Debian GNU/Linux forky/sid (forky) x86_64
- Configure summary: HIP backend enabled, hipBLASLt GEMM enabled
Steps to reproduce
- Configure with the HIP backend enabled (
build_hip build dir)
cmake --build build_hip (or make)
Error
Every HIP object fails identically, e.g. concat.cu, acc.cu, argmax.cu, ...:
In file included from external/ggml/src/ggml-cuda/concat.cu:1:
In file included from external/ggml/src/ggml-cuda/concat.cuh:1:
external/ggml/src/ggml-cuda/common.cuh:1461:28: error: use of undeclared identifier
'cudaStreamCreateWithPriority'; did you mean 'hipStreamCreateWithPriority'?
1461 | CUDA_CHECK(cudaStreamCreateWithPriority(
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.4/include/hip/hip_runtime_api.h:2835:12: note:
'hipStreamCreateWithPriority' declared here
1 error generated when compiling for gfx1150.
Build aborts with gmake: *** [Makefile:156: all] Error 2.
Cause
src/ggml-cuda/common.cuh (around line 1461) creates priority streams via
cudaStreamCreateWithPriority, but vendors/hip.h — the table of
#define cudaXxx hipXxx shims used when compiling with HIP — is missing the
mappings for the stream-priority APIs.
Suggested fix
The below fixes worked for me.
Add the following to src/ggml-cuda/vendors/hip.h alongside the other
cudaStream* defines:
#define cudaStreamCreateWithPriority hipStreamCreateWithPriority
#define cudaStreamGetPriority hipStreamGetPriority
#define cudaDeviceGetStreamPriorityRange hipDeviceGetStreamPriorityRange
The CUDA and HIP signatures for these three functions match 1:1, so the alias
shim is sufficient (this is the same mechanism the rest of the file uses).
With these defines added locally, the HIP build compiles past this point.
Workaround for anyone hitting this
cd external/ggml
# append the three defines next to the other cudaStream aliases in
# src/ggml-cuda/vendors/hip.h, then rebuild
HIP build fails: undeclared
cudaStreamCreateWithPriority(missing alias in vendors/hip.h)Summary
Building audio.cpp with the HIP backend fails to compile the ggml CUDA sources.
common.cuhcallscudaStreamCreateWithPriority, but that symbol has noCUDA→HIP alias in
external/ggml/src/ggml-cuda/vendors/hip.h, so every.cufile that includes
common.cuhfails withuse of undeclared identifier(the compiler itself suggests the correct HIP equivalent).
Environment
8ceb983/opt/rocm-7.2.4)/opt/rocm/lib/llvm/bin/clang)Steps to reproduce
build_hipbuild dir)cmake --build build_hip(ormake)Error
Every HIP object fails identically, e.g.
concat.cu,acc.cu,argmax.cu, ...:Build aborts with
gmake: *** [Makefile:156: all] Error 2.Cause
src/ggml-cuda/common.cuh(around line 1461) creates priority streams viacudaStreamCreateWithPriority, butvendors/hip.h— the table of#define cudaXxx hipXxxshims used when compiling with HIP — is missing themappings for the stream-priority APIs.
Suggested fix
The below fixes worked for me.
Add the following to
src/ggml-cuda/vendors/hip.halongside the othercudaStream*defines:The CUDA and HIP signatures for these three functions match 1:1, so the alias
shim is sufficient (this is the same mechanism the rest of the file uses).
With these defines added locally, the HIP build compiles past this point.
Workaround for anyone hitting this