HRX CI: run on self-hosted ROCm GPU runners for gfx1151/gfx1201

.github/workflows/build-android.yml

@@ -16,6 +16,7 @@ on:
     ]
 
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/build-android.yml',

.github/workflows/build-apple.yml

@@ -19,6 +19,7 @@ on:
     ]
 
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/build-apple.yml',

.github/workflows/build-cann.yml

@@ -16,6 +16,7 @@ on:
     ]
 
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/build-cann.yml',

.github/workflows/build-hrx.yml

@@ -1,6 +1,9 @@
 name: HRX CI
 
 on:
+  push:
+    branches:
+      - hrx-integration
   pull_request:
     types: [opened, synchronize, reopened]
     branches:
@@ -16,40 +19,95 @@ concurrency:
 
 jobs:
   hrx-build:
-    name: ubuntu-latest
-    runs-on: ubuntu-latest
+    name: ${{ matrix.name }} (${{ matrix.gpu_target }})
+    strategy:
+      # Per-arch failures are independent; keep both signals.
+      fail-fast: false
+      matrix:
+        include:
+          - name: gfx1151_strix-halo
+            gpu_target: gfx1151
+            runs_on: linux-gfx1151-gpu-rocm
+            extra_rocm_artifacts: rocwmma_dev_gfx1151
+          - name: gfx1201_9070
+            gpu_target: gfx1201
+            runs_on: linux-gfx120X-gpu-rocm
+            extra_rocm_artifacts: rocwmma_dev_gfx120X-all
+    runs-on: ${{ matrix.runs_on }}
+    timeout-minutes: 10
+    defaults:
+      run:
+        # --noprofile skips loading ~/.bash_profile
+        # --norc      skips loading ~/.bashrc
+        # -exo pipefail:
+        #   -e exit immediately if any command fails
+        #   -x print each command before executing
+        #   -o pipefail ensure pipeline fails if any command in it fails
+        # {0} github actions placeholder for the command to run
+        shell: bash --noprofile --norc -exo pipefail {0}
+    container:
+      image: ghcr.io/rocm/no_rocm_image_ubuntu24_04@sha256:fba5f55a122dbb15925e98c51fe65bffe88c36e11ebb25b73daf2bea04202dc3
+      # --user 0:0: actions running inside the container need to write to some
+      #             files set up outside the container by the runner agent.  In
+      #             June 2026 some runner agents set those files up with
+      #             permissions for a "tester" user with UID/GID 1001, which
+      #             matches the user in no_rocm_image_ubuntu24_04, and some are
+      #             set up as root; accessing a file owned by root with user
+      #             "tester" gives an EACCES. Running as root is the common
+      #             denominator.
+      # --device kfd/dri: GPU access.
+      options: >-
+        --user 0:0
+        --device /dev/kfd
+        --device /dev/dri
     env:
-      HRX_WORK_DIR: ${{ github.workspace }}
+      # Write out of tree build/install to a non-bind-mounted directory. In June
+      # 2026 a debug test showed some runners had assets dated from 6 months ago
+      # in the bind-mounted github workspace, it seems like nothing cleans it up
+      # currently. Writing to an non-bind-mounted directory ensures that
+      # rocm-artifacts/builds are cleaned up when docker container is removed.
+      HRX_WORK_DIR: /work
       # Public location for bench tooling (rsuderman/llamacpp_ci and fork
       # AaronStGeorge/llamacpp_ci)
       #TODO: switch to ROCm/llamacpp-hrx-bench once it is open sourced.
       BENCH_REPOSITORY: AaronStGeorge/llamacpp_ci
       BENCH_REF: 'main'
+      # actions/checkout@v6 does auto-cleanup, an in-tree build would be
+      # auto-cleaned as well.
       BENCH_DIR: ${{ github.workspace }}/bench
       LLAMA_SRC_DIR: ${{ github.workspace }}/llama-src
-      HRX_ARTIFACT_SET: 'core-with-upstream-hip'
+      LLAMA_BUILD_DIR: /work/llama-build
+      HRX_EXTRA_ROCM_ARTIFACTS: ${{ matrix.extra_rocm_artifacts }}
+      GGML_HRX_AMDGPU_TARGETS: ${{ matrix.gpu_target }}
       CCACHE_COMPILERCHECK: content
 
     steps:
