Merge upstream/main into merge-cuda-hip

Author: Abdennacer-Badaoui

.github/workflows/test-runner.yml

@@ -26,7 +26,7 @@ on:
       gpu_type:
         type: string
         default: ""
-        description: "GPU type for CUDA testing: T4, L40S"
+        description: "GPU type for CUDA testing: T4, A10,L40S"
       # cpu_type currently only affects linux x64 CPU testing to select specific CPU architectures
       cpu_type:
         type: string
@@ -65,11 +65,14 @@ jobs:
                   T4)
                     TEST_RUNNER="bandb-aws-g4dn-4xlarge-plus-use1-public-80"
                     ;;
+                  A10)
+                    TEST_RUNNER="bandb-aws-g5-4xlarge-plus-use1-public-80"
+                    ;;
                   L40S)
                     TEST_RUNNER="bandb-aws-g6e-4xlarge-plus-use1-public-80"
                     ;;
                   *)
-                    echo "::error::Must specify gpu_type (T4 or L40S) for linux-x64 cuda backend"
+                    echo "::error::Must specify gpu_type (T4, A10, L40S) for linux-x64 cuda backend"
                     exit 1
                     ;;
                 esac
@@ -164,7 +167,7 @@ jobs:
         run: bash .github/scripts/build-cuda.sh
         env:
           cuda_version: ${{ inputs.cuda_version }}
-          cuda_targets: "75;89"
+          cuda_targets: "75;80;89"
 
       - name: Upload build artifact
         uses: actions/upload-artifact@v4

.github/workflows/tests-nightly.yml

@@ -56,7 +56,7 @@ jobs:
       matrix:
         # Linux x64 cross-product
         platform: [linux-x64]
-        gpu_type: [T4, L40S]
+        gpu_type: [T4, A10, L40S]
         cuda_version: ["11.8.0", "12.6.3", "12.8.1", "13.0.2"]
 
         include:

.github/workflows/tests-pr.yml

@@ -64,7 +64,7 @@ jobs:
       fail-fast: false
       matrix:
         platform: [linux-x64]
-        gpu_type: [T4, L40S]
+        gpu_type: [T4, A10, L40S]
         cuda_version: ["11.8.0", "12.8.1", "13.0.2"]
 
         include:

CLAUDE.md

