From efd04b607db4c507ac49ea37fa88dbb8cd09a0c6 Mon Sep 17 00:00:00 2001 From: sstamenk Date: Fri, 14 Aug 2026 14:56:00 +0000 Subject: [PATCH 1/5] Fix ROCm library version detection Use ROCm release metadata consistently at build and runtime while keeping packaged-binary fallbacks explicit. --- .github/scripts/build-rocm.sh | 61 +++++---- CMakeLists.txt | 114 +++++++++++---- bitsandbytes/cextension.py | 127 ++++++++++++----- bitsandbytes/cuda_specs.py | 25 +++- bitsandbytes/diagnostics/cuda.py | 17 ++- bitsandbytes/diagnostics/main.py | 3 +- csrc/compat.cuh | 2 - csrc/ops.cu | 6 - docs/source/errors.mdx | 9 +- docs/source/installation.mdx | 6 +- tests/test_cuda_setup_evaluator.py | 213 +++++++++++++++-------------- 11 files changed, 365 insertions(+), 218 deletions(-) diff --git a/.github/scripts/build-rocm.sh b/.github/scripts/build-rocm.sh index 7f971824a..c4981616e 100644 --- a/.github/scripts/build-rocm.sh +++ b/.github/scripts/build-rocm.sh @@ -1,28 +1,28 @@ #!/bin/bash set -xeuo pipefail -: "${RUNNER_OS:?RUNNER_OS must be set (Linux/Windows)}" -: "${ROCM_VERSION:?ROCM_VERSION must be set}" - rocm_version_at_least() { - local required_version="$1" - local current_major current_minor required_major required_minor - - IFS=. read -r current_major current_minor _ <<< "${ROCM_VERSION}" - IFS=. read -r required_major required_minor _ <<< "${required_version}" + local required_major required_minor - if ((current_major > required_major)); then - return 0 - fi - if ((current_major < required_major)); then - return 1 - fi - if ((current_minor >= required_minor)); then - return 0 - fi - return 1 + IFS=. read -r required_major required_minor <<< "$1" + ((rocm_version_major > required_major || + (rocm_version_major == required_major && rocm_version_minor >= required_minor))) } +if [[ "${RUNNER_OS:-}" != "Linux" && "${RUNNER_OS:-}" != "Windows" ]]; then + echo "Invalid RUNNER_OS '${RUNNER_OS:-}'; expected Linux or Windows." >&2 + exit 1 +fi + +if [[ ! "${ROCM_VERSION:-}" =~ ^([0-9]+)\.([0-9]+)(\.[0-9]+)?$ ]]; then + echo "Invalid ROCM_VERSION '${ROCM_VERSION:-}'; expected a dotted ROCm release." >&2 + exit 1 +fi + +rocm_version_major="$((10#${BASH_REMATCH[1]}))" +rocm_version_minor="$((10#${BASH_REMATCH[2]}))" +rocm_version_tag="${rocm_version_major}${rocm_version_minor}" + bnb_rocm_arch="gfx90a;gfx942;gfx1100;gfx1101;gfx1102;gfx1103" # ROCm 6.4+ - Add RDNA4 and RDNA3.5 targets. Note we assume >=6.4.4. @@ -35,14 +35,9 @@ if rocm_version_at_least "7.0"; then bnb_rocm_arch="${bnb_rocm_arch};gfx950" fi -# ROCm 7.14+ - Add CDNA1 and RDNA2 targets. -if rocm_version_at_least "7.14"; then - bnb_rocm_arch="${bnb_rocm_arch};gfx908;gfx1030;gfx1031;gfx1032;gfx1033;gfx1034;gfx1035;gfx1036" -fi - -# ROCm 7.14+ - Add CDNA5 (gfx1250). +# ROCm 7.14+ - Add CDNA1, CDNA5, and RDNA2 targets. if rocm_version_at_least "7.14"; then - bnb_rocm_arch="${bnb_rocm_arch};gfx1250" + bnb_rocm_arch="${bnb_rocm_arch};gfx908;gfx1030;gfx1031;gfx1032;gfx1033;gfx1034;gfx1035;gfx1036;gfx1250" fi if [ "${RUNNER_OS}" == "Linux" ]; then @@ -55,7 +50,7 @@ if [ "${RUNNER_OS}" == "Linux" ]; then docker run --rm -i \ -w /src -v "$PWD:/src" "$image" sh -c \ "pip install cmake==3.31.6 \ - && cmake -DCOMPUTE_BACKEND=hip -DCMAKE_BUILD_TYPE=MinSizeRel -DCMAKE_HIP_FLAGS=\"--offload-compress\" -DBNB_ROCM_ARCH=\"${bnb_rocm_arch}\" . \ + && cmake -DCOMPUTE_BACKEND=hip -DROCM_VERSION=\"${ROCM_VERSION}\" -DCMAKE_BUILD_TYPE=MinSizeRel -DCMAKE_HIP_FLAGS=\"--offload-compress\" -DBNB_ROCM_ARCH=\"${bnb_rocm_arch}\" . \ && cmake --build . --parallel" else bnb_rocm_arch="gfx1100;gfx1101;gfx1102;gfx1150;gfx1151;gfx1200;gfx1201" @@ -84,6 +79,7 @@ else cmake -G Ninja \ -DCOMPUTE_BACKEND=hip \ + -DROCM_VERSION="${ROCM_VERSION}" \ -DBNB_ROCM_ARCH="${bnb_rocm_arch}" \ -DCMAKE_BUILD_TYPE=MinSizeRel \ -DCMAKE_HIP_FLAGS="--offload-compress" \ @@ -94,4 +90,15 @@ fi output_dir="output/${RUNNER_OS}/X64" mkdir -p "${output_dir}" -(shopt -s nullglob && cp bitsandbytes/*.{so,dylib,dll} "${output_dir}") + +shopt -s nullglob +libraries=(bitsandbytes/libbitsandbytes_rocm${rocm_version_tag}.{so,dylib,dll}) +shopt -u nullglob + +if [ "${#libraries[@]}" -eq 0 ]; then + expected_pattern="bitsandbytes/libbitsandbytes_rocm${rocm_version_tag}.{so,dylib,dll}" + echo "Expected ROCm ${ROCM_VERSION} library was not built: ${expected_pattern}" >&2 + exit 1 +fi + +cp "${libraries[@]}" "${output_dir}/" diff --git a/CMakeLists.txt b/CMakeLists.txt index 950f8d213..9e6353744 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -10,17 +10,21 @@ # Separate by semicolons, i.e. `-DCOMPUTE_CAPABILITY=89;90;100;120` # Check your compute capability here: https://developer.nvidia.com/cuda-gpus # - PTXAS_VERBOSE: Pass the `-v` option to the PTX Assembler -# - ROCM_VERSION: Override the ROCm version shortcode used in the output library name. -# Useful when PyTorch was built against a different ROCm version than the -# system install. For example, `-DROCM_VERSION=70` produces -# libbitsandbytes_rocm70.so even if the system has ROCm 7.2. +# - ROCM_VERSION: Override the ROCm release used in the output library name. Accepts a +# dotted release or shortcode. For example, `7.14.0` and `714` +# both produce libbitsandbytes_rocm714.so. cmake_minimum_required(VERSION 3.22.1) # On Windows with HIP backend, auto-detect compilers from ROCM_PATH before project() if(WIN32 AND COMPUTE_BACKEND STREQUAL "hip") - if(DEFINED ENV{ROCM_PATH}) - file(TO_CMAKE_PATH "$ENV{ROCM_PATH}" ROCM_PATH) + if(NOT DEFINED ENV{ROCM_PATH} OR "$ENV{ROCM_PATH}" STREQUAL "") + message(FATAL_ERROR + "ROCM_PATH must be set for HIP builds on Windows. " + "After 'rocm-sdk init', set it from 'rocm-sdk path --root'. " + "PowerShell: $env:ROCM_PATH = (rocm-sdk path --root)" + ) endif() + file(TO_CMAKE_PATH "$ENV{ROCM_PATH}" ROCM_PATH) if(ROCM_PATH AND NOT DEFINED CMAKE_CXX_COMPILER) set(CMAKE_CXX_COMPILER "${ROCM_PATH}/lib/llvm/bin/clang++.exe") endif() @@ -282,20 +286,75 @@ elseif(BUILD_HIP) string(APPEND BNB_OUTPUT_NAME "_rocm") - # get hip version - execute_process(COMMAND hipconfig --version OUTPUT_VARIABLE HIP_CONFIG_VERSION) - string(REGEX MATCH "[0-9]+\\.[0-9]+" HIP_VERSION "${HIP_CONFIG_VERSION}") - string(REPLACE "." "" HIP_VERSION_SHORT "${HIP_VERSION}") + # HIP and ROCm releases can diverge, so use the ROCm release for the filename. + set(ROCM_VERSION "" CACHE STRING "ROCm release used in the output library name") + set(_ROCM_VERSION_RAW "${ROCM_VERSION}") + set(_ROCM_VERSION_SOURCE "-DROCM_VERSION") + + if(NOT _ROCM_VERSION_RAW) + if(NOT ROCM_PATH) + if(DEFINED ENV{ROCM_PATH}) + file(TO_CMAKE_PATH "$ENV{ROCM_PATH}" ROCM_PATH) + elseif(CMAKE_HIP_COMPILER_ROCM_ROOT) + set(ROCM_PATH "${CMAKE_HIP_COMPILER_ROCM_ROOT}") + elseif(WIN32) + message(FATAL_ERROR "ROCM_PATH must be set for HIP builds on Windows") + else() + message(WARNING "ROCM_PATH is not set; falling back to /opt/rocm") + set(ROCM_PATH "/opt/rocm") + endif() + endif() - # Expose a cache variable that the user can set to override the ROCm version in the library name - set(ROCM_VERSION "${HIP_VERSION_SHORT}" CACHE STRING "Expected ROCm Version Shortcode") + foreach(_VERSION_FILE "${ROCM_PATH}/.info/version" "${ROCM_PATH}/core/.info/version") + if(NOT _ROCM_VERSION_RAW AND EXISTS "${_VERSION_FILE}") + file(READ "${_VERSION_FILE}" _ROCM_VERSION_RAW) + set(_ROCM_VERSION_SOURCE "${_VERSION_FILE}") + endif() + endforeach() + + if(NOT _ROCM_VERSION_RAW) + execute_process( + COMMAND rocm-sdk version + OUTPUT_VARIABLE _ROCM_VERSION_RAW + ERROR_QUIET + OUTPUT_STRIP_TRAILING_WHITESPACE + ) + set(_ROCM_VERSION_SOURCE "rocm-sdk version") + endif() + + if(NOT _ROCM_VERSION_RAW) + message(FATAL_ERROR + "Could not determine the ROCm version from ${ROCM_PATH}/.info/version, " + "${ROCM_PATH}/core/.info/version, or 'rocm-sdk version'. " + "Set ROCM_PATH correctly or pass -DROCM_VERSION=." + ) + endif() + endif() - message(STATUS "ROCm Version: ${HIP_VERSION_SHORT} (from hipconfig)") - if(NOT ROCM_VERSION STREQUAL "${HIP_VERSION_SHORT}") - message(WARNING "Overriding ROCm version in library name: ${HIP_VERSION_SHORT} -> ${ROCM_VERSION}") + string(STRIP "${_ROCM_VERSION_RAW}" _ROCM_VERSION_RAW) + if(_ROCM_VERSION_RAW MATCHES "^([0-9]+)\\.([0-9]+)") + set(_ROCM_VERSION_MAJOR "${CMAKE_MATCH_1}") + set(_ROCM_VERSION_MINOR "${CMAKE_MATCH_2}") + elseif(_ROCM_VERSION_RAW MATCHES "^([6-9])([0-9]+)$") + set(_ROCM_VERSION_MAJOR "${CMAKE_MATCH_1}") + set(_ROCM_VERSION_MINOR "${CMAKE_MATCH_2}") + elseif(_ROCM_VERSION_RAW MATCHES "^([1-5][0-9])([0-9]+)$") + set(_ROCM_VERSION_MAJOR "${CMAKE_MATCH_1}") + set(_ROCM_VERSION_MINOR "${CMAKE_MATCH_2}") + else() + message(FATAL_ERROR + "Could not parse ROCm version '${_ROCM_VERSION_RAW}'. " + "Pass -DROCM_VERSION=7.14.0 (or 714)." + ) endif() - string(APPEND BNB_OUTPUT_NAME "${ROCM_VERSION}") + set(ROCM_VERSION "${_ROCM_VERSION_MAJOR}.${_ROCM_VERSION_MINOR}") + set(_ROCM_VERSION_TAG "${_ROCM_VERSION_MAJOR}${_ROCM_VERSION_MINOR}") + message(STATUS + "ROCm Release: ${_ROCM_VERSION_MAJOR}.${_ROCM_VERSION_MINOR}; " + "library suffix: rocm${_ROCM_VERSION_TAG} (from ${_ROCM_VERSION_SOURCE})" + ) + string(APPEND BNB_OUTPUT_NAME "${_ROCM_VERSION_TAG}") add_compile_definitions(__HIP_PLATFORM_AMD__) add_compile_definitions(__HIP_PLATFORM_HCC__) add_compile_definitions(BUILD_HIP) @@ -416,11 +475,16 @@ if(BUILD_CUDA) ) endif() if(BUILD_HIP) - # Determine ROCM_PATH from environment variable, fallback to /opt/rocm on Linux - if(DEFINED ENV{ROCM_PATH}) - file(TO_CMAKE_PATH "$ENV{ROCM_PATH}" ROCM_PATH) - else() - set(ROCM_PATH /opt/rocm) + # ROCM_PATH was resolved during version detection; retain it for package discovery. + if(NOT ROCM_PATH) + if(DEFINED ENV{ROCM_PATH}) + file(TO_CMAKE_PATH "$ENV{ROCM_PATH}" ROCM_PATH) + elseif(WIN32) + message(FATAL_ERROR "ROCM_PATH must be set for ROCm builds on Windows") + else() + message(WARNING "ROCM_PATH is not set; falling back to /opt/rocm") + set(ROCM_PATH "/opt/rocm") + endif() endif() list(APPEND CMAKE_PREFIX_PATH ${ROCM_PATH}) macro(find_package_and_print_version PACKAGE_NAME) @@ -452,12 +516,8 @@ if(BUILD_HIP) set_source_files_properties(${GPU_FILES} PROPERTIES LANGUAGE HIP) set_target_properties(bitsandbytes PROPERTIES LINKER_LANGUAGE CXX) - if(HIP_VERSION VERSION_LESS "6.1") - target_compile_definitions(bitsandbytes PUBLIC NO_HIPBLASLT) - else() - find_package(hipblaslt) - target_link_libraries(bitsandbytes PUBLIC roc::hipblaslt) - endif() + find_package(hipblaslt REQUIRED) + target_link_libraries(bitsandbytes PUBLIC roc::hipblaslt) endif() if(BUILD_XPU) set(SYCL_LINK_FLAGS "-fsycl;--offload-compress;-fsycl-targets=spir64_gen,spir64;-Xs;-device pvc,xe-lpg,ats-m150 -options ' -cl-intel-enable-auto-large-GRF-mode -cl-poison-unsupported-fp64-kernels -cl-intel-greater-than-4GB-buffer-required'") diff --git a/bitsandbytes/cextension.py b/bitsandbytes/cextension.py index e234f20d3..efc92363f 100644 --- a/bitsandbytes/cextension.py +++ b/bitsandbytes/cextension.py @@ -25,58 +25,88 @@ def get_cuda_bnb_library_path(cuda_specs: CUDASpecs) -> Path: When no override is set, selects from packaged libraries using the following priority: 1. Exact version match. - 2. Highest packaged version <= runtime version, same major (e.g. runtime 12.9, ship 12.8). - 3. Lowest packaged version > runtime version, same major (e.g. runtime 12.0, ship 12.1). - No cross-major fallback: if no same-major library exists, returns the exact non-existent - path so the caller raises a clear "not found" error. - A warning is logged when falling back. Override env vars bypass selection entirely - and load the named version with no fallback. The returned path is not guaranteed to - exist when no packaged libs are found, or when an override names an absent version. + 2. Same-major fallback, preferring the newest older binary, then the + lowest newer binary. + 3. For ROCm only, cross-major fallback using the same older-first policy. + CUDA never falls back across major versions. + A warning is logged when falling back. Overrides select the requested filename + directly. The returned path is not guaranteed to exist when no packaged libs + are found, or when an override names an absent version. """ is_hip = bool(torch.version.hip) prefix = "rocm" if is_hip else "cuda" override_var = "BNB_ROCM_VERSION" if is_hip else "BNB_CUDA_VERSION" + other_override_var = "BNB_CUDA_VERSION" if is_hip else "BNB_ROCM_VERSION" + + if os.environ.get(other_override_var): + logger.warning( + "%s is ignored because PyTorch is using %s; use %s instead.", + other_override_var, + "ROCm" if is_hip else "CUDA", + override_var, + ) override_value = os.environ.get(override_var) if override_value is not None: - if not override_value.isdigit(): - raise RuntimeError(f"{override_var}={override_value!r}: value must be digits only (e.g. '124' for 12.4).") - library_name = f"libbitsandbytes_{prefix}{override_value}{DYNAMIC_LIBRARY_SUFFIX}" + try: + override_version = _parse_version_override(override_value, is_hip) + except ValueError as error: + example = "7.14 or 714" if is_hip else "12.8 or 128" + raise RuntimeError( + f"{override_var}={override_value!r}: expected a dotted version or shortcode ({example})." + ) from error + + version_tag = _format_native_version_tag(override_version) + library_name = f"libbitsandbytes_{prefix}{version_tag}{DYNAMIC_LIBRARY_SUFFIX}" + override_path = PACKAGE_DIR / library_name logger.warning( - f"WARNING: {override_var}={override_value} environment variable detected; loading {library_name}.\n" + f"WARNING: {override_var}={override_value} environment variable detected; " + f"loading {override_path.name}.\n" f"This overrides automatic {'ROCm' if is_hip else 'CUDA'} version selection.\n" f"If this was unintended clear the variable and retry: unset {override_var}\n", ) - return PACKAGE_DIR / library_name + return override_path available = _find_cuda_libs(prefix, is_hip) runtime_version = cuda_specs.cuda_version_tuple + runtime_tag = _format_native_version_tag(runtime_version) if not available: - return PACKAGE_DIR / f"libbitsandbytes_{prefix}{cuda_specs.cuda_version_string}{DYNAMIC_LIBRARY_SUFFIX}" + return PACKAGE_DIR / f"libbitsandbytes_{prefix}{runtime_tag}{DYNAMIC_LIBRARY_SUFFIX}" if runtime_version in available: return available[runtime_version] - lower = [v for v in available if v[0] == runtime_version[0] and v < runtime_version] + missing_path = PACKAGE_DIR / f"libbitsandbytes_{prefix}{runtime_tag}{DYNAMIC_LIBRARY_SUFFIX}" + same_major = [version for version in available if version[0] == runtime_version[0]] + cross_major = False + lower = [version for version in same_major if version < runtime_version] if lower: selected = max(lower) + elif same_major: + selected = min(same_major) + elif is_hip: + lower = [version for version in available if version < runtime_version] + selected = max(lower) if lower else min(available) + cross_major = True else: - higher_same = [v for v in available if v[0] == runtime_version[0] and v > runtime_version] - if higher_same: - selected = min(higher_same) - else: - # No same-major library available. Return the non-existent exact path so - # get_native_library() raises a clear "not found" error. - return PACKAGE_DIR / f"libbitsandbytes_{prefix}{cuda_specs.cuda_version_string}{DYNAMIC_LIBRARY_SUFFIX}" - - logger.warning( - f"No prebuilt binary for {'ROCm' if is_hip else 'CUDA'} " - f"{runtime_version[0]}.{runtime_version[1]}, loading " - f"{'ROCm' if is_hip else 'CUDA'} {selected[0]}.{selected[1]} instead. " - f"Set {override_var} to override." - ) + return missing_path + + if cross_major: + logger.warning( + f"No prebuilt binary for ROCm {runtime_version[0]}.{runtime_version[1]}, loading " + f"ROCm {selected[0]}.{selected[1]} across major releases. This binary may be incompatible " + "or may not contain code for your GPU architecture. " + f"Set {override_var} to override or compile from source." + ) + else: + logger.warning( + f"No prebuilt binary for {'ROCm' if is_hip else 'CUDA'} " + f"{runtime_version[0]}.{runtime_version[1]}, loading " + f"{'ROCm' if is_hip else 'CUDA'} {selected[0]}.{selected[1]} instead. " + f"Set {override_var} to override." + ) return available[selected] @@ -124,26 +154,47 @@ def __init__(self, lib: ct.CDLL): lib.cget_managed_ptr.restype = ct.c_void_p -def _split_cuda_version(compact: str, is_hip: bool) -> tuple[int, int]: - """Split a compact CUDA/ROCm version string from a library filename into (major, minor). +def _format_native_version_tag(version: tuple[int, int]) -> str: + major, minor = version + return f"{major}{minor}" + + +def _split_cuda_version(version_tag: str, is_hip: bool) -> tuple[int, int]: + """Split a CUDA/ROCm library filename tag into (major, minor). CUDA: major is always 2 digits (11, 12, 13...), e.g. '118' -> (11, 8), '132' -> (13, 2). - ROCm: major is always 1 digit for now (6, 7...), e.g. '72' -> (7, 2), '713' -> (7, 13). - Note: revisit if ROCm major reaches 10. + ROCm: supported majors 6-9 use one digit, e.g. '72' -> (7, 2) + and '713' -> (7, 13). Tags starting with 1-5 reserve two major digits. """ if is_hip: - return int(compact[:1]), int(compact[1:]) - return int(compact[:2]), int(compact[2:]) + if len(version_tag) >= 3 and version_tag[0] in "12345": + return int(version_tag[:2]), int(version_tag[2:]) + return int(version_tag[:1]), int(version_tag[1:]) + return int(version_tag[:2]), int(version_tag[2:]) + + +def _parse_version_override(value: str, is_hip: bool) -> tuple[int, int]: + dotted = re.fullmatch(r"(\d+)\.(\d+)(?:\.\d+(?:[A-Za-z0-9+_.-]*)?)?", value) + if dotted: + return int(dotted.group(1)), int(dotted.group(2)) + if value.isdigit(): + return _split_cuda_version(value, is_hip) + raise ValueError(f"Invalid version override: {value}") def _find_cuda_libs(prefix: str, is_hip: bool) -> dict[tuple[int, int], Path]: """Return a {(major, minor): Path} mapping for all packaged CUDA/ROCm library files.""" result = {} for lib in PACKAGE_DIR.glob(f"libbitsandbytes_{prefix}*{DYNAMIC_LIBRARY_SUFFIX}"): - match = re.search(rf"{prefix}(\d+)", lib.name) + match = re.fullmatch( + rf"libbitsandbytes_{re.escape(prefix)}(\d+){re.escape(DYNAMIC_LIBRARY_SUFFIX)}", + lib.name, + ) if match: try: - result[_split_cuda_version(match.group(1), is_hip)] = lib + version_tag = match.group(1) + version = _split_cuda_version(version_tag, is_hip) + result[version] = lib except (ValueError, IndexError): continue return result @@ -153,11 +204,11 @@ def get_available_cuda_binary_versions() -> list[str]: """Get formatted CUDA/ROCm versions from existing library files.""" is_hip = bool(torch.version.hip) prefix = "rocm" if is_hip else "cuda" - return sorted(f"{major}.{minor}" for major, minor in _find_cuda_libs(prefix, is_hip)) + return [f"{major}.{minor}" for major, minor in sorted(_find_cuda_libs(prefix, is_hip))] def parse_cuda_version(version_str: str) -> str: - """Convert a raw version code string (e.g. '118', '713') to a dotted version (e.g. '11.8', '7.13').""" + """Convert a compact version tag (e.g. '118', '714') to a dotted version.""" if version_str.isdigit(): is_hip = bool(torch.version.hip) try: diff --git a/bitsandbytes/cuda_specs.py b/bitsandbytes/cuda_specs.py index 25ce3cd1e..4811af1f5 100644 --- a/bitsandbytes/cuda_specs.py +++ b/bitsandbytes/cuda_specs.py @@ -24,15 +24,32 @@ def get_compute_capabilities() -> list[tuple[int, int]]: return sorted(torch.cuda.get_device_capability(torch.cuda.device(i)) for i in range(torch.cuda.device_count())) +@lru_cache(None) +def get_rocm_version() -> Optional[str]: + """Get the ROCm release used to build PyTorch, with a legacy HIP fallback.""" + version = getattr(torch.version, "rocm", None) + if version is not None: + return version + + version = getattr(torch.version, "hip", None) + logging.getLogger(__name__).warning( + "torch.version.rocm is unavailable; falling back to legacy torch.version.hip=%s for ROCm library selection. " + "HIP and ROCm versions may differ.", + version, + ) + return version + + @lru_cache(None) def get_cuda_version_tuple() -> Optional[tuple[int, int]]: - """Get CUDA/HIP version as a tuple of (major, minor).""" + """Get the CUDA or ROCm release as a tuple of (major, minor).""" try: if torch.version.cuda: version_str = torch.version.cuda - elif torch.version.hip: - version_str = torch.version.hip else: + version_str = get_rocm_version() + + if not version_str: return None parts = version_str.split(".") @@ -44,7 +61,7 @@ def get_cuda_version_tuple() -> Optional[tuple[int, int]]: def get_cuda_version_string() -> Optional[str]: - """Get CUDA/HIP version as a string.""" + """Get the compact CUDA/ROCm version string retained for API compatibility.""" version_tuple = get_cuda_version_tuple() if version_tuple is None: return None diff --git a/bitsandbytes/diagnostics/cuda.py b/bitsandbytes/diagnostics/cuda.py index 655da84a0..ef436c954 100644 --- a/bitsandbytes/diagnostics/cuda.py +++ b/bitsandbytes/diagnostics/cuda.py @@ -6,7 +6,7 @@ import torch from bitsandbytes.cextension import HIP_ENVIRONMENT, get_cuda_bnb_library_path -from bitsandbytes.cuda_specs import CUDASpecs +from bitsandbytes.cuda_specs import CUDASpecs, get_rocm_version from bitsandbytes.diagnostics.utils import print_dedented CUDART_PATH_PREFERRED_ENVVARS = ("CONDA_PREFIX", "LD_LIBRARY_PATH") @@ -137,7 +137,13 @@ def _print_cuda_diagnostics(cuda_specs: CUDASpecs) -> None: def _print_hip_diagnostics(cuda_specs: CUDASpecs) -> None: - print(f"PyTorch settings found: ROCM_VERSION={cuda_specs.cuda_version_string}") + rocm_major, rocm_minor = cuda_specs.cuda_version_tuple + print( + "PyTorch settings found: " + f"ROCm={getattr(torch.version, 'rocm', None) or 'N/A'}, " + f"HIP={getattr(torch.version, 'hip', None) or 'N/A'}, " + f"binary suffix=rocm{rocm_major}{rocm_minor}" + ) rocm_override = os.environ.get("BNB_ROCM_VERSION") if rocm_override: @@ -153,11 +159,10 @@ def _print_hip_diagnostics(cuda_specs: CUDASpecs) -> None: """, ) - hip_major, hip_minor = cuda_specs.cuda_version_tuple - if (hip_major, hip_minor) < (6, 1): + if (rocm_major, rocm_minor) < (6, 3): print_dedented( """ - WARNING: bitsandbytes is fully supported only from ROCm 6.1. + WARNING: bitsandbytes is fully supported only from ROCm 6.3. """, ) @@ -171,7 +176,7 @@ def print_diagnostics(cuda_specs: CUDASpecs) -> None: def print_runtime_diagnostics() -> None: backend = "ROCm" if HIP_ENVIRONMENT else "CUDA" - runtime_version = torch.version.hip if HIP_ENVIRONMENT else torch.version.cuda + runtime_version = get_rocm_version() if HIP_ENVIRONMENT else torch.version.cuda override_var = "BNB_ROCM_VERSION" if HIP_ENVIRONMENT else "BNB_CUDA_VERSION" override_example = "72" if HIP_ENVIRONMENT else "122" diff --git a/bitsandbytes/diagnostics/main.py b/bitsandbytes/diagnostics/main.py index a64925c06..3584472ef 100644 --- a/bitsandbytes/diagnostics/main.py +++ b/bitsandbytes/diagnostics/main.py @@ -58,8 +58,9 @@ def show_environment(): print(f"PyTorch: {torch.__version__}") print(f" CUDA: {torch.version.cuda or 'N/A'}") + print(f" ROCm: {getattr(torch.version, 'rocm', None) or 'N/A'}") print(f" HIP: {torch.version.hip or 'N/A'}") - print(f" XPU: {getattr(torch.version, 'xpu', 'N/A') or 'N/A'}") + print(f" XPU: {getattr(torch.version, 'xpu', None) or 'N/A'}") print("Related packages:") for pkg in _RELATED_PACKAGES: diff --git a/csrc/compat.cuh b/csrc/compat.cuh index f8c307c2a..de745c368 100644 --- a/csrc/compat.cuh +++ b/csrc/compat.cuh @@ -144,9 +144,7 @@ using bnb_bfloat162 = __nv_bfloat162; #if BNB_HIP -#ifndef NO_HIPBLASLT #include -#endif using bnb_blasLt_handle_t = hipblasLtHandle_t; using bnb_blasLt_matmul_desc_t = hipblasLtMatmulDesc_t; diff --git a/csrc/ops.cu b/csrc/ops.cu index 16eed4e81..b67731175 100644 --- a/csrc/ops.cu +++ b/csrc/ops.cu @@ -284,11 +284,6 @@ int igemmlt( bnb_blasLt_handle_t ltHandle, int m, int n, int k, const int8_t* A, const int8_t* B, void* C, float* row_scale, int lda, int ldb, int ldc, bnb_stream_t stream ) { - -#if BNB_HIP && defined(NO_HIPBLASLT) - return ERR_NOT_IMPLEMENTED; -#else - // Calculate C = A^T @ B, in col-major layout. // // Use the IMMA kernels requires: @@ -400,7 +395,6 @@ int igemmlt( printf("error detected"); return has_error; -#endif // NO_HIPBLASLT } int fill_up_to_nearest_multiple(int value, int multiple) { diff --git a/docs/source/errors.mdx b/docs/source/errors.mdx index 987488770..a50611396 100644 --- a/docs/source/errors.mdx +++ b/docs/source/errors.mdx @@ -23,16 +23,19 @@ If this does not work, please open an issue and paste the printed environment if ## Library not found: version mismatch -The library filename encodes the version: `libbitsandbytes_cuda{major}{minor}` for CUDA, `libbitsandbytes_rocm{major}{minor}` for ROCm. bitsandbytes selects which one to load based on what PyTorch reports: +The library filename encodes the version by concatenating major and minor: `libbitsandbytes_cuda{major}{minor}` for CUDA and `libbitsandbytes_rocm{major}{minor}` for ROCm. bitsandbytes selects which one to load based on what PyTorch reports: ```python import torch print(torch.version.cuda) # e.g. "12.8" -> looks for libbitsandbytes_cuda128 -print(torch.version.hip) # e.g. "7.2" -> looks for libbitsandbytes_rocm72 +print(getattr(torch.version, "rocm", None) or torch.version.hip) +# e.g. "7.14" -> looks for libbitsandbytes_rocm714 ``` bitsandbytes will automatically fall back to the closest available pre-compiled version if an exact match is not found, and log a warning. For example, if your PyTorch was built with CUDA 12.9 but bitsandbytes only ships 12.8, it will load 12.8 automatically. +ROCm may also fall back across major release numbers, preferring the newest available lower release and then the oldest newer release. Compatibility is not guaranteed: a binary from another major release may not contain code for your GPU architecture. A cross-major fallback therefore emits a stronger warning. Compile from source or set `BNB_ROCM_VERSION` if the selected binary fails to load or launch kernels. + If you see an error like `No compatible CUDA library found`, it means no compatible pre-compiled library could be found at all. To resolve this: 1. **Compile from source** to produce a library matching your exact toolkit version. See the [installation guide](installation) for instructions. @@ -44,4 +47,4 @@ If you see an error like `No compatible CUDA library found`, it means no compati # Windows (cmd) set BNB_CUDA_VERSION=128 ``` - The value must be digits only, e.g. `128` for CUDA 12.8 or `72` for ROCm 7.2. + Both backends accept dotted versions and compact shortcodes, for example `12.8` or `128` for CUDA and `7.14` or `714` for ROCm. diff --git a/docs/source/installation.mdx b/docs/source/installation.mdx index aa928dac6..e82c25a06 100644 --- a/docs/source/installation.mdx +++ b/docs/source/installation.mdx @@ -176,6 +176,8 @@ pip install bitsandbytes bitsandbytes can be compiled from ROCm 6.3 - ROCm 7.14.0. See the `CMakeLists.txt` for additional options. +CMake uses the ROCm release, rather than the HIP SDK version, to name the native library. It detects the release from the active ROCm installation's `.info/version` file or `rocm-sdk version`. If neither is available, pass the version explicitly as a dotted release such as `-DROCM_VERSION=7.14.0` or its shortcode, `-DROCM_VERSION=714`. The major and minor versions are concatenated in the filename, for example `libbitsandbytes_rocm714.so`. + @@ -220,7 +222,7 @@ export ROCM_PATH="$(rocm-sdk path --root)" export PATH="${ROCM_PATH}/bin:${PATH}" git clone https://github.com/bitsandbytes-foundation/bitsandbytes.git && cd bitsandbytes/ # Use the same GPU architecture selected above. Separate multiple architectures with semicolons. -cmake -G Ninja -DCOMPUTE_BACKEND=hip -DBNB_ROCM_ARCH="gfx1100" -DCMAKE_BUILD_TYPE=Release -DCMAKE_HIP_COMPILER_ROCM_ROOT="${ROCM_PATH}" -S . +cmake -G Ninja -DCOMPUTE_BACKEND=hip -DROCM_VERSION=7.14.0 -DBNB_ROCM_ARCH="gfx1100" -DCMAKE_BUILD_TYPE=Release -DCMAKE_HIP_COMPILER_ROCM_ROOT="${ROCM_PATH}" -S . cmake --build . --config Release pip install . ``` @@ -242,7 +244,7 @@ rocm-sdk init export ROCM_PATH="$(rocm-sdk path --root)" export PATH="${ROCM_PATH}/bin:${PATH}" git clone https://github.com/bitsandbytes-foundation/bitsandbytes.git && cd bitsandbytes/ -cmake -G Ninja -DCOMPUTE_BACKEND=hip -DBNB_ROCM_ARCH="gfx1100" -DCMAKE_BUILD_TYPE=Release -S . +cmake -G Ninja -DCOMPUTE_BACKEND=hip -DROCM_VERSION=7.2.1 -DBNB_ROCM_ARCH="gfx1100" -DCMAKE_BUILD_TYPE=Release -S . cmake --build . --config Release pip install . ``` diff --git a/tests/test_cuda_setup_evaluator.py b/tests/test_cuda_setup_evaluator.py index 56a52736e..81e62ba3b 100644 --- a/tests/test_cuda_setup_evaluator.py +++ b/tests/test_cuda_setup_evaluator.py @@ -2,134 +2,143 @@ from unittest.mock import patch import pytest +import torch from bitsandbytes.cextension import get_cuda_bnb_library_path from bitsandbytes.consts import DYNAMIC_LIBRARY_SUFFIX from bitsandbytes.cuda_specs import CUDASpecs -@pytest.fixture -def cuda120_spec() -> CUDASpecs: - """Simulates torch+cuda12.0 and a representative Ampere-class capability.""" +def specs(version: tuple[int, int]) -> CUDASpecs: return CUDASpecs( - cuda_version_string="120", - highest_compute_capability=(8, 6), - cuda_version_tuple=(12, 0), - ) - - -@pytest.fixture -def rocm70_spec() -> CUDASpecs: - """Simulates torch+rocm7.0.""" - return CUDASpecs( - cuda_version_string="70", + cuda_version_string=f"{version[0]}{version[1]}", highest_compute_capability=(0, 0), - cuda_version_tuple=(7, 0), + cuda_version_tuple=version, ) @pytest.mark.parametrize( - "spec,fake_libs,hip_version,expected_name,expect_warning", + "backend,backend_version,runtime_version,available,expected,warning", [ - # exact match - ( - CUDASpecs(cuda_version_string="124", highest_compute_capability=(8, 6), cuda_version_tuple=(12, 4)), - {(12, 4): Path(f"libbitsandbytes_cuda124{DYNAMIC_LIBRARY_SUFFIX}")}, - None, - f"libbitsandbytes_cuda124{DYNAMIC_LIBRARY_SUFFIX}", - False, - ), - # forward fallback within major: 12.0 -> 12.1 - ( - CUDASpecs(cuda_version_string="120", highest_compute_capability=(8, 6), cuda_version_tuple=(12, 0)), - { - (12, 1): Path(f"libbitsandbytes_cuda121{DYNAMIC_LIBRARY_SUFFIX}"), - (12, 4): Path(f"libbitsandbytes_cuda124{DYNAMIC_LIBRARY_SUFFIX}"), - }, - None, - f"libbitsandbytes_cuda121{DYNAMIC_LIBRARY_SUFFIX}", - True, - ), - # backward fallback: 12.9 -> 12.8 - ( - CUDASpecs(cuda_version_string="129", highest_compute_capability=(8, 9), cuda_version_tuple=(12, 9)), - { - (12, 4): Path(f"libbitsandbytes_cuda124{DYNAMIC_LIBRARY_SUFFIX}"), - (12, 8): Path(f"libbitsandbytes_cuda128{DYNAMIC_LIBRARY_SUFFIX}"), - }, - None, - f"libbitsandbytes_cuda128{DYNAMIC_LIBRARY_SUFFIX}", - True, - ), - # ROCm double-digit minor: 7.13 -> 7.2 - ( - CUDASpecs(cuda_version_string="713", highest_compute_capability=(0, 0), cuda_version_tuple=(7, 13)), - {(7, 2): Path(f"libbitsandbytes_rocm72{DYNAMIC_LIBRARY_SUFFIX}")}, - "7.13.0", - f"libbitsandbytes_rocm72{DYNAMIC_LIBRARY_SUFFIX}", - True, - ), - # no same-major match: 11.8 with only 12.x -> non-existent exact path, no warning - ( - CUDASpecs(cuda_version_string="118", highest_compute_capability=(7, 5), cuda_version_tuple=(11, 8)), - {(12, 1): Path("libbitsandbytes_cuda121.so"), (12, 4): Path("libbitsandbytes_cuda124.so")}, - None, - f"libbitsandbytes_cuda118{DYNAMIC_LIBRARY_SUFFIX}", - False, - ), - # no libs at all -> non-existent exact path, no warning - ( - CUDASpecs(cuda_version_string="129", highest_compute_capability=(8, 9), cuda_version_tuple=(12, 9)), - {}, - None, - f"libbitsandbytes_cuda129{DYNAMIC_LIBRARY_SUFFIX}", - False, - ), + # Exact match. + ("cuda", "12.4", (12, 4), [(12, 4)], (12, 4), False), + # Same-major fallback to the newest older binary. + ("cuda", "12.9", (12, 9), [(12, 4), (12, 8)], (12, 8), True), + # Same-major fallback to the oldest newer binary. + ("cuda", "12.0", (12, 0), [(12, 1), (12, 4)], (12, 1), True), + # CUDA does not fall back across major versions. + ("cuda", "11.8", (11, 8), [(12, 1)], None, False), + # ROCm same-major fallback with a double-digit minor. + ("hip", "7.13.0", (7, 13), [(7, 2), (7, 14)], (7, 2), True), + # ROCm same-major fallback to the newest older binary. + ("hip", "7.9.0", (7, 9), [(7, 2), (7, 14)], (7, 2), True), + # ROCm/HIP version-line divergence with an older cross-major fallback. + ("hip", "7.16.0", (12, 1), [(8, 0), (7, 14)], (8, 0), True), + # ROCm cross-major fallback to the newest older binary. + ("hip", "8.0.0", (8, 0), [(7, 14)], (7, 14), True), + # ROCm cross-major fallback to the oldest newer binary. + ("hip", "6.4.0", (6, 4), [(7, 0)], (7, 0), True), + # No packaged libraries returns the requested path without a warning. + ("hip", "7.14.0", (7, 14), [], None, False), ], ) -def test_version_selection(monkeypatch, caplog, spec, fake_libs, hip_version, expected_name, expect_warning): - """Library selection: exact match, fallback, no-same-major, no-libs.""" +def test_library_selection( + backend, + backend_version, + runtime_version, + available, + expected, + warning, + monkeypatch, + caplog, +): monkeypatch.delenv("BNB_CUDA_VERSION", raising=False) monkeypatch.delenv("BNB_ROCM_VERSION", raising=False) - is_hip = spec.cuda_version_tuple[0] < 10 + other_backend = "cuda" if backend == "hip" else "hip" + prefix = "rocm" if backend == "hip" else "cuda" + paths = { + version: Path(f"libbitsandbytes_{prefix}{version[0]}{version[1]}{DYNAMIC_LIBRARY_SUFFIX}") + for version in available + } with ( - patch("torch.version.hip", hip_version if is_hip else None), - patch("bitsandbytes.cextension._find_cuda_libs", return_value=fake_libs), + patch.object(torch.version, backend, backend_version), + patch.object(torch.version, other_backend, None), + patch("bitsandbytes.cextension._find_cuda_libs", return_value=paths), + caplog.at_level("WARNING"), ): - with caplog.at_level("WARNING"): - result = get_cuda_bnb_library_path(spec) - assert result.name == expected_name - if expect_warning: - assert caplog.text + result = get_cuda_bnb_library_path(specs(runtime_version)) + + if expected is None: + tag = f"{runtime_version[0]}{runtime_version[1]}" + assert result.name == f"libbitsandbytes_{prefix}{tag}{DYNAMIC_LIBRARY_SUFFIX}" else: - assert not caplog.text + assert result == paths[expected] + assert bool(caplog.text) is warning -def test_override(monkeypatch, cuda120_spec, caplog): - """BNB_CUDA_VERSION overrides path selection.""" - monkeypatch.setenv("BNB_CUDA_VERSION", "110") - with patch("bitsandbytes.cextension._find_cuda_libs", return_value={}): - with caplog.at_level("WARNING"): - result = get_cuda_bnb_library_path(cuda120_spec) - assert result.stem == "libbitsandbytes_cuda110" - assert "BNB_CUDA_VERSION" in caplog.text +@pytest.mark.parametrize( + "backend,backend_version,version,override,expected_stem", + [ + ("hip", "7.0.0", (7, 0), "72", "libbitsandbytes_rocm72"), + ("hip", "7.0.0", (7, 0), "7.2", "libbitsandbytes_rocm72"), + ("cuda", "12.0", (12, 0), "128", "libbitsandbytes_cuda128"), + ("cuda", "12.0", (12, 0), "12.8", "libbitsandbytes_cuda128"), + ("cuda", "12.0", (12, 0), "12.8.1", "libbitsandbytes_cuda128"), + ], +) +def test_override_formats(monkeypatch, backend, backend_version, version, override, expected_stem): + other_backend = "cuda" if backend == "hip" else "hip" + override_var = "BNB_ROCM_VERSION" if backend == "hip" else "BNB_CUDA_VERSION" + other_var = "BNB_CUDA_VERSION" if backend == "hip" else "BNB_ROCM_VERSION" + monkeypatch.setenv(override_var, override) + monkeypatch.delenv(other_var, raising=False) + with ( + patch.object(torch.version, backend, backend_version), + patch.object(torch.version, other_backend, None), + ): + assert get_cuda_bnb_library_path(specs(version)).stem == expected_stem -def test_rocm_override(monkeypatch, rocm70_spec, caplog): - """BNB_ROCM_VERSION overrides path selection.""" - monkeypatch.setenv("BNB_ROCM_VERSION", "72") +@pytest.mark.parametrize( + "backend,backend_version,version", + [ + ("cuda", "12.0", (12, 0)), + ("hip", "7.2.0", (7, 2)), + ], +) +def test_opposite_backend_override_warns(monkeypatch, caplog, backend, backend_version, version): + other_backend = "hip" if backend == "cuda" else "cuda" + correct_var = "BNB_CUDA_VERSION" if backend == "cuda" else "BNB_ROCM_VERSION" + wrong_var = "BNB_ROCM_VERSION" if backend == "cuda" else "BNB_CUDA_VERSION" + monkeypatch.setenv(wrong_var, "72") + monkeypatch.delenv(correct_var, raising=False) with ( - patch("torch.version.hip", "7.0.0"), + patch.object(torch.version, backend, backend_version), + patch.object(torch.version, other_backend, None), patch("bitsandbytes.cextension._find_cuda_libs", return_value={}), + caplog.at_level("WARNING"), ): - with caplog.at_level("WARNING"): - result = get_cuda_bnb_library_path(rocm70_spec) - assert result.stem == "libbitsandbytes_rocm72" - assert "BNB_ROCM_VERSION" in caplog.text + get_cuda_bnb_library_path(specs(version)) + assert f"{wrong_var} is ignored" in caplog.text + assert f"use {correct_var} instead" in caplog.text -def test_override_invalid_format(monkeypatch, cuda120_spec): - """Override value must be digits only (e.g. '124'), not dotted or alphanumeric.""" - monkeypatch.setenv("BNB_CUDA_VERSION", "12.4") - with pytest.raises(RuntimeError, match="digits only"): - get_cuda_bnb_library_path(cuda120_spec) +@pytest.mark.parametrize( + "backend,backend_version,version", + [ + ("hip", "7.0.0", (7, 0)), + ("cuda", "12.0", (12, 0)), + ], +) +def test_invalid_override(monkeypatch, backend, backend_version, version): + other_backend = "cuda" if backend == "hip" else "hip" + override_var = "BNB_ROCM_VERSION" if backend == "hip" else "BNB_CUDA_VERSION" + other_var = "BNB_CUDA_VERSION" if backend == "hip" else "BNB_ROCM_VERSION" + monkeypatch.setenv(override_var, "not-a-version") + monkeypatch.delenv(other_var, raising=False) + with ( + patch.object(torch.version, backend, backend_version), + patch.object(torch.version, other_backend, None), + pytest.raises(RuntimeError, match="dotted version"), + ): + get_cuda_bnb_library_path(specs(version)) From 237702f492bb0511bec59bb7145f342bf2186d83 Mon Sep 17 00:00:00 2001 From: sstamenk Date: Fri, 14 Aug 2026 15:20:55 +0000 Subject: [PATCH 2/5] Clarify native library fallback order Document the current same-major and ROCm cross-major selection priority more explicitly. --- bitsandbytes/cextension.py | 13 ++++++------- 1 file changed, 6 insertions(+), 7 deletions(-) diff --git a/bitsandbytes/cextension.py b/bitsandbytes/cextension.py index efc92363f..b72762861 100644 --- a/bitsandbytes/cextension.py +++ b/bitsandbytes/cextension.py @@ -25,13 +25,12 @@ def get_cuda_bnb_library_path(cuda_specs: CUDASpecs) -> Path: When no override is set, selects from packaged libraries using the following priority: 1. Exact version match. - 2. Same-major fallback, preferring the newest older binary, then the - lowest newer binary. - 3. For ROCm only, cross-major fallback using the same older-first policy. - CUDA never falls back across major versions. - A warning is logged when falling back. Overrides select the requested filename - directly. The returned path is not guaranteed to exist when no packaged libs - are found, or when an override names an absent version. + 2. Highest packaged version <= runtime version, same major. + 3. Lowest packaged version > runtime version, same major. + 4. For ROCm only, repeat the same older-first selection across major versions. + CUDA does not fall back across major versions. A warning is logged when falling back. + Overrides select the requested filename directly. The returned path is not guaranteed + to exist when no packaged libraries are found or an override names an absent version. """ is_hip = bool(torch.version.hip) prefix = "rocm" if is_hip else "cuda" From bf8b7a81d6c4381540caacd24e70825ee0bd9e68 Mon Sep 17 00:00:00 2001 From: sstamenk Date: Fri, 14 Aug 2026 15:25:32 +0000 Subject: [PATCH 3/5] Add examples to library fallback docs Illustrate older and newer same-major fallback ordering in the loader docstring. --- bitsandbytes/cextension.py | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/bitsandbytes/cextension.py b/bitsandbytes/cextension.py index b72762861..9421470f0 100644 --- a/bitsandbytes/cextension.py +++ b/bitsandbytes/cextension.py @@ -25,8 +25,8 @@ def get_cuda_bnb_library_path(cuda_specs: CUDASpecs) -> Path: When no override is set, selects from packaged libraries using the following priority: 1. Exact version match. - 2. Highest packaged version <= runtime version, same major. - 3. Lowest packaged version > runtime version, same major. + 2. Highest packaged version <= runtime version, same major (e.g. runtime 12.9, packaged 12.8). + 3. Lowest packaged version > runtime version, same major (e.g. runtime 12.0, packaged 12.1). 4. For ROCm only, repeat the same older-first selection across major versions. CUDA does not fall back across major versions. A warning is logged when falling back. Overrides select the requested filename directly. The returned path is not guaranteed From f9634e12c8325336926f84415a1c0b7421df31ac Mon Sep 17 00:00:00 2001 From: sstamenk Date: Fri, 14 Aug 2026 15:48:54 +0000 Subject: [PATCH 4/5] Fix ROCm artifact collection Only pass shared libraries that exist to the packaging copy step. --- .github/scripts/build-rocm.sh | 8 +++++--- 1 file changed, 5 insertions(+), 3 deletions(-) diff --git a/.github/scripts/build-rocm.sh b/.github/scripts/build-rocm.sh index c4981616e..5d45dec99 100644 --- a/.github/scripts/build-rocm.sh +++ b/.github/scripts/build-rocm.sh @@ -91,9 +91,11 @@ fi output_dir="output/${RUNNER_OS}/X64" mkdir -p "${output_dir}" -shopt -s nullglob -libraries=(bitsandbytes/libbitsandbytes_rocm${rocm_version_tag}.{so,dylib,dll}) -shopt -u nullglob +libraries=() +for extension in so dylib dll; do + library="bitsandbytes/libbitsandbytes_rocm${rocm_version_tag}.${extension}" + [ -f "${library}" ] && libraries+=("${library}") +done if [ "${#libraries[@]}" -eq 0 ]; then expected_pattern="bitsandbytes/libbitsandbytes_rocm${rocm_version_tag}.{so,dylib,dll}" From e199d158ee72acd6dc23497d2b0ef9432dfa254c Mon Sep 17 00:00:00 2001 From: sstamenk Date: Wed, 26 Aug 2026 15:05:50 +0000 Subject: [PATCH 5/5] Drop ROCm version logic and harden HIP path --- .github/scripts/build-rocm.sh | 44 +++++------ CMakeLists.txt | 115 ++++++++++------------------- bitsandbytes/cextension.py | 16 ++-- bitsandbytes/cuda_specs.py | 25 +------ bitsandbytes/diagnostics/cuda.py | 34 +++++---- csrc/compat.cuh | 2 + csrc/ops.cu | 6 ++ docs/source/errors.mdx | 11 +-- docs/source/installation.mdx | 10 +-- tests/test_cuda_setup_evaluator.py | 92 ++++++++++++----------- 10 files changed, 160 insertions(+), 195 deletions(-) diff --git a/.github/scripts/build-rocm.sh b/.github/scripts/build-rocm.sh index 5d45dec99..b6eff1250 100644 --- a/.github/scripts/build-rocm.sh +++ b/.github/scripts/build-rocm.sh @@ -1,27 +1,25 @@ #!/bin/bash set -xeuo pipefail -rocm_version_at_least() { - local required_major required_minor - - IFS=. read -r required_major required_minor <<< "$1" - ((rocm_version_major > required_major || - (rocm_version_major == required_major && rocm_version_minor >= required_minor))) -} - if [[ "${RUNNER_OS:-}" != "Linux" && "${RUNNER_OS:-}" != "Windows" ]]; then - echo "Invalid RUNNER_OS '${RUNNER_OS:-}'; expected Linux or Windows." >&2 + echo "Invalid RUNNER_OS '${RUNNER_OS:-}'; expected Linux or Windows." >&2 exit 1 fi if [[ ! "${ROCM_VERSION:-}" =~ ^([0-9]+)\.([0-9]+)(\.[0-9]+)?$ ]]; then - echo "Invalid ROCM_VERSION '${ROCM_VERSION:-}'; expected a dotted ROCm release." >&2 + echo "Invalid ROCM_VERSION '${ROCM_VERSION:-}'; expected a dotted ROCm release such as 7.14." >&2 exit 1 fi rocm_version_major="$((10#${BASH_REMATCH[1]}))" rocm_version_minor="$((10#${BASH_REMATCH[2]}))" -rocm_version_tag="${rocm_version_major}${rocm_version_minor}" + +rocm_version_at_least() { + local required_major required_minor + IFS=. read -r required_major required_minor <<< "$1" + ((rocm_version_major > required_major || + (rocm_version_major == required_major && rocm_version_minor >= required_minor))) +} bnb_rocm_arch="gfx90a;gfx942;gfx1100;gfx1101;gfx1102;gfx1103" @@ -35,9 +33,14 @@ if rocm_version_at_least "7.0"; then bnb_rocm_arch="${bnb_rocm_arch};gfx950" fi -# ROCm 7.14+ - Add CDNA1, CDNA5, and RDNA2 targets. +# ROCm 7.14+ - Add CDNA1 and RDNA2 targets. if rocm_version_at_least "7.14"; then - bnb_rocm_arch="${bnb_rocm_arch};gfx908;gfx1030;gfx1031;gfx1032;gfx1033;gfx1034;gfx1035;gfx1036;gfx1250" + bnb_rocm_arch="${bnb_rocm_arch};gfx908;gfx1030;gfx1031;gfx1032;gfx1033;gfx1034;gfx1035;gfx1036" +fi + +# ROCm 7.14+ - Add CDNA5 (gfx1250). +if rocm_version_at_least "7.14"; then + bnb_rocm_arch="${bnb_rocm_arch};gfx1250" fi if [ "${RUNNER_OS}" == "Linux" ]; then @@ -50,7 +53,7 @@ if [ "${RUNNER_OS}" == "Linux" ]; then docker run --rm -i \ -w /src -v "$PWD:/src" "$image" sh -c \ "pip install cmake==3.31.6 \ - && cmake -DCOMPUTE_BACKEND=hip -DROCM_VERSION=\"${ROCM_VERSION}\" -DCMAKE_BUILD_TYPE=MinSizeRel -DCMAKE_HIP_FLAGS=\"--offload-compress\" -DBNB_ROCM_ARCH=\"${bnb_rocm_arch}\" . \ + && cmake -DCOMPUTE_BACKEND=hip -DCMAKE_BUILD_TYPE=MinSizeRel -DCMAKE_HIP_FLAGS=\"--offload-compress\" -DBNB_ROCM_ARCH=\"${bnb_rocm_arch}\" . \ && cmake --build . --parallel" else bnb_rocm_arch="gfx1100;gfx1101;gfx1102;gfx1150;gfx1151;gfx1200;gfx1201" @@ -79,7 +82,6 @@ else cmake -G Ninja \ -DCOMPUTE_BACKEND=hip \ - -DROCM_VERSION="${ROCM_VERSION}" \ -DBNB_ROCM_ARCH="${bnb_rocm_arch}" \ -DCMAKE_BUILD_TYPE=MinSizeRel \ -DCMAKE_HIP_FLAGS="--offload-compress" \ @@ -90,17 +92,17 @@ fi output_dir="output/${RUNNER_OS}/X64" mkdir -p "${output_dir}" - libraries=() +shopt -s nullglob for extension in so dylib dll; do - library="bitsandbytes/libbitsandbytes_rocm${rocm_version_tag}.${extension}" - [ -f "${library}" ] && libraries+=("${library}") + for library in bitsandbytes/libbitsandbytes_rocm*."${extension}"; do + libraries+=("${library}") + done done +shopt -u nullglob if [ "${#libraries[@]}" -eq 0 ]; then - expected_pattern="bitsandbytes/libbitsandbytes_rocm${rocm_version_tag}.{so,dylib,dll}" - echo "Expected ROCm ${ROCM_VERSION} library was not built: ${expected_pattern}" >&2 + echo "No ROCm-backend library was built (expected bitsandbytes/libbitsandbytes_rocm*.{so,dylib,dll})." >&2 exit 1 fi - cp "${libraries[@]}" "${output_dir}/" diff --git a/CMakeLists.txt b/CMakeLists.txt index 9e6353744..9a46467fb 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -10,9 +10,10 @@ # Separate by semicolons, i.e. `-DCOMPUTE_CAPABILITY=89;90;100;120` # Check your compute capability here: https://developer.nvidia.com/cuda-gpus # - PTXAS_VERBOSE: Pass the `-v` option to the PTX Assembler -# - ROCM_VERSION: Override the ROCm release used in the output library name. Accepts a -# dotted release or shortcode. For example, `7.14.0` and `714` -# both produce libbitsandbytes_rocm714.so. +# - HIP_VERSION: Override the HIP version used in the ROCm-backend output library name. +# Accepts a dotted version or compact shortcode. For example, +# `-DHIP_VERSION=7.14` and `-DHIP_VERSION=714` both produce +# libbitsandbytes_rocm714.so. cmake_minimum_required(VERSION 3.22.1) # On Windows with HIP backend, auto-detect compilers from ROCM_PATH before project() @@ -274,6 +275,11 @@ elseif(BUILD_HIP) set(CMAKE_HIP_ARCHITECTURES "gfx90a;gfx942;gfx1100;gfx1101;gfx1102;gfx1103;gfx1150;gfx1151;gfx1152;gfx1153;gfx1200;gfx1201") endif() + find_program(HIPCONFIG_EXECUTABLE hipconfig) + if(NOT HIPCONFIG_EXECUTABLE) + message(FATAL_ERROR "hipconfig was not found on PATH; a ROCm/HIP SDK is required for HIP builds") + endif() + enable_language(HIP) message(STATUS "HIP Compiler: ${CMAKE_HIP_COMPILER}") message(STATUS "HIP Targets: ${CMAKE_HIP_ARCHITECTURES}") @@ -286,75 +292,32 @@ elseif(BUILD_HIP) string(APPEND BNB_OUTPUT_NAME "_rocm") - # HIP and ROCm releases can diverge, so use the ROCm release for the filename. - set(ROCM_VERSION "" CACHE STRING "ROCm release used in the output library name") - set(_ROCM_VERSION_RAW "${ROCM_VERSION}") - set(_ROCM_VERSION_SOURCE "-DROCM_VERSION") - - if(NOT _ROCM_VERSION_RAW) - if(NOT ROCM_PATH) - if(DEFINED ENV{ROCM_PATH}) - file(TO_CMAKE_PATH "$ENV{ROCM_PATH}" ROCM_PATH) - elseif(CMAKE_HIP_COMPILER_ROCM_ROOT) - set(ROCM_PATH "${CMAKE_HIP_COMPILER_ROCM_ROOT}") - elseif(WIN32) - message(FATAL_ERROR "ROCM_PATH must be set for HIP builds on Windows") - else() - message(WARNING "ROCM_PATH is not set; falling back to /opt/rocm") - set(ROCM_PATH "/opt/rocm") - endif() - endif() - - foreach(_VERSION_FILE "${ROCM_PATH}/.info/version" "${ROCM_PATH}/core/.info/version") - if(NOT _ROCM_VERSION_RAW AND EXISTS "${_VERSION_FILE}") - file(READ "${_VERSION_FILE}" _ROCM_VERSION_RAW) - set(_ROCM_VERSION_SOURCE "${_VERSION_FILE}") - endif() - endforeach() - - if(NOT _ROCM_VERSION_RAW) - execute_process( - COMMAND rocm-sdk version - OUTPUT_VARIABLE _ROCM_VERSION_RAW - ERROR_QUIET - OUTPUT_STRIP_TRAILING_WHITESPACE - ) - set(_ROCM_VERSION_SOURCE "rocm-sdk version") - endif() + execute_process( + COMMAND "${HIPCONFIG_EXECUTABLE}" --version + OUTPUT_VARIABLE HIPCONFIG_VERSION + OUTPUT_STRIP_TRAILING_WHITESPACE + COMMAND_ERROR_IS_FATAL ANY + ) - if(NOT _ROCM_VERSION_RAW) - message(FATAL_ERROR - "Could not determine the ROCm version from ${ROCM_PATH}/.info/version, " - "${ROCM_PATH}/core/.info/version, or 'rocm-sdk version'. " - "Set ROCM_PATH correctly or pass -DROCM_VERSION=." - ) - endif() + if(NOT HIPCONFIG_VERSION MATCHES "([0-9]+)\\.([0-9]+)") + message(FATAL_ERROR "Could not parse HIP version: ${HIPCONFIG_VERSION}") endif() + set(DETECTED_HIP_VERSION_TAG "${CMAKE_MATCH_1}${CMAKE_MATCH_2}") - string(STRIP "${_ROCM_VERSION_RAW}" _ROCM_VERSION_RAW) - if(_ROCM_VERSION_RAW MATCHES "^([0-9]+)\\.([0-9]+)") - set(_ROCM_VERSION_MAJOR "${CMAKE_MATCH_1}") - set(_ROCM_VERSION_MINOR "${CMAKE_MATCH_2}") - elseif(_ROCM_VERSION_RAW MATCHES "^([6-9])([0-9]+)$") - set(_ROCM_VERSION_MAJOR "${CMAKE_MATCH_1}") - set(_ROCM_VERSION_MINOR "${CMAKE_MATCH_2}") - elseif(_ROCM_VERSION_RAW MATCHES "^([1-5][0-9])([0-9]+)$") - set(_ROCM_VERSION_MAJOR "${CMAKE_MATCH_1}") - set(_ROCM_VERSION_MINOR "${CMAKE_MATCH_2}") + set(HIP_VERSION "${DETECTED_HIP_VERSION_TAG}" CACHE STRING + "HIP version used in the ROCm-backend library name" + ) + if(HIP_VERSION MATCHES "^([0-9]+)\\.([0-9]+)(\\.[0-9]+)?$") + set(HIP_VERSION_TAG "${CMAKE_MATCH_1}${CMAKE_MATCH_2}") + elseif(HIP_VERSION MATCHES "^[0-9]+$") + set(HIP_VERSION_TAG "${HIP_VERSION}") else() - message(FATAL_ERROR - "Could not parse ROCm version '${_ROCM_VERSION_RAW}'. " - "Pass -DROCM_VERSION=7.14.0 (or 714)." - ) + message(FATAL_ERROR "HIP_VERSION must be dotted or compact, such as 7.14 or 714") endif() - set(ROCM_VERSION "${_ROCM_VERSION_MAJOR}.${_ROCM_VERSION_MINOR}") - set(_ROCM_VERSION_TAG "${_ROCM_VERSION_MAJOR}${_ROCM_VERSION_MINOR}") - message(STATUS - "ROCm Release: ${_ROCM_VERSION_MAJOR}.${_ROCM_VERSION_MINOR}; " - "library suffix: rocm${_ROCM_VERSION_TAG} (from ${_ROCM_VERSION_SOURCE})" - ) - string(APPEND BNB_OUTPUT_NAME "${_ROCM_VERSION_TAG}") + message(STATUS "HIP ${HIPCONFIG_VERSION}; library suffix: rocm${HIP_VERSION_TAG}") + + string(APPEND BNB_OUTPUT_NAME "${HIP_VERSION_TAG}") add_compile_definitions(__HIP_PLATFORM_AMD__) add_compile_definitions(__HIP_PLATFORM_HCC__) add_compile_definitions(BUILD_HIP) @@ -475,16 +438,16 @@ if(BUILD_CUDA) ) endif() if(BUILD_HIP) - # ROCM_PATH was resolved during version detection; retain it for package discovery. - if(NOT ROCM_PATH) - if(DEFINED ENV{ROCM_PATH}) - file(TO_CMAKE_PATH "$ENV{ROCM_PATH}" ROCM_PATH) - elseif(WIN32) - message(FATAL_ERROR "ROCM_PATH must be set for ROCm builds on Windows") - else() - message(WARNING "ROCM_PATH is not set; falling back to /opt/rocm") - set(ROCM_PATH "/opt/rocm") - endif() + # Determine ROCM_PATH from an existing CMake value or environment variable. + if(ROCM_PATH) + file(TO_CMAKE_PATH "${ROCM_PATH}" ROCM_PATH) + elseif(DEFINED ENV{ROCM_PATH} AND NOT "$ENV{ROCM_PATH}" STREQUAL "") + file(TO_CMAKE_PATH "$ENV{ROCM_PATH}" ROCM_PATH) + elseif(WIN32) + message(FATAL_ERROR "ROCM_PATH must be set for HIP builds on Windows") + else() + message(WARNING "ROCM_PATH is not set; falling back to /opt/rocm") + set(ROCM_PATH /opt/rocm) endif() list(APPEND CMAKE_PREFIX_PATH ${ROCM_PATH}) macro(find_package_and_print_version PACKAGE_NAME) diff --git a/bitsandbytes/cextension.py b/bitsandbytes/cextension.py index 9421470f0..37935d7d6 100644 --- a/bitsandbytes/cextension.py +++ b/bitsandbytes/cextension.py @@ -27,8 +27,10 @@ def get_cuda_bnb_library_path(cuda_specs: CUDASpecs) -> Path: 1. Exact version match. 2. Highest packaged version <= runtime version, same major (e.g. runtime 12.9, packaged 12.8). 3. Lowest packaged version > runtime version, same major (e.g. runtime 12.0, packaged 12.1). - 4. For ROCm only, repeat the same older-first selection across major versions. - CUDA does not fall back across major versions. A warning is logged when falling back. + 4. For the ROCm backend only, repeat the same older-first selection across HIP major versions. + ROCm-backend selection uses the version tuple reported by ``torch.version.hip``; it does + not infer a ROCm release from ``torch.version.rocm``. CUDA does not fall back across major + versions. A warning is logged when falling back. Overrides select the requested filename directly. The returned path is not guaranteed to exist when no packaged libraries are found or an override names an absent version. """ @@ -94,16 +96,16 @@ def get_cuda_bnb_library_path(cuda_specs: CUDASpecs) -> Path: if cross_major: logger.warning( - f"No prebuilt binary for ROCm {runtime_version[0]}.{runtime_version[1]}, loading " - f"ROCm {selected[0]}.{selected[1]} across major releases. This binary may be incompatible " + f"No prebuilt ROCm-backend binary for HIP {runtime_version[0]}.{runtime_version[1]}, loading " + f"HIP {selected[0]}.{selected[1]} across major versions. This binary may be incompatible " "or may not contain code for your GPU architecture. " f"Set {override_var} to override or compile from source." ) else: logger.warning( - f"No prebuilt binary for {'ROCm' if is_hip else 'CUDA'} " + f"No prebuilt binary for {'ROCm-backend HIP' if is_hip else 'CUDA'} " f"{runtime_version[0]}.{runtime_version[1]}, loading " - f"{'ROCm' if is_hip else 'CUDA'} {selected[0]}.{selected[1]} instead. " + f"{'HIP' if is_hip else 'CUDA'} {selected[0]}.{selected[1]} instead. " f"Set {override_var} to override." ) return available[selected] @@ -325,7 +327,7 @@ def _format_lib_error_message( "You have two options:\n" "1. COMPILE FROM SOURCE as mentioned here:\n" " https://huggingface.co/docs/bitsandbytes/main/en/installation?backend=AMD+ROCm#amd-gpu\n" - "2. Use BNB_ROCM_VERSION to specify a DIFFERENT ROCm version from the detected one, matching the version the library was built with.\n\n" + "2. Use BNB_ROCM_VERSION to select a DIFFERENT HIP-version suffix for the ROCm-backend binary, matching the version the library was built with.\n\n" ) ) diff --git a/bitsandbytes/cuda_specs.py b/bitsandbytes/cuda_specs.py index 4811af1f5..25ce3cd1e 100644 --- a/bitsandbytes/cuda_specs.py +++ b/bitsandbytes/cuda_specs.py @@ -24,32 +24,15 @@ def get_compute_capabilities() -> list[tuple[int, int]]: return sorted(torch.cuda.get_device_capability(torch.cuda.device(i)) for i in range(torch.cuda.device_count())) -@lru_cache(None) -def get_rocm_version() -> Optional[str]: - """Get the ROCm release used to build PyTorch, with a legacy HIP fallback.""" - version = getattr(torch.version, "rocm", None) - if version is not None: - return version - - version = getattr(torch.version, "hip", None) - logging.getLogger(__name__).warning( - "torch.version.rocm is unavailable; falling back to legacy torch.version.hip=%s for ROCm library selection. " - "HIP and ROCm versions may differ.", - version, - ) - return version - - @lru_cache(None) def get_cuda_version_tuple() -> Optional[tuple[int, int]]: - """Get the CUDA or ROCm release as a tuple of (major, minor).""" + """Get CUDA/HIP version as a tuple of (major, minor).""" try: if torch.version.cuda: version_str = torch.version.cuda + elif torch.version.hip: + version_str = torch.version.hip else: - version_str = get_rocm_version() - - if not version_str: return None parts = version_str.split(".") @@ -61,7 +44,7 @@ def get_cuda_version_tuple() -> Optional[tuple[int, int]]: def get_cuda_version_string() -> Optional[str]: - """Get the compact CUDA/ROCm version string retained for API compatibility.""" + """Get CUDA/HIP version as a string.""" version_tuple = get_cuda_version_tuple() if version_tuple is None: return None diff --git a/bitsandbytes/diagnostics/cuda.py b/bitsandbytes/diagnostics/cuda.py index ef436c954..760166129 100644 --- a/bitsandbytes/diagnostics/cuda.py +++ b/bitsandbytes/diagnostics/cuda.py @@ -6,7 +6,7 @@ import torch from bitsandbytes.cextension import HIP_ENVIRONMENT, get_cuda_bnb_library_path -from bitsandbytes.cuda_specs import CUDASpecs, get_rocm_version +from bitsandbytes.cuda_specs import CUDASpecs from bitsandbytes.diagnostics.utils import print_dedented CUDART_PATH_PREFERRED_ENVVARS = ("CONDA_PREFIX", "LD_LIBRARY_PATH") @@ -137,13 +137,7 @@ def _print_cuda_diagnostics(cuda_specs: CUDASpecs) -> None: def _print_hip_diagnostics(cuda_specs: CUDASpecs) -> None: - rocm_major, rocm_minor = cuda_specs.cuda_version_tuple - print( - "PyTorch settings found: " - f"ROCm={getattr(torch.version, 'rocm', None) or 'N/A'}, " - f"HIP={getattr(torch.version, 'hip', None) or 'N/A'}, " - f"binary suffix=rocm{rocm_major}{rocm_minor}" - ) + print(f"PyTorch settings found: HIP_VERSION={cuda_specs.cuda_version_string}") rocm_override = os.environ.get("BNB_ROCM_VERSION") if rocm_override: @@ -155,16 +149,24 @@ def _print_hip_diagnostics(cuda_specs: CUDASpecs) -> None: f""" No compatible ROCm library found (tried: {binary_path.name}). You may need to compile from source: https://huggingface.co/docs/bitsandbytes/main/en/installation#rocm-compile - Use BNB_ROCM_VERSION to force a specific version if needed. + Use BNB_ROCM_VERSION to force a specific HIP-version suffix if needed. """, ) - if (rocm_major, rocm_minor) < (6, 3): - print_dedented( - """ - WARNING: bitsandbytes is fully supported only from ROCm 6.3. - """, - ) + rocm_version = getattr(torch.version, "rocm", None) + if rocm_version is not None: + try: + rocm_major, rocm_minor = map(int, rocm_version.split(".")[:2]) + except (ValueError, IndexError): + pass + else: + if (rocm_major, rocm_minor) < (6, 4): + print_dedented( + """ + WARNING: ROCm 6.4 or newer is required when compiling bitsandbytes from source. + Current prebuilt Linux binaries begin at ROCm 6.4.4. + """, + ) def print_diagnostics(cuda_specs: CUDASpecs) -> None: @@ -176,7 +178,7 @@ def print_diagnostics(cuda_specs: CUDASpecs) -> None: def print_runtime_diagnostics() -> None: backend = "ROCm" if HIP_ENVIRONMENT else "CUDA" - runtime_version = get_rocm_version() if HIP_ENVIRONMENT else torch.version.cuda + runtime_version = torch.version.hip if HIP_ENVIRONMENT else torch.version.cuda override_var = "BNB_ROCM_VERSION" if HIP_ENVIRONMENT else "BNB_CUDA_VERSION" override_example = "72" if HIP_ENVIRONMENT else "122" diff --git a/csrc/compat.cuh b/csrc/compat.cuh index de745c368..f8c307c2a 100644 --- a/csrc/compat.cuh +++ b/csrc/compat.cuh @@ -144,7 +144,9 @@ using bnb_bfloat162 = __nv_bfloat162; #if BNB_HIP +#ifndef NO_HIPBLASLT #include +#endif using bnb_blasLt_handle_t = hipblasLtHandle_t; using bnb_blasLt_matmul_desc_t = hipblasLtMatmulDesc_t; diff --git a/csrc/ops.cu b/csrc/ops.cu index b67731175..16eed4e81 100644 --- a/csrc/ops.cu +++ b/csrc/ops.cu @@ -284,6 +284,11 @@ int igemmlt( bnb_blasLt_handle_t ltHandle, int m, int n, int k, const int8_t* A, const int8_t* B, void* C, float* row_scale, int lda, int ldb, int ldc, bnb_stream_t stream ) { + +#if BNB_HIP && defined(NO_HIPBLASLT) + return ERR_NOT_IMPLEMENTED; +#else + // Calculate C = A^T @ B, in col-major layout. // // Use the IMMA kernels requires: @@ -395,6 +400,7 @@ int igemmlt( printf("error detected"); return has_error; +#endif // NO_HIPBLASLT } int fill_up_to_nearest_multiple(int value, int multiple) { diff --git a/docs/source/errors.mdx b/docs/source/errors.mdx index a50611396..20ae8ac8c 100644 --- a/docs/source/errors.mdx +++ b/docs/source/errors.mdx @@ -23,18 +23,19 @@ If this does not work, please open an issue and paste the printed environment if ## Library not found: version mismatch -The library filename encodes the version by concatenating major and minor: `libbitsandbytes_cuda{major}{minor}` for CUDA and `libbitsandbytes_rocm{major}{minor}` for ROCm. bitsandbytes selects which one to load based on what PyTorch reports: +The library filename encodes the CUDA or HIP version by concatenating major and minor: `libbitsandbytes_cuda{major}{minor}` for CUDA and `libbitsandbytes_rocm{major}{minor}` for ROCm. bitsandbytes selects which one to load based on what PyTorch reports: ```python import torch print(torch.version.cuda) # e.g. "12.8" -> looks for libbitsandbytes_cuda128 -print(getattr(torch.version, "rocm", None) or torch.version.hip) -# e.g. "7.14" -> looks for libbitsandbytes_rocm714 +print(torch.version.hip) # e.g. "7.14" -> looks for libbitsandbytes_rocm714 ``` +The ROCm distribution version and HIP version can diverge. `torch.version.rocm` is informational release metadata, while `torch.version.hip` identifies the HIP version that PyTorch was built against. bitsandbytes uses the HIP `(major, minor)` tuple for its ROCm library suffix so that build-time and runtime lookup use the same version line. The `rocm` portion of the filename identifies the backend; its numeric suffix represents the HIP major and minor version. bitsandbytes does not infer a ROCm release from this HIP tuple. + bitsandbytes will automatically fall back to the closest available pre-compiled version if an exact match is not found, and log a warning. For example, if your PyTorch was built with CUDA 12.9 but bitsandbytes only ships 12.8, it will load 12.8 automatically. -ROCm may also fall back across major release numbers, preferring the newest available lower release and then the oldest newer release. Compatibility is not guaranteed: a binary from another major release may not contain code for your GPU architecture. A cross-major fallback therefore emits a stronger warning. Compile from source or set `BNB_ROCM_VERSION` if the selected binary fails to load or launch kernels. +For the ROCm backend, selection may also fall back across HIP major versions, preferring the newest available lower version and then the oldest newer version. This compares HIP tuples only and does not imply compatibility between the corresponding ROCm SDK releases. A binary from another HIP major version may also lack code for your GPU architecture, so a cross-major fallback emits a stronger warning. Compile from source or set `BNB_ROCM_VERSION` to select a different HIP-version suffix if the selected binary fails to load or launch kernels. If you see an error like `No compatible CUDA library found`, it means no compatible pre-compiled library could be found at all. To resolve this: @@ -47,4 +48,4 @@ If you see an error like `No compatible CUDA library found`, it means no compati # Windows (cmd) set BNB_CUDA_VERSION=128 ``` - Both backends accept dotted versions and compact shortcodes, for example `12.8` or `128` for CUDA and `7.14` or `714` for ROCm. + Both backends accept dotted versions and compact shortcodes, for example `12.8` or `128` for CUDA and `7.14` or `714` for the ROCm backend. diff --git a/docs/source/installation.mdx b/docs/source/installation.mdx index b368ce379..6d6a2973e 100644 --- a/docs/source/installation.mdx +++ b/docs/source/installation.mdx @@ -178,16 +178,14 @@ pip install bitsandbytes ### Compile from Source[[rocm-compile]] -bitsandbytes can be compiled from ROCm 6.3 - ROCm 7.14.0. See the `CMakeLists.txt` for additional options. - -CMake uses the ROCm release, rather than the HIP SDK version, to name the native library. It detects the release from the active ROCm installation's `.info/version` file or `rocm-sdk version`. If neither is available, pass the version explicitly as a dotted release such as `-DROCM_VERSION=7.14.0` or its shortcode, `-DROCM_VERSION=714`. The major and minor versions are concatenated in the filename, for example `libbitsandbytes_rocm714.so`. +bitsandbytes can be compiled from ROCm 6.4 - ROCm 7.14. See the `CMakeLists.txt` for additional options. To compile from source, you need CMake >= **3.31.6** and Python >= **3.10** installed. Make sure you have a compiler installed to compile C++ (`gcc`, `make`, headers, etc.). -You should also have a ROCm installation (system-wide or via Docker). The current minimum supported version is **6.3**. +You should also have a ROCm installation (system-wide or via Docker). The current minimum supported version is **6.4**. ```bash # Install bitsandbytes from source @@ -226,7 +224,7 @@ export ROCM_PATH="$(rocm-sdk path --root)" export PATH="${ROCM_PATH}/bin:${PATH}" git clone https://github.com/bitsandbytes-foundation/bitsandbytes.git && cd bitsandbytes/ # Use the same GPU architecture selected above. Separate multiple architectures with semicolons. -cmake -G Ninja -DCOMPUTE_BACKEND=hip -DROCM_VERSION=7.14.0 -DBNB_ROCM_ARCH="gfx1100" -DCMAKE_BUILD_TYPE=Release -DCMAKE_HIP_COMPILER_ROCM_ROOT="${ROCM_PATH}" -S . +cmake -G Ninja -DCOMPUTE_BACKEND=hip -DBNB_ROCM_ARCH="gfx1100" -DCMAKE_BUILD_TYPE=Release -DCMAKE_HIP_COMPILER_ROCM_ROOT="${ROCM_PATH}" -S . cmake --build . --config Release pip install . ``` @@ -248,7 +246,7 @@ rocm-sdk init export ROCM_PATH="$(rocm-sdk path --root)" export PATH="${ROCM_PATH}/bin:${PATH}" git clone https://github.com/bitsandbytes-foundation/bitsandbytes.git && cd bitsandbytes/ -cmake -G Ninja -DCOMPUTE_BACKEND=hip -DROCM_VERSION=7.2.1 -DBNB_ROCM_ARCH="gfx1100" -DCMAKE_BUILD_TYPE=Release -S . +cmake -G Ninja -DCOMPUTE_BACKEND=hip -DBNB_ROCM_ARCH="gfx1100" -DCMAKE_BUILD_TYPE=Release -S . cmake --build . --config Release pip install . ``` diff --git a/tests/test_cuda_setup_evaluator.py b/tests/test_cuda_setup_evaluator.py index 81e62ba3b..0781e1ca9 100644 --- a/tests/test_cuda_setup_evaluator.py +++ b/tests/test_cuda_setup_evaluator.py @@ -22,36 +22,36 @@ def specs(version: tuple[int, int]) -> CUDASpecs: [ # Exact match. ("cuda", "12.4", (12, 4), [(12, 4)], (12, 4), False), - # Same-major fallback to the newest older binary. - ("cuda", "12.9", (12, 9), [(12, 4), (12, 8)], (12, 8), True), # Same-major fallback to the oldest newer binary. ("cuda", "12.0", (12, 0), [(12, 1), (12, 4)], (12, 1), True), - # CUDA does not fall back across major versions. - ("cuda", "11.8", (11, 8), [(12, 1)], None, False), + # Same-major fallback to the newest older binary. + ("cuda", "12.9", (12, 9), [(12, 4), (12, 8)], (12, 8), True), # ROCm same-major fallback with a double-digit minor. - ("hip", "7.13.0", (7, 13), [(7, 2), (7, 14)], (7, 2), True), - # ROCm same-major fallback to the newest older binary. + ("hip", "7.13.0", (7, 13), [(7, 2)], (7, 2), True), + # ROCm same-major fallback prefers the newest older binary. ("hip", "7.9.0", (7, 9), [(7, 2), (7, 14)], (7, 2), True), - # ROCm/HIP version-line divergence with an older cross-major fallback. - ("hip", "7.16.0", (12, 1), [(8, 0), (7, 14)], (8, 0), True), # ROCm cross-major fallback to the newest older binary. ("hip", "8.0.0", (8, 0), [(7, 14)], (7, 14), True), # ROCm cross-major fallback to the oldest newer binary. ("hip", "6.4.0", (6, 4), [(7, 0)], (7, 0), True), + # CUDA does not fall back across major versions. + ("cuda", "11.8", (11, 8), [(12, 1), (12, 4)], None, False), # No packaged libraries returns the requested path without a warning. + ("cuda", "12.9", (12, 9), [], None, False), ("hip", "7.14.0", (7, 14), [], None, False), ], ) -def test_library_selection( +def test_version_selection( + monkeypatch, + caplog, backend, backend_version, runtime_version, available, expected, warning, - monkeypatch, - caplog, ): + """Library selection: exact match, fallback, no-same-major, no-libs.""" monkeypatch.delenv("BNB_CUDA_VERSION", raising=False) monkeypatch.delenv("BNB_ROCM_VERSION", raising=False) other_backend = "cuda" if backend == "hip" else "hip" @@ -60,9 +60,11 @@ def test_library_selection( version: Path(f"libbitsandbytes_{prefix}{version[0]}{version[1]}{DYNAMIC_LIBRARY_SUFFIX}") for version in available } + # ROCm release metadata is deliberately unrelated to HIP-based library selection. with ( patch.object(torch.version, backend, backend_version), patch.object(torch.version, other_backend, None), + patch.object(torch.version, "rocm", "10.0.0", create=True), patch("bitsandbytes.cextension._find_cuda_libs", return_value=paths), caplog.at_level("WARNING"), ): @@ -77,68 +79,72 @@ def test_library_selection( @pytest.mark.parametrize( - "backend,backend_version,version,override,expected_stem", + "backend,backend_version,runtime_version,override,expected_stem", [ ("hip", "7.0.0", (7, 0), "72", "libbitsandbytes_rocm72"), ("hip", "7.0.0", (7, 0), "7.2", "libbitsandbytes_rocm72"), + ("hip", "7.0.0", (7, 0), "714", "libbitsandbytes_rocm714"), + ("hip", "10.0.0", (10, 0), "1014", "libbitsandbytes_rocm1014"), ("cuda", "12.0", (12, 0), "128", "libbitsandbytes_cuda128"), ("cuda", "12.0", (12, 0), "12.8", "libbitsandbytes_cuda128"), ("cuda", "12.0", (12, 0), "12.8.1", "libbitsandbytes_cuda128"), ], ) -def test_override_formats(monkeypatch, backend, backend_version, version, override, expected_stem): +def test_override_formats(monkeypatch, caplog, backend, backend_version, runtime_version, override, expected_stem): other_backend = "cuda" if backend == "hip" else "hip" override_var = "BNB_ROCM_VERSION" if backend == "hip" else "BNB_CUDA_VERSION" - other_var = "BNB_CUDA_VERSION" if backend == "hip" else "BNB_ROCM_VERSION" + other_override_var = "BNB_CUDA_VERSION" if backend == "hip" else "BNB_ROCM_VERSION" monkeypatch.setenv(override_var, override) - monkeypatch.delenv(other_var, raising=False) + monkeypatch.delenv(other_override_var, raising=False) with ( patch.object(torch.version, backend, backend_version), patch.object(torch.version, other_backend, None), + patch("bitsandbytes.cextension._find_cuda_libs", return_value={}), + caplog.at_level("WARNING"), ): - assert get_cuda_bnb_library_path(specs(version)).stem == expected_stem + result = get_cuda_bnb_library_path(specs(runtime_version)) + assert result.stem == expected_stem + assert override_var in caplog.text @pytest.mark.parametrize( - "backend,backend_version,version", - [ - ("cuda", "12.0", (12, 0)), - ("hip", "7.2.0", (7, 2)), - ], + "backend,torch_version,runtime_version", + [("cuda", "12.0", (12, 0)), ("hip", "7.2.0", (7, 2))], ) -def test_opposite_backend_override_warns(monkeypatch, caplog, backend, backend_version, version): - other_backend = "hip" if backend == "cuda" else "cuda" - correct_var = "BNB_CUDA_VERSION" if backend == "cuda" else "BNB_ROCM_VERSION" - wrong_var = "BNB_ROCM_VERSION" if backend == "cuda" else "BNB_CUDA_VERSION" - monkeypatch.setenv(wrong_var, "72") - monkeypatch.delenv(correct_var, raising=False) +def test_override_invalid_format(monkeypatch, backend, torch_version, runtime_version): + """Reject malformed overrides for both backends.""" + other_backend = "cuda" if backend == "hip" else "hip" + override_var = "BNB_ROCM_VERSION" if backend == "hip" else "BNB_CUDA_VERSION" + other_override_var = "BNB_CUDA_VERSION" if backend == "hip" else "BNB_ROCM_VERSION" + monkeypatch.setenv(override_var, "not-a-version") + monkeypatch.delenv(other_override_var, raising=False) with ( - patch.object(torch.version, backend, backend_version), + patch.object(torch.version, backend, torch_version), patch.object(torch.version, other_backend, None), - patch("bitsandbytes.cextension._find_cuda_libs", return_value={}), - caplog.at_level("WARNING"), + pytest.raises(RuntimeError, match="dotted version"), ): - get_cuda_bnb_library_path(specs(version)) - assert f"{wrong_var} is ignored" in caplog.text - assert f"use {correct_var} instead" in caplog.text + get_cuda_bnb_library_path(specs(runtime_version)) @pytest.mark.parametrize( - "backend,backend_version,version", + "backend,backend_version,runtime_version,wrong_var,correct_var", [ - ("hip", "7.0.0", (7, 0)), - ("cuda", "12.0", (12, 0)), + ("cuda", "12.0", (12, 0), "BNB_ROCM_VERSION", "BNB_CUDA_VERSION"), + ("hip", "7.2.0", (7, 2), "BNB_CUDA_VERSION", "BNB_ROCM_VERSION"), ], ) -def test_invalid_override(monkeypatch, backend, backend_version, version): +def test_opposite_backend_override_warns( + monkeypatch, caplog, backend, backend_version, runtime_version, wrong_var, correct_var +): other_backend = "cuda" if backend == "hip" else "hip" - override_var = "BNB_ROCM_VERSION" if backend == "hip" else "BNB_CUDA_VERSION" - other_var = "BNB_CUDA_VERSION" if backend == "hip" else "BNB_ROCM_VERSION" - monkeypatch.setenv(override_var, "not-a-version") - monkeypatch.delenv(other_var, raising=False) + monkeypatch.setenv(wrong_var, "72") + monkeypatch.delenv(correct_var, raising=False) with ( patch.object(torch.version, backend, backend_version), patch.object(torch.version, other_backend, None), - pytest.raises(RuntimeError, match="dotted version"), + patch("bitsandbytes.cextension._find_cuda_libs", return_value={}), + caplog.at_level("WARNING"), ): - get_cuda_bnb_library_path(specs(version)) + get_cuda_bnb_library_path(specs(runtime_version)) + assert f"{wrong_var} is ignored" in caplog.text + assert f"use {correct_var} instead" in caplog.text