hipBLAS on CLBlast/OpenCL via chipStar SVM bridge

.github/workflows/ci.yml

@@ -0,0 +1,160 @@
+# chipBLAS continuous integration.
+#
+# GitHub-hosted runners have no GPU, so we cannot run the test binaries.
+# We also cannot install chipStar (requires HIP toolchain + OpenCL/L0 drivers
+# + a working SPIR-V compiler stack). When chipStar is unavailable the
+# workflow falls back to a smoke job that verifies the source tree is
+# well-formed and CMake can parse the project.
+#
+# Full testing requires a self-hosted runner with chipStar pre-installed and
+# an OpenCL device accessible.  Point CHIPSTAR_DIR at the install prefix.
+#
+# Self-hosted runner .env requirements:
+#   CHIPSTAR_DIR=<chipStar-install-prefix>
+#   CHIP_BE=opencl
+#   # OpenCL ICD loader install prefix (if separate from chipStar):
+#   OCL_ICD_DIR=<ocl-icd-loader-install-prefix>
+#   # On Mali/devices needing profiling disabled:
+#   CHIP_OCL_DISABLE_QUEUE_PROFILING=on
+#   # chipBLAS SVM wrap (USE_HOST_PTR): workflow Test steps set
+#   # CHIP_OCL_USE_ALLOC_STRATEGY=svm unless you override in the runner env.
+#   # On Mali-G52 (no fp64) set:
+#   CHIPBLAS_CMAKE_EXTRA=-DCHIPBLAS_HAS_FP64=OFF
+
+name: ci
+
+on:
+  push:
+    branches: ['**']
+  pull_request:
+
+jobs:
+  smoke:
+    name: smoke (cmake parse)
+    runs-on: ubuntu-22.04
+    steps:
+      - uses: actions/checkout@v4
+        with:
+          submodules: recursive
+
+      - name: Install build basics
+        run: |
+          sudo apt-get update
+          sudo apt-get install -y cmake ninja-build build-essential \
+                                  ocl-icd-opencl-dev opencl-headers
+
+      - name: Verify source tree
+        run: |
+          test -f CMakeLists.txt
+          test -d src
+          test -d include/hipblas
+          cmake --version
+
+  build:
+    name: build (best-effort, vendored CLBlast)
+    runs-on: ubuntu-22.04
+    # Don't fail the workflow if chipStar isn't available; this job is
+    # informational.  A self-hosted runner with chipStar should set
+    # continue-on-error: false.
+    continue-on-error: true
+    steps:
+      - uses: actions/checkout@v4
+        with:
+          submodules: recursive
+
+      - name: Install build basics
+        run: |
+          sudo apt-get update
+          sudo apt-get install -y cmake ninja-build build-essential \
+                                  ocl-icd-opencl-dev opencl-headers
+
+      - name: Probe for chipStar
+        run: |
+          if [ -z "${CHIPSTAR_DIR:-}" ] || [ ! -d "${CHIPSTAR_DIR}" ]; then
+            echo "chipStar not found (CHIPSTAR_DIR unset). Skipping configure/build."
+            echo "SKIP_BUILD=1" >> $GITHUB_ENV
+          fi
+
+      - name: Configure
+        if: env.SKIP_BUILD != '1'
+        run: |
+          cmake -S . -B build -G Ninja \
+            -DCMAKE_PREFIX_PATH="${CHIPSTAR_DIR}" \
+            -DCHIPBLAS_USE_VENDORED_CLBLAST=ON \
+            -DCHIPBLAS_BUILD_TESTS=ON
+
+      - name: Build
+        if: env.SKIP_BUILD != '1'
+        run: cmake --build build -j
+
+      - name: Test
+        if: env.SKIP_BUILD != '1'
+        working-directory: build
+        run: |
+          CHIP_BE=opencl \
+          LD_LIBRARY_PATH="${CHIPSTAR_DIR}/lib:${PWD}" \
+          ctest --output-on-failure -j1
+
+      # For devices without fp64 (e.g. Mali-G52), set:
+      #   CHIP_OCL_DISABLE_QUEUE_PROFILING=on
+      # in the runner environment if the device requires it.
+
+  test-opencl-linux:
+    name: test (linux-arm64, chipStar/OpenCL)
+    runs-on: [self-hosted, chipstar, opencl, linux]
+    steps:
+      - uses: actions/checkout@v4
+        with:
+          submodules: recursive
+
+      - name: Configure
+        run: |
+          cmake -S . -B build -G Ninja \
+            -DCMAKE_PREFIX_PATH="${CHIPSTAR_DIR}" \
+            -DCHIPBLAS_USE_VENDORED_CLBLAST=ON \
+            -DCHIPBLAS_BUILD_TESTS=ON \
+            ${CHIPBLAS_CMAKE_EXTRA:-}
+
+      - name: Build
+        run: cmake --build build -j
+
+      - name: Test
+        working-directory: build
+        env:
+          CHIP_OCL_USE_ALLOC_STRATEGY: svm
+          CHIPBLAS_SKIP_HALF_API_SURFACE: "1"
+        run: |
+          CHIP_BE=opencl \
+          LD_LIBRARY_PATH="${CHIPSTAR_DIR}/lib:${PWD}" \
+          ctest --output-on-failure -j1
+
+  test-opencl-macos:
+    name: test (macos-arm64, chipStar/PoCL)
+    runs-on: [self-hosted, chipstar, opencl, macos]
+    steps:
+      - uses: actions/checkout@v4
+        with:
+          submodules: recursive
+
+      - name: Configure
+        run: |
+          cmake -S . -B build -G Ninja \
+            -DCMAKE_PREFIX_PATH="${CHIPSTAR_DIR}${OCL_ICD_DIR:+;${OCL_ICD_DIR}}" \
+            -DCHIPBLAS_USE_VENDORED_CLBLAST=ON \
+            -DCHIPBLAS_BUILD_TESTS=ON \
+            ${CHIPBLAS_CMAKE_EXTRA:-}
+
+      - name: Build
+        run: cmake --build build -j
+
+      # Pastrami's PoCL path reports CLBlast kNoHalfPrecision (-2045) for
+      # hipblasHalf* routines; skip them in api_surface via env (see test file).
+      - name: Test
+        working-directory: build
+        env:
+          CHIP_OCL_USE_ALLOC_STRATEGY: svm
+          CHIPBLAS_SKIP_HALF_API_SURFACE: "1"
+        run: |
+          CHIP_BE=opencl \
+          DYLD_FALLBACK_LIBRARY_PATH="${CHIPSTAR_DIR}/lib:${PWD}:${DYLD_FALLBACK_LIBRARY_PATH:-}" \
+          ctest --output-on-failure -j1

.gitmodules

@@ -0,0 +1,6 @@
+[submodule "third_party/CLBlast"]
+	path = third_party/CLBlast
+	url = https://github.com/CNugteren/CLBlast.git
+[submodule "third_party/OpenBLAS"]
+	path = third_party/OpenBLAS
+	url = https://github.com/OpenMathLib/OpenBLAS.git

CMakeLists.txt

@@ -0,0 +1,126 @@
+cmake_minimum_required(VERSION 3.20)
+project(chipBLAS
+    VERSION 0.1.0
+    DESCRIPTION "Portable hipBLAS on chipStar + CLBlast"
+    LANGUAGES C CXX)
+
+set(CMAKE_CXX_STANDARD 17)
+set(CMAKE_CXX_STANDARD_REQUIRED ON)
+set(CMAKE_CXX_EXTENSIONS OFF)
+
+option(CHIPBLAS_BUILD_TESTS "Build chipBLAS tests" ON)
+option(CHIPBLAS_USE_VENDORED_CLBLAST
+       "Build CLBlast from third_party/CLBlast (off = use system CLBlast)" ON)
+
+# On macOS, prevent any find_package/find_library from selecting Apple's
+# OpenCL.framework. Linking CLBlast against the framework while the
+# loader-from-chipStar is active at runtime hands queues across two
+# different OpenCL stacks and crashes in dynamic_cast on the first
+# clRetain*. Must be set before any find_package(OpenCL).
+if(APPLE)
+    set(CMAKE_FIND_FRAMEWORK NEVER)
+endif()
+
+# chipStar's CHIPTargets.cmake transitively requires Threads::Threads.
+set(THREADS_PREFER_PTHREAD_FLAG ON)
+find_package(Threads REQUIRED)
+
+# --- chipStar -----------------------------------------------------------------
+# chipStar's install ships CHIPTargets.cmake (no Config wrapper); locate it
+# via CHIP_DIR or by searching CMAKE_PREFIX_PATH.
+find_path(CHIP_TARGETS_DIR CHIPTargets.cmake
+    HINTS
+        ${CHIP_DIR}
+        ${CMAKE_PREFIX_PATH}
+    PATH_SUFFIXES cmake/CHIP)
+if(NOT CHIP_TARGETS_DIR)
+    message(FATAL_ERROR
+        "Could not locate chipStar's CHIPTargets.cmake. "
+        "Pass -DCHIP_DIR=<chipStar-install>/cmake/CHIP "
+        "or -DCMAKE_PREFIX_PATH=<chipStar-install>.")
+endif()
+include(${CHIP_TARGETS_DIR}/CHIPTargets.cmake)
+message(STATUS "chipBLAS: using chipStar from ${CHIP_TARGETS_DIR}")
+
+# --- OpenCL -------------------------------------------------------------------
+# CLBlast is OpenCL-only, so we always need OpenCL headers + ICD loader.
+find_package(OpenCL REQUIRED)
+
+# On macOS, CLBlast's clblast.h does `#include <OpenCL/opencl.h>` (the
+# framework-style spelling). Apple's framework header uses Apple-only
+# deprecation macros that don't survive a non-Apple OpenCL stack. Force
+# that include to the Khronos <CL/opencl.h> via a one-line shim, by
+# putting our shim dir first on every target's include path.
+if(APPLE)
+    include_directories(BEFORE
+        ${CMAKE_CURRENT_SOURCE_DIR}/cmake/opencl_shim
+        ${OpenCL_INCLUDE_DIRS})
+endif()
+
+# --- CLBlast ------------------------------------------------------------------
+if(CHIPBLAS_USE_VENDORED_CLBLAST)
+    if(NOT EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/third_party/CLBlast/CMakeLists.txt)
+        message(FATAL_ERROR
+            "third_party/CLBlast is empty. Run:\n"
+            "  git submodule add https://github.com/CNugteren/CLBlast.git "
+            "third_party/CLBlast\n"
+            "or pass -DCHIPBLAS_USE_VENDORED_CLBLAST=OFF to use the system "
+            "package.")
+    endif()
+    # Suppress CLBlast's own tests/samples — we only want the library.
+    set(TESTS    OFF CACHE BOOL "" FORCE)
+    set(CLIENTS  OFF CACHE BOOL "" FORCE)
+    set(SAMPLES  OFF CACHE BOOL "" FORCE)
+    set(TUNERS   OFF CACHE BOOL "" FORCE)
+    set(NETLIB   OFF CACHE BOOL "" FORCE)
+    add_subdirectory(third_party/CLBlast EXCLUDE_FROM_ALL)
+    set(CLBLAST_TARGET clblast)
+else()
+    find_package(CLBlast REQUIRED)
+    set(CLBLAST_TARGET clblast)
+endif()
+
+# --- Library ------------------------------------------------------------------
+set(_CHIPBLAS_SRC
+    src/hipblas_common.cc
+    src/hipblas_ocl.cc
+    src/hipblas_tuning.cc
+    src/hipblas_l1.cc
+    src/hipblas_l2.cc
+    src/hipblas_l3.cc
+    src/hipblas_extras.cc
+)
+
+add_library(hipblas SHARED ${_CHIPBLAS_SRC})
+target_include_directories(hipblas
+    PUBLIC
+        $<BUILD_INTERFACE:${CMAKE_CURRENT_SOURCE_DIR}/include>
+        $<INSTALL_INTERFACE:include>
+    PRIVATE
+        ${CMAKE_CURRENT_SOURCE_DIR}/src)
+target_compile_definitions(hipblas
+    PUBLIC __HIP_PLATFORM_SPIRV__)
+target_link_libraries(hipblas
+    PUBLIC  CHIP::CHIP
+    PRIVATE OpenCL::OpenCL ${CLBLAST_TARGET})
+
+set_target_properties(hipblas PROPERTIES
+    OUTPUT_NAME hipblas
+    SOVERSION   ${PROJECT_VERSION_MAJOR}
+    VERSION     ${PROJECT_VERSION})
+
+# --- Install ------------------------------------------------------------------
+include(GNUInstallDirs)
+install(TARGETS hipblas
+    EXPORT chipBLASTargets
+    LIBRARY DESTINATION ${CMAKE_INSTALL_LIBDIR}
+    ARCHIVE DESTINATION ${CMAKE_INSTALL_LIBDIR}
+    RUNTIME DESTINATION ${CMAKE_INSTALL_BINDIR})
+install(DIRECTORY include/hipblas
+    DESTINATION ${CMAKE_INSTALL_INCLUDEDIR})
+
+# --- Tests --------------------------------------------------------------------
+if(CHIPBLAS_BUILD_TESTS)
+    enable_testing()
+    add_subdirectory(test)
+endif()