AISpot

llama.cpp 修復 RDNA3/RDNA4 上 HIP bf16 WMMA 編譯失敗

llama.cpp releases版本更新地端推論開源

llama.cpp 修復了 HIP 後端在 AMD RDNA3(gfx1150/1151)與 RDNA4(gfx1200/1201)上 bf16 WMMA 的編譯失敗。原因是 __builtin_amdgcn_wmma_f32_16x16x16_bf16_w32 等內建要求 A/B 運算元為 short 向量而非 __bf16,原先按 __bf16 宣告分片會直接報錯。

原標題:b11847
閱讀原文

評分38 / 40(平均 39,門檻 60)
狀態未入選

原文

hip: fix bf16 WMMA operand vector type for gfx11/gfx12 ( #30885 ) The __builtin_amdgcn_wmma_f32_16x16x16_bf16_w32 (and _gfx12) builtins take their A/B operands as vectors of short (_ExtVector<16, short> / _ExtVector<8, short>), not __bf16. Declaring the fragments as ext_vector_type(N) __bf16 fails to compile: error: cannot initialize a parameter of type 'short attribute ((ext_vector_type(16)))' with an lvalue of type 'const bf16x16_t' (vector of 16 '__bf16' values) short and __bf16 are both 16-bit and the value is only ever passed to the builtin via reinterpret_cast over raw bytes, so changing the element type is bit-identical. The operand type has been short since the builtin was introduced, so this works across LLVM/ROCm versions without a version gate. Fixes HIP builds for RDNA3 (gfx1150/1151) and RDNA4 (gfx1200/1201). Signed-off-by: Pablo Cholaky [email protected] Website: https://llama.app Attestations: https://github.com/ggml-org/llama.cpp/attestations/59316876 macOS/iOS: macOS Apple Silicon (arm64) macOS Apple Silicon (arm64, KleidiAI enabled) DISABLED macOS Intel (x64) iOS XCFramework Linux: Ubuntu x64 (CPU) Ubuntu arm64 (CPU) Ubuntu s390x (CPU) Ubuntu x64 (Vulkan) Ubuntu arm64 (Vulkan) Ubuntu x64 (CUDA 12) - CUDA 12.8 libraries Ubuntu x64 (CUDA 13) - CUDA 13.4 libraries Ubuntu arm64 (CUDA 13) - CUDA 13.4 libraries Ubuntu x64 (ROCm 10.0) Ubuntu x64 (OpenVINO) Ubuntu x64 (SYCL FP32) Ubuntu x64 (SYCL FP16) Linux arm64 (Snapdragon: CPU, Adreno GPU, Hexagon NPU) - setup guide Android: Android arm64 (CPU) Android arm64 (Snapdragon: CPU, Adreno GPU, Hexagon NPU) - setup guide Windows: Windows x64 (CPU) Windows arm64 (CPU) Windows arm64 (OpenCL Adreno) Windows x64 (CUDA 12) - CUDA 12.4 DLLs Windows x64 (CUDA 13) - CUDA 13.4 DLLs Windows arm64 (CUDA 13) - CUDA 13.4 DLLs Windows x64 (Vulkan) Windows arm64 (Vulkan) Windows x64 (OpenVINO) Windows x64 (SYCL) Windows x64 (ROCm 10.0) openEuler: DISABLED openEuler x86 (310p) openEuler x86 (910b, ACL Graph) openEuler aarch64 (310p) openEuler aarch64 (910b, ACL Graph) UI: UI

相關報導

llama.cpp releasesAI 評分37

llama.cpp 修復 ARM NEON fp32→fp16 未走硬體指令

llama.cpp 建置 b11761 修復了 ARM NEON 分支只選中 fp16→fp32、漏掉 fp32→fp16 的問題:此前該方向一直靜默回退到軟體實現,已定義的 fcvt 硬體指令無人呼叫。補丁讓該分支同時選中兩個方向,並加入 #error,任何定義了 COMPUTE 宏卻未選中對應實現的平台分支會直接編譯失敗。fp16→fp32 方向按 PR 30343 仍刻意走查表,因此保留 !defined(F16C) 例外;

llama.cpp releasesAI 評分42

llama.cpp 修復 HIP 建置:補充位元組 permute 回退

llama.cpp 修復了 HIP 後端在缺少 __builtin_amdgcn_perm 時的建置失敗。由於 amdgcnspirv 會拒絕未經 __builtin_amdgcn_is_invocable 保護的目標專用 builtin,此前 145 個 HIP 檔案中有 66 個無法編譯。此次改動把 perm 呼叫封裝進一個輔助函式,目標支援時保留 v_perm_b32,否則改用可移植的位元組選擇實現;輔助函式被移到 vendors/hip.h。

llama.cpp releasesAI 評分36

llama.cpp 修復 Vulkan 後端 RDNA4 矩陣向量調優

llama.cpp 發布 b11389 版本,修復了 Vulkan 後端在 RDNA4 架構上的矩陣向量運算調優問題(PR #29934)。該版本繼續提供覆蓋 macOS、Linux、Windows、Android 等平台的預編譯二進位制,包括 Vulkan、CUDA、ROCm、SYCL 等後端,其中 macOS Apple Silicon 的 KleidiAI 啟用版和 openEuler 建置被停用。