@@ -8,7 +8,9 @@ git worktree add ~/git/bnb-fix-<NUMBER> -b fix/issue-<NUMBER>
 cd ~/git/bnb-fix-<NUMBER>
 ```
 
-This keeps the main checkout clean and allows parallel sessions. If you are already inside a worktree directory, you do not need to create another one. Full guide: `agents/worktree_guide.md`
+This keeps the main checkout clean and allows parallel sessions. If you are already inside a worktree directory, you do not need to create another one.
+
+**Before creating a worktree**, check the worktree registry for existing ones — see the Git Worktrees section in `~/.claude/CLAUDE.md`. Bitsandbytes-specific naming conventions: `agents/worktree_guide.md`. General worktree guide: `~/git/lab_tools/worktree_guide.md`.
 
 # MANDATORY: Check for existing PRs before starting work
 

CMakeLists.txt

@@ -10,8 +10,38 @@
 #                        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.
 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})
+        set(ROCM_PATH $ENV{ROCM_PATH})
+    endif()
+    if(ROCM_PATH AND NOT DEFINED CMAKE_CXX_COMPILER)
+        set(CMAKE_CXX_COMPILER "${ROCM_PATH}/lib/llvm/bin/clang++.exe")
+    endif()
+    if(ROCM_PATH AND NOT DEFINED CMAKE_HIP_COMPILER)
+        set(CMAKE_HIP_COMPILER "${ROCM_PATH}/lib/llvm/bin/clang++.exe")
+    endif()
+    # On Windows, the HIP compiler needs explicit paths to find device libraries.
+    if(ROCM_PATH)
+        find_path(ROCM_DEVICE_LIB_PATH
+            NAMES oclc_abi_version_400.bc ocml.bc
+            PATHS "${ROCM_PATH}/amdgcn/bitcode"
+                  "${ROCM_PATH}/lib/llvm/amdgcn/bitcode"
+            NO_DEFAULT_PATH
+        )
+        set(CMAKE_HIP_FLAGS "--rocm-path=${ROCM_PATH}")
+        if(ROCM_DEVICE_LIB_PATH)
+            set(CMAKE_HIP_FLAGS "${CMAKE_HIP_FLAGS} --rocm-device-lib-path=${ROCM_DEVICE_LIB_PATH}")
+        endif()
+    endif()
+endif()
+
 project(bitsandbytes LANGUAGES CXX)
 
 # If run without specifying a build type, default to using the Release configuration:
@@ -199,17 +229,18 @@ if(BUILD_CUDA)
     string(APPEND BNB_OUTPUT_NAME "_cuda${CUDA_VERSION_SHORT}")
     add_compile_definitions(BUILD_CUDA)
 elseif(BUILD_HIP)
-    enable_language(HIP)
-    message(STATUS "HIP Compiler: ${CMAKE_HIP_COMPILER}")
+    # Set target architectures before enable_language(HIP), which would otherwise
+    # auto-detect a single GPU and override the defaults.
     if(DEFINED BNB_ROCM_ARCH)
       set(CMAKE_HIP_ARCHITECTURES ${BNB_ROCM_ARCH})
-    else()
-      if (NOT AMDGPU_TARGETS AND NOT CMAKE_HIP_ARCHITECTURES)
-        set(CMAKE_HIP_ARCHITECTURES "gfx90a;gfx942;gfx1100;gfx1101;gfx1150;gfx1151;gfx1200;gfx1201")
-      elseif (AMDGPU_TARGETS AND NOT CMAKE_HIP_ARCHITECTURES)
-        set(CMAKE_HIP_ARCHITECTURES ${AMDGPU_TARGETS})
-      endif()
+    elseif(AMDGPU_TARGETS AND NOT CMAKE_HIP_ARCHITECTURES)
+      set(CMAKE_HIP_ARCHITECTURES ${AMDGPU_TARGETS})
+    elseif(NOT CMAKE_HIP_ARCHITECTURES)
+      set(CMAKE_HIP_ARCHITECTURES "gfx90a;gfx942;gfx1100;gfx1101;gfx1150;gfx1151;gfx1200;gfx1201")
     endif()
+
+    enable_language(HIP)
+    message(STATUS "HIP Compiler: ${CMAKE_HIP_COMPILER}")
     message(STATUS "HIP Targets: ${CMAKE_HIP_ARCHITECTURES}")
 
     list(APPEND SRC_FILES ${GPU_FILES})
@@ -221,7 +252,15 @@ elseif(BUILD_HIP)
     string(REGEX MATCH "[0-9]+\\.[0-9]+" HIP_VERSION "${HIP_CONFIG_VERSION}")
     string(REPLACE "." "" HIP_VERSION_SHORT "${HIP_VERSION}")
 
-    string(APPEND BNB_OUTPUT_NAME "${HIP_VERSION_SHORT}")
+    # 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")
+
+    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}")
+    endif()
+
+    string(APPEND BNB_OUTPUT_NAME "${ROCM_VERSION}")
     add_compile_definitions(__HIP_PLATFORM_AMD__)
     add_compile_definitions(__HIP_PLATFORM_HCC__)
     add_compile_definitions(BUILD_HIP)
@@ -262,6 +301,8 @@ endif()
 if(WIN32)
     # Export all symbols
     set(CMAKE_WINDOWS_EXPORT_ALL_SYMBOLS ON)
+    # Prevent Windows SDK min/max macros from conflicting with std::min/std::max
+    add_compile_definitions(NOMINMAX)
 endif()
 
 if(MSVC)
@@ -314,10 +355,11 @@ if(BUILD_CUDA)
     )
 endif()
 if(BUILD_HIP)
-    if(NOT DEFINED ENV{ROCM_PATH})
-      set(ROCM_PATH /opt/rocm)
-    else()
+    # Determine ROCM_PATH from environment variable, fallback to /opt/rocm on Linux
+    if(DEFINED ENV{ROCM_PATH})
       set(ROCM_PATH $ENV{ROCM_PATH})
+    else()
+      set(ROCM_PATH /opt/rocm)
     endif()
     list(APPEND CMAKE_PREFIX_PATH ${ROCM_PATH})
     macro(find_package_and_print_version PACKAGE_NAME)
@@ -329,14 +371,23 @@ if(BUILD_HIP)
     find_package_and_print_version(hipsparse REQUIRED)
 
     ## hacky way of excluding hip::amdhip64 (with it linked many tests unexpectedly fail e.g. adam8bit because of inaccuracies)
-    set_target_properties(hip::host PROPERTIES INTERFACE_LINK_LIBRARIES "")
-    set_target_properties(hip-lang::host PROPERTIES INTERFACE_LINK_LIBRARIES "")
-    set(CMAKE_HIP_IMPLICIT_LINK_LIBRARIES "")
+    ## On Windows, we need to link amdhip64 explicitly
+    if(NOT WIN32)
+        set_target_properties(hip::host PROPERTIES INTERFACE_LINK_LIBRARIES "")
+        set_target_properties(hip-lang::host PROPERTIES INTERFACE_LINK_LIBRARIES "")
+        set(CMAKE_HIP_IMPLICIT_LINK_LIBRARIES "")
+    endif()
 
     target_include_directories(bitsandbytes PRIVATE ${CMAKE_SOURCE_DIR} ${CMAKE_SOURCE_DIR}/include ${ROCM_PATH}/include /include)
     target_link_directories(bitsandbytes PRIVATE ${ROCM_PATH}/lib /lib)
     target_link_libraries(bitsandbytes PUBLIC roc::hipblas hip::hiprand roc::hipsparse)
 
+    # On Windows, rocblas is not pulled in transitively by roc::hipblas
+    # and is needed because ops_hip.cuh uses rocblas_handle directly.
+    if(WIN32)
+        target_link_libraries(bitsandbytes PUBLIC rocblas)
+    endif()
+
     target_compile_definitions(bitsandbytes PUBLIC BNB_USE_HIP)
     set_source_files_properties(${GPU_FILES} PROPERTIES LANGUAGE HIP)
     set_target_properties(bitsandbytes PROPERTIES LINKER_LANGUAGE CXX)

agents/architecture_guide.md

@@ -329,6 +329,7 @@ GPU-specific functions are actually invoked.
 ### Environment variables
 
 - `BNB_CUDA_VERSION` — Override the auto-detected CUDA version for library selection
+    - `BNB_ROCM_VERSION` is the ROCm equivalent
 - Standard CUDA env vars (`CUDA_HOME`, `LD_LIBRARY_PATH`) affect library discovery
 
 ---

agents/dispatch_guide.md

@@ -12,6 +12,27 @@ python3 agents/fetch_issues.py
 
 Read `agents/github_tools_guide.md` for the full reference on how to use the query tools.
 
+## Step 0: Check Existing Reviews on Open PRs
+
+Before looking at issues, check whether there are open PRs from external contributors that need review. But **do not assume a PR needs review just because it is open** — check whether a review has already been posted.
+
+```bash
+# List open PRs
+gh pr list --state open --limit 30
+
+# For each external contributor PR, check for existing reviews
+gh api repos/bitsandbytes-foundation/bitsandbytes/pulls/<NUMBER>/reviews \
+  --jq '.[] | "\(.user.login) | \(.state) | \(.submitted_at)"'
+```
+
+A PR only needs a new review if:
+
+- **No review exists at all** from a maintainer or agent
+- **The author has pushed new commits** since the last review (check commit dates vs review dates)
+- **The author has responded to review feedback** and the review needs a re-review
+
+If a review already exists and the author has not responded or pushed changes, the ball is in the author's court — skip that PR. Do not generate a prompt to re-review work that has already been reviewed.
+
 ## Step 1: Find Candidate Issues
 
 Start by getting the landscape of open issues: