ROCm 7 build fails: #3267 compat probes test for macros that HIP 7 delivers as enum members and functions
メンテナーはふだん 1 日以内に返信
まだ誰も着手していません。
評価
- 難易度
- 3/5
- 見積もり時間
- 1〜2日
- 初心者へのやさしさ
- 55/100
- issue の種類
- バグ
- 明瞭さ
- 明確に書かれている
- 活発さ
- 活発
- 技術スタック
- cmake, cpp
- 領域
- backend, build-system
調査の方向性
The issue is in two files: src/vt/rocm/rocm_matmul_hipblaslt.hip and include/vt/rocm/hip_shfl_compat.h. Start by checking the HIP_VERSION_MAJOR macro and the existing #ifndef probes. Look at the ROCm 7 headers to confirm HIPBLAS_COMPUTE_32F is an enum and __shfl_sync are functions. The fix involves replacing #ifndef with version checks. Test the build with the ROCm 7 toolchain after making changes.
索引モデルが issue の本文から書いたものです。
説明
Environment
- OS: Fedora 44
- ROCm 7.1.52802 (
HIP_VERSION_MAJOR= 7), clang from the ROCm toolchain - Targets:
gfx1102(Radeon RX 7700S 8GB VRAM) andgfx1103(Radeon 780M, UMA) - Host: Ryzen 9 7940HS, Framework 16, 64 GB DDR5 RAM, ~54GB usable for the host
- Tree:
mainafter #3267,cmakeHIP preset, tests excluded
Problem
The ROCm 5.7 compatibility layer from #3267 uses #ifndef probes. Two of
them misfire on ROCm 7, and the build fails for both RDNA3 targets.
-
src/vt/rocm/rocm_matmul_hipblaslt.hipprobes#ifndef HIPBLAS_COMPUTE_32F.
On ROCm 7 that name is an enumerator ofhipblasComputeType_tin
hipblas-common.h, not a macro. The probe reads true, so the pre-ROCm6
block activates and typedefshipblasDatatype_t, a type ROCm 7 removed.
Result: 20 compile errors in that file. -
include/vt/rocm/hip_shfl_compat.hprobes#ifndef __shfl_down_syncand
#ifndef __shfl_sync. On ROCm 6 and later those names exist as functions
inamd_warp_sync_functions.h, invisible to#ifndef. The macro shims
activate anyway and collide with the real declarations.
Root cause
#ifndef answers "does a macro with this name exist". HIP delivers API
changes as enumerators and functions, so the probe tests the wrong property.
Every ROCm release that moves a name from macro to enum or function form
silently re-arms these branches.
Proposed fix
Gate version-dependent shims on HIP_VERSION_MAJOR, and keep #ifndef
probes only for names that stay macros across releases:
// hip_shfl_compat.h: the _sync wrappers exist as functions on ROCm 6+
#if !defined(HIP_VERSION_MAJOR) || HIP_VERSION_MAJOR < 6
...existing shims...
#endif
// rocm_matmul_hipblaslt.hip: enum form landed in ROCm 6,
// hipblasDatatype_t was removed in ROCm 7
#if !defined(HIPBLAS_COMPUTE_32F) && \
(!defined(HIP_VERSION_MAJOR) || HIP_VERSION_MAJOR < 7)
...existing typedef and aliases...
#endif
Other propositions
Maybe... we should also think about shared header for ROCm versioning handling?
Other notes
I already have local version with fix. Tested on Granite 4.2 3b safetensors, fp8 and bf16. Results on iGPU and dGPU:
| Device | bf16 | fp8 |
|---|---|---|
| 7700S | 22.8 t/s, 6.9G | 22.3 t/s, 3.9G |
| 780M | 10.2 t/s | 8.8 t/s |
- 主要言語
- C++
- スター
- 423
- フォーク
- 53
- 平均マージ
- 1日 5時間
- マージ済み PR（30日）
- 376
環境構築
はじめの一歩
- issue を最後まで読み、次にプロジェクトのコントリビューションガイドを読みます。
- 着手することを issue にコメントします — 二人が同じ作業をするのを防げます。
- リポジトリをフォークし、ブランチを切って変更します。
- issue 番号を参照したプルリクエストを送ります。
mudler/vllm.cpp のほかの issue
-
難易度 2/5 1〜3時間 初心者へのやさしさ 84/100
メンテナーはふだん 1 日以内に返信
-
難易度 2/5 1〜3時間 初心者へのやさしさ 82/100
メンテナーはふだん 1 日以内に返信
-
難易度 2/5 1〜3時間 初心者へのやさしさ 70/100
メンテナーはふだん 1 日以内に返信
-
難易度 1/5 1時間未満 初心者へのやさしさ 88/100
メンテナーはふだん 1 日以内に返信
-
難易度 2/5 1〜3時間 初心者へのやさしさ 84/100
メンテナーはふだん 1 日以内に返信
mudler/vllm.cpp の issue をすべて見る
似ている issue
-
難易度 2/5 半日 初心者へのやさしさ 84/100
-
難易度 2/5 1〜3時間 初心者へのやさしさ 84/100
メンテナーはふだん 1 日以内に返信
-
難易度 1/5 1〜3時間 初心者へのやさしさ 88/100
ROCm/rocm-libraries#12703 ·
メンテナーはふだん 2 日以内に返信
-
bug
難易度 1/5 1〜3時間 初心者へのやさしさ 88/100
isl-org/Open3D#7585 · コメント 1 件 ·
メンテナーはふだん 2 日以内に返信
-
Feature request
難易度 2/5 1〜3時間 初心者へのやさしさ 76/100
qbittorrent/qBittorrent#24975 ·
メンテナーはふだん 3 日以内に返信