mirror/LocalAI
1
0 Fork 0
mirror of https://github.com/mudler/LocalAI.git synced 2026-10-09 22:54:42 -04:00
Code Issues Packages Projects Releases Wiki Activity

fix(rocm): point *_TENSILE_LIBPATH at the architecture folder (#12411)

Browse Source 
Newer ROCm releases split the rocBLAS and hipBLASLt TensileLibrary data
into one folder per GPU architecture (lib/hipblaslt/library/gfx1151/...).
The run.sh of the ROCm-capable backends exports ROCBLAS_TENSILE_LIBPATH and
HIPBLASLT_TENSILE_LIBPATH as the parent folder, and both libraries only look
directly in that folder, so they miss every kernel:

  rocblaslt error: Cannot read ".../hipblaslt/library/TensileLibrary_lazy_gfx1151.dat"
  hipModuleLoad failed: .../hipblaslt/library/Kernels.so-000-gfx1151.hsaco

and fall back to slower code paths.

rocm_tensile_dir keeps the folder when it has files at the top level (the
older flat layout) or several architecture folders, and descends into the
architecture folder when the bundle carries exactly one.

Assisted-by: Claude:claude-opus-5-5

Signed-off-by: Stefan Walcz <stefan.walcz@walcz.de>
This commit is contained in:
Stefan Walcz authored and GitHub committed 2026-10-05 01:38:09 +02:00
1 parent e70c409a93
commit 9b0b7da96e
3 files changed
+57 -6
Download Patch File Download Diff File Expand all files Collapse all files

No files matched your search

backend/cpp/bonsai/run.sh
+19 -2
View File
@@ -25,6 +25,23 @@ if [ -n "$LLAMACPP_GRPC_SERVERS" ]; then
fi
fi
# Newer ROCm releases split the TensileLibrary data per GPU architecture
# (library/gfx1151/...). rocBLAS and hipBLASLt look for their kernels directly in
# the directory the *_TENSILE_LIBPATH variable names, so pointing at the parent
# makes them miss every kernel ("Cannot read TensileLibrary_lazy_gfx1151.dat")
# and fall back to slower paths. Point at the architecture folder when the bundle
# carries exactly one and nothing at the top level; otherwise keep the directory.
rocm_tensile_dir() {
local dir="$1"
if [ -z "$(find "$dir" -maxdepth 1 -type f -print -quit)" ]; then
set -- "$dir"/gfx*/
if [ "$#" -eq 1 ] && [ -d "$1" ]; then
dir="${1%/}"
fi
fi
echo "$dir"
}
# Extend ld library path with the dir where this script is located/lib
if [ "$(uname)" == "Darwin" ]; then
export DYLD_LIBRARY_PATH="$CURDIR"/lib:$DYLD_LIBRARY_PATH
@@ -32,13 +49,13 @@ else
export LD_LIBRARY_PATH="$CURDIR"/lib:$LD_LIBRARY_PATH
# Tell rocBLAS where to find TensileLibrary data (GPU kernel tuning files)
if [ -d "$CURDIR/lib/rocblas/library" ]; then
export ROCBLAS_TENSILE_LIBPATH="$CURDIR"/lib/rocblas/library
export ROCBLAS_TENSILE_LIBPATH="$(rocm_tensile_dir "$CURDIR"/lib/rocblas/library)"
fi
# Same for hipBLASLt (rocblaslt): the bundled libhipblaslt.so resolves its
# TensileLibrary_lazy_gfx*.dat kernel data relative to itself, so point it at
# the bundled data or it falls back to slow generic kernels (issue #10660).
if [ -d "$CURDIR/lib/hipblaslt/library" ]; then
export HIPBLASLT_TENSILE_LIBPATH="$CURDIR"/lib/hipblaslt/library
export HIPBLASLT_TENSILE_LIBPATH="$(rocm_tensile_dir "$CURDIR"/lib/hipblaslt/library)"
fi
# Backends built for Intel GPUs carry a copy of the Intel graphics driver,
# and libze_loader is only there in those builds. Level Zero looks for a
backend/cpp/llama-cpp/run.sh
+19 -2
View File
@@ -27,6 +27,23 @@ if [ -n "$LLAMACPP_GRPC_SERVERS" ]; then
fi
fi
# Newer ROCm releases split the TensileLibrary data per GPU architecture
# (library/gfx1151/...). rocBLAS and hipBLASLt look for their kernels directly in
# the directory the *_TENSILE_LIBPATH variable names, so pointing at the parent
# makes them miss every kernel ("Cannot read TensileLibrary_lazy_gfx1151.dat")
# and fall back to slower paths. Point at the architecture folder when the bundle
# carries exactly one and nothing at the top level; otherwise keep the directory.
rocm_tensile_dir() {
local dir="$1"
if [ -z "$(find "$dir" -maxdepth 1 -type f -print -quit)" ]; then
set -- "$dir"/gfx*/
if [ "$#" -eq 1 ] && [ -d "$1" ]; then
dir="${1%/}"
fi
fi
echo "$dir"
}
# Extend ld library path with the dir where this script is located/lib
if [ "$(uname)" == "Darwin" ]; then
export DYLD_LIBRARY_PATH="$CURDIR"/lib:$DYLD_LIBRARY_PATH
@@ -35,13 +52,13 @@ else
export LD_LIBRARY_PATH="$CURDIR"/lib:$LD_LIBRARY_PATH
# Tell rocBLAS where to find TensileLibrary data (GPU kernel tuning files)
if [ -d "$CURDIR/lib/rocblas/library" ]; then
export ROCBLAS_TENSILE_LIBPATH="$CURDIR"/lib/rocblas/library
export ROCBLAS_TENSILE_LIBPATH="$(rocm_tensile_dir "$CURDIR"/lib/rocblas/library)"
fi
# Same for hipBLASLt (rocblaslt): the bundled libhipblaslt.so resolves its
# TensileLibrary_lazy_gfx*.dat kernel data relative to itself, so point it at
# the bundled data or it falls back to slow generic kernels (issue #10660).
if [ -d "$CURDIR/lib/hipblaslt/library" ]; then
export HIPBLASLT_TENSILE_LIBPATH="$CURDIR"/lib/hipblaslt/library
export HIPBLASLT_TENSILE_LIBPATH="$(rocm_tensile_dir "$CURDIR"/lib/hipblaslt/library)"
fi
# Backends built for Intel GPUs carry a copy of the Intel graphics driver,
# and libze_loader is only there in those builds. Level Zero looks for a
backend/cpp/turboquant/run.sh
+19 -2
View File
@@ -28,6 +28,23 @@ if [ -n "$LLAMACPP_GRPC_SERVERS" ]; then
fi
fi
# Newer ROCm releases split the TensileLibrary data per GPU architecture
# (library/gfx1151/...). rocBLAS and hipBLASLt look for their kernels directly in
# the directory the *_TENSILE_LIBPATH variable names, so pointing at the parent
# makes them miss every kernel ("Cannot read TensileLibrary_lazy_gfx1151.dat")
# and fall back to slower paths. Point at the architecture folder when the bundle
# carries exactly one and nothing at the top level; otherwise keep the directory.
rocm_tensile_dir() {
local dir="$1"
if [ -z "$(find "$dir" -maxdepth 1 -type f -print -quit)" ]; then
set -- "$dir"/gfx*/
if [ "$#" -eq 1 ] && [ -d "$1" ]; then
dir="${1%/}"
fi
fi
echo "$dir"
}
# Extend ld library path with the dir where this script is located/lib
if [ "$(uname)" == "Darwin" ]; then
export DYLD_LIBRARY_PATH="$CURDIR"/lib:$DYLD_LIBRARY_PATH
@@ -35,13 +52,13 @@ else
export LD_LIBRARY_PATH="$CURDIR"/lib:$LD_LIBRARY_PATH
# Tell rocBLAS where to find TensileLibrary data (GPU kernel tuning files)
if [ -d "$CURDIR/lib/rocblas/library" ]; then
export ROCBLAS_TENSILE_LIBPATH="$CURDIR"/lib/rocblas/library
export ROCBLAS_TENSILE_LIBPATH="$(rocm_tensile_dir "$CURDIR"/lib/rocblas/library)"
fi
# Same for hipBLASLt (rocblaslt): the bundled libhipblaslt.so resolves its
# TensileLibrary_lazy_gfx*.dat kernel data relative to itself, so point it at
# the bundled data or it falls back to slow generic kernels (issue #10660).
if [ -d "$CURDIR/lib/hipblaslt/library" ]; then
export HIPBLASLT_TENSILE_LIBPATH="$CURDIR"/lib/hipblaslt/library
export HIPBLASLT_TENSILE_LIBPATH="$(rocm_tensile_dir "$CURDIR"/lib/hipblaslt/library)"
fi
# Backends built for Intel GPUs carry a copy of the Intel graphics driver,
# and libze_loader is only there in those builds. Level Zero looks for a