-      - name: Checkout llama.cpp (under test)
-        uses: actions/checkout@v6
-        with:
-          path: llama-src
-
       - name: Checkout bench tooling
         uses: actions/checkout@v6
         with:
           repository: ${{ env.BENCH_REPOSITORY }}
           ref: ${{ env.BENCH_REF }}
           path: bench
 
+      - name: Runner info
+        env:
+          MATRIX_NAME: ${{ matrix.name }}
+          MATRIX_GPU_TARGET: ${{ matrix.gpu_target }}
+          MATRIX_RUNS_ON: ${{ matrix.runs_on }}
+        run: "${BENCH_DIR}/scripts/hrx/runner-info.sh"
+
+      - name: Checkout llama.cpp (under test)
+        uses: actions/checkout@v6
+        with:
+          path: llama-src
+
       - name: Install ROCm build dependencies
         run: "${BENCH_DIR}/scripts/hrx/install-rocm-deps.sh"
 
       - name: ccache
         uses: ggml-org/ccache-action@v1.2.21
         with:
-          key: ubuntu-latest-${{ env.HRX_ARTIFACT_SET }}
+          key: ${{ matrix.gpu_target }}
           evict-old-files: 1d
           save: ${{ github.event_name == 'pull_request' }}
 
@@ -68,28 +126,30 @@ jobs:
       - name: Build llama.cpp with HRX
         run: "${BENCH_DIR}/scripts/hrx/build-llama-hrx.sh"
 
-      - name: Run sample MUL_MAT correctness config on CPU
+      - name: Run sample MUL_MAT correctness config on HRX
         run: |
           . "${BENCH_DIR}/scripts/hrx/env.sh"
+          . "${BENCH_DIR}/scripts/hrx/runtime-env.sh"
           "${BENCH_DIR}/tools/run-op-test.py" \
             --test-backend-ops "${LLAMA_BUILD_DIR}/bin/test-backend-ops" \
             --test-file "${BENCH_DIR}/benchmark-configs/test/mul_mat_f16.txt" \
             --op MUL_MAT \
-            --backend CPU \
-            --output benchmark-results/sample-mul-mat-f16-cpu-test.jsonl
+            --backend HRX0 \
+            --output benchmark-results/sample-mul-mat-f16-hrx-test.jsonl
 
-      - name: Run sample MUL_MAT benchmark config on CPU
+      - name: Run sample MUL_MAT benchmark config on HRX
         run: |
           . "${BENCH_DIR}/scripts/hrx/env.sh"
+          . "${BENCH_DIR}/scripts/hrx/runtime-env.sh"
           "${BENCH_DIR}/tools/run-op-perf.py" \
             --test-backend-ops "${LLAMA_BUILD_DIR}/bin/test-backend-ops" \
             --test-file "${BENCH_DIR}/benchmark-configs/test/mul_mat_f16.txt" \
             --op MUL_MAT \
-            --backend CPU \
-            --output benchmark-results/sample-mul-mat-f16-cpu-perf.jsonl
+            --backend HRX0 \
+            --output benchmark-results/sample-mul-mat-f16-hrx-perf.jsonl
 
-      - name: Upload CPU benchmark results
+      - name: Upload benchmark results
         uses: actions/upload-artifact@v5
         with:
-          name: cpu-benchmark-results
+          name: benchmark-results-${{ matrix.name }}
           path: benchmark-results/

.github/workflows/build-riscv.yml

@@ -16,6 +16,7 @@ on:
     ]
 
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/build-riscv.yml',

.github/workflows/build-self-hosted.yml

@@ -24,6 +24,7 @@ on:
     ]
 
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/build-self-hosted.yml',

.github/workflows/build-vulkan.yml

@@ -18,6 +18,7 @@ on:
     ]
 
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/build-vulkan.yml',

.github/workflows/build.yml

@@ -25,6 +25,7 @@ on:
     ]
 
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/build.yml',

.github/workflows/check-vendor.yml

@@ -11,6 +11,7 @@ on:
     ]
 
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       'vendor/**',

.github/workflows/copilot-setup-steps.yml

@@ -5,9 +5,11 @@ name: "Copilot Setup Steps"
 on:
   workflow_dispatch:
   push:
+    branches-ignore: ['hrx-integration']
     paths:
       - .github/workflows/copilot-setup-steps.yml
   pull_request:
+    branches-ignore: ['hrx-integration']
     paths:
       - .github/workflows/copilot-setup-steps.yml
 

.github/workflows/hip-quality-check.yml

@@ -13,6 +13,7 @@ on:
     ]
 
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/hip-quality-check.yml',

.github/workflows/labeler.yml

@@ -1,6 +1,7 @@
 name: "Pull Request Labeler"
 on:
-- pull_request_target
+  pull_request_target:
+    branches-ignore: ['hrx-integration']
 
 jobs:
   labeler:

.github/workflows/pre-tokenizer-hashes.yml

@@ -2,10 +2,12 @@ name: Check Pre-Tokenizer Hashes
 
 on:
     push:
+        branches-ignore: ['hrx-integration']
         paths:
             - 'convert_hf_to_gguf.py'
             - 'convert_hf_to_gguf_update.py'
     pull_request:
+        branches-ignore: ['hrx-integration']
         paths:
             - 'convert_hf_to_gguf.py'
             - 'convert_hf_to_gguf_update.py'

.github/workflows/python-check-requirements.yml

@@ -2,12 +2,14 @@ name: Python check requirements.txt
 
 on:
   push:
+    branches-ignore: ['hrx-integration']
     paths:
       - '.github/workflows/python-check-requirements.yml'
       - 'scripts/check-requirements.sh'
       - 'convert*.py'
       - '**/requirements*.txt'
   pull_request:
+    branches-ignore: ['hrx-integration']
     paths:
       - '.github/workflows/python-check-requirements.yml'
       - 'scripts/check-requirements.sh'

.github/workflows/python-lint.yml

@@ -9,6 +9,7 @@ on:
       '**/*.py'
     ]
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/python-lint.yml',

.github/workflows/python-type-check.yml

@@ -2,13 +2,15 @@ name: Python Type-Check
 
 on:
   push:
+    branches-ignore: ['hrx-integration']
     paths:
       - '.github/workflows/python-type-check.yml'
       - 'ty.toml'
       - '**.py'
       - '**/requirements*.txt'
       # - 'pyrightconfig.json'
   pull_request:
+    branches-ignore: ['hrx-integration']
     paths:
       - '.github/workflows/python-type-check.yml'
       - 'ty.toml'

.github/workflows/server-webui.yml

@@ -17,6 +17,7 @@ on:
       'tools/server/public/**'
     ]
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/server-webui.yml',

.github/workflows/server.yml

@@ -28,6 +28,7 @@ on:
       'tools/server/**.*'
     ]
   pull_request:
+    branches-ignore: ['hrx-integration']
     types: [opened, synchronize, reopened]
     paths: [
       '.github/workflows/server.yml',

.github/workflows/update-ops-docs.yml

@@ -2,11 +2,13 @@ name: Update Operations Documentation
 
 on:
     push:
+        branches-ignore: ['hrx-integration']
         paths:
             - 'docs/ops.md'
             - 'docs/ops/**'
             - 'scripts/create_ops_docs.py'
     pull_request:
+        branches-ignore: ['hrx-integration']
         paths:
             - 'docs/ops.md'
             - 'docs/ops/**'

ggml/src/ggml-hrx/ggml-hrx.cpp

@@ -2521,7 +2521,7 @@ static bool ggml_backend_hrx_load_catalog_provider(
                         executable, export_ordinal, &export_info)) &&
                     export_info.binding_count == entry->binding_count &&
                     export_info.parameter_count == entry->parameter_count &&
-                    export_info.constant_count * sizeof(uint32_t) == entry->constants_size;
+                    export_info.constant_byte_length == entry->constants_size;
     if (!ok) {
         GGML_LOG_WARN(
             "%s: HRX catalog kernel %s has unsupported ABI "
@@ -2530,7 +2530,7 @@ static bool ggml_backend_hrx_load_catalog_provider(
             entry->name,
             export_info.binding_count,
             entry->binding_count,
-            export_info.constant_count,
+            export_info.constant_byte_length,
             entry->constants_size,
             export_info.parameter_count,
             entry->parameter_count,