CI: nightly-cuda-core + nightly-numba-cuda-mlir + PTDS fix (#2283)

* CI: add nightly-cuda-core and nightly-numba-cuda-mlir modes

nightly-cuda-core: test the released cuda-core from PyPI against
main-built pathfinder and cuda-bindings, catching the "core released ×
bindings main" gap documented in issue #1955. Runs on linux-64 (a100)
and win-64 (a100 MCDM).

nightly-numba-cuda-mlir: MLIR-backend companion to nightly-numba-cuda.
Installs main pathfinder+bindings+core plus numba-cuda-mlir from PyPI,
runs numba-cuda-mlir's own test suite from the matching git tag.
Linux amd64/arm64 x CUDA 12.9.1 / 13.3.0.

Both modes fetch the released version's tests from git tags because
the respective wheels do not ship test_*.py files. Includes
tag-not-found fallback (log warning + exit 0) to avoid red-lining the
nightly on a freshly-cut PyPI release that hasn't been pushed to git
yet.

* ci/test-matrix.yml: fix CUDA_PYTHON_CUDA_PER_THREAD_DEFAULT_STREAM typo

The two ENV overrides intended to exercise the per-thread default
stream code path were misspelled (missing the CUDA_ segment), so the
env var was silently ignored and the PTDS coverage added in #1972 had
no effect. Rename to the correct
CUDA_PYTHON_CUDA_PER_THREAD_DEFAULT_STREAM.

Refs #971.

* cuda_pathfinder: pin nvshmem to <3.7 (was previously excluding only 3.7.0)

nvidia-nvshmem-cu{12,13} 3.7.x breaks the main branch, not only 3.7.0. Widen the exclusion from an exact-version bump to <3.7 so 3.7.x and above are avoided until we can move forward.

* nightly-numba-cuda-mlir: swap arm64 for win-64 coverage, use rtxpro6000

Drop the linux-aarch64 rows and instead add win-64 coverage with the
same CUDA 12.9.1 / 13.3.0 pair. Switch all four rows from GPU l4 to
rtxpro6000. Windows rows use DRIVER_MODE MCDM, matching the existing
rtxpro6000 CUDA 13.3.0 patterns.

* Temporarily add push trigger to ci-nightly.yml for testing

Remove before merging.

* CI: switch nightly-{cuda-core,numba-cuda-mlir} to actions/checkout for tests

The initial approach used git inside the ubuntu:24.04 container to fetch
the released version's test suite, but git is not installed on that
container (install_unix_deps only pulls in jq/wget/g++/etc.) and its
absence made the run steps silently skip via the tag-not-fetchable
fallback. On Windows, git archive of just the cuda_core subtree also hit
a dangling-symlink extraction failure (cuda_core/.git_archival.txt).

Refactor to:

- run-tests: just install wheels and expose the resolved release version
  (CUDA_CORE_RELEASED_VER / NUMBA_CUDA_MLIR_VER) and cuda-core test-group
  name via GITHUB_ENV. No more git operations.
- test-wheel-{linux,windows}.yml: add an actions/checkout step per mode
  that pulls the matching release tag into a subdirectory
  (cuda-core-released / numba-cuda-mlir-released), then the follow-up
  test step installs that tag's test dep-group and runs pytest.

For numba-cuda-mlir also pass --ignore=tests/benchmarks
--ignore=tests/doc_examples to pytest: those directories import the
`numba` package at module top and would fail collection, which is
cuSIMT's expected behavior (see NVIDIA/numba-cuda-mlir#136 — cuSIMT
intentionally does not depend on numba).

* CI: pin numpy<2.5 (mlir) and pytest<9.1 (cuda-core released tests)

Two nightly failure fixups after the first green iteration:

nightly-numba-cuda-mlir: numba-cuda-mlir 0.4.0 has an inverted guard
that registers an overload of np.row_stack on NumPy 2.x, and NumPy 2.5
removed that name entirely, so test collection fails with
"AttributeError: module 'numpy' has no attribute 'row_stack'". Cap
numpy to <2.5. See NVIDIA/numba-cuda-mlir#154.

nightly-cuda-core: released cuda-core v1.0.1's test suite uses a
parametrize argvalues pattern that pytest 9.1 rejects
("in parametrize the number of names (1)... must be equal to the
number of values (3)"). The main-side fix was #2212 but it has not
shipped in a cuda-core release yet. Cap pytest to <9.1 for the
released-cuda-core test run only.

* CI: deselect known pre-existing failures in nightly-cuda-core and nightly-numba-cuda-mlir

Applied only in the affected nightly-* pytest invocations; the released
source trees under test are unmodified.

nightly-numba-cuda-mlir (all 10 tests deselected are from cuSIMT):

  * CudaArraySetting::{test_no_sync_default_stream, test_no_sync_supplied_stream, test_sync}
    TestCudaArrayInterface::{test_consume_no_sync, test_consume_sync,
                             test_launch_no_sync, test_launch_sync,
                             test_launch_sync_two_streams, test_fortran_contiguous}
      Serial-pytest contamination of numba_cuda_mlir.cuda.cudadrv from an
      xfailed test in test_nrt_comprehensive.py. Upstream CI runs with
      `pytest -n auto --dist loadscope`, which isolates the offending
      side effect in a separate xdist worker; our nightly runs serially
      and hits the pollution. See NVIDIA/numba-cuda-mlir#135.
  * TestLinkerDumpAssembly::test_nvjitlink_jit_with_linkable_code_lto_dump_assembly_warn
      Subprocess-invokes `cuobjdump`, which isn't on PATH in the base
      ubuntu:24.04 container. Filed as an upstream skip-guard bug.

nightly-cuda-core (3 tests deselected are pre-existing v1.0.1 issues):

  * test_enum_coverage.py::test_wrapper_covers_all_binding_members[NvlinkVersion]
      Expected drift: main cuda-bindings adds NvlinkVersion.VERSION_6_0
      which v1.0.1's wrapper mapping predates. This mode intentionally
      pairs released core with main bindings, so this coverage-style
      test will stay red here until a cuda-core release catches up.
  * test_rlcompleter_patch.py::test_opt_out_env_var_disables_patch_even_when_interactive
      Environment-dependent test: expects rlcompleter to crash without
      the tab-completion patch, but on Windows MCDM the pre-patch
      behavior is clean. Passes on Linux, fails on Windows MCDM.
  * test_memory.py::test_non_managed_resources_report_not_managed[pinned]
      Same underlying "Failed to allocate memory from pool" error that
      v1.0.1 already xfails in the sibling test_pinned_memory_resource_initialization
      (TODO(#9999)). cuda-python main has since fixed the parametrized
      case to route through _allocate_pinned_buffer_or_xfail(), but that
      fix hasn't shipped in a cuda-core release yet.

* CI: tighten deselects to per-platform failing sets

Previously applied the same list on both Linux and Windows workflows,
which over-deselected — some tests only fail on one platform because
the underlying issues (serial-pytest test-order in mlir, MCDM-only
behavior in cuda-core) are platform-specific.

Now:

nightly-numba-cuda-mlir
  linux-64: TestCudaArrayInterface::{test_consume_no_sync,
    test_consume_sync, test_launch_no_sync, test_launch_sync,
    test_launch_sync_two_streams, test_fortran_contiguous}
    + TestLinkerDumpAssembly::test_nvjitlink_jit_with_linkable_code_lto_dump_assembly_warn.
  win-64: CudaArraySetting::{test_no_sync_default_stream,
    test_no_sync_supplied_stream, test_sync}
    + TestCudaArrayInterface::test_fortran_contiguous.

Test-order contamination in numba-cuda-mlir#135 surfaces different
tests depending on collection order (linux-64 vs win-64 exercise
different subsets), so the per-platform lists differ. cuobjdump-based
TestLinkerDumpAssembly only fires on Linux because the ubuntu:24.04
container's PATH lacks cuobjdump; Windows runners ship it with the
local CTK.

nightly-cuda-core
  linux-64: test_enum_coverage.py::test_wrapper_covers_all_binding_members[NvlinkVersion].
  win-64: NvlinkVersion (same as Linux)
    + test_rlcompleter_patch.py::test_opt_out_env_var_disables_patch_even_when_interactive
    + test_memory.py::test_non_managed_resources_report_not_managed[pinned].

rlcompleter and pinned mempool tests only fail on Windows MCDM.
NvlinkVersion fails on both (expected drift for the mode).

* CI: version-gate the nightly-mode deselects so they auto-clean

Each deselect is now wrapped in a bash conditional keyed on the
installed release version. When a newer numba-cuda-mlir or cuda-core
release ships with the referenced fix, the nightly picks it up
automatically, the guard evaluates false, and the deselect drops — so
the tests run against the new release. If they still fail we hear
about it loudly rather than silently masking a regression.

Current guards:
- numba-cuda-mlir #135 tests + cuobjdump TestLinkerDumpAssembly:
  applied when installed numba-cuda-mlir version <= 0.4.0.
- cuda-core NvlinkVersion / rlcompleter opt-out / pinned mempool:
  applied when installed cuda-core version <= 1.0.1.

Structure keeps one conditional block per (mode, platform) with a
comment above each deselect explaining the tracking issue.

* CI: broaden mlir deselect list to full #135 union across platforms

The previous per-platform-tight lists were incomplete: NVIDIA/numba-cuda-mlir#135's
import-time contamination poisons whichever tests reference
cuda.cudadrv.driver AFTER the polluting xfail runs, and collection
order varies between runs. Two consecutive Windows CI runs failed on
different subsets (3 slicing tests one run, 5 interface tests the
next).

Deselect the full union of #135-listed tests + test_fortran_contiguous
(observed to hit the same contamination) on both Linux and Windows.
Same version guard (<= 0.4.0) still applies, so the whole block drops
automatically when a newer numba-cuda-mlir release ships with the
root-cause fix.

Linux keeps the extra cuobjdump deselect (Linux-only environment
issue).

* Revert "cuda_pathfinder: pin nvshmem to <3.7 (was previously excluding only 3.7.0)"

This reverts commit 2a42aa7.

* Revert "Temporarily add push trigger to ci-nightly.yml for testing"

This reverts commit a0ccd19.

Author: leofang

.github/workflows/ci-nightly.yml

@@ -209,6 +209,82 @@ jobs:
       test-mode: nightly-numba-cuda
       matrix_filter: 'map(select(.ENV.MODE == "nightly-numba-cuda"))'
 
+  # ── numba-cuda-mlir tests ──
+
+  test-numba-cuda-mlir-linux-64:
+    name: "Nightly numba-cuda-mlir (linux-64)"
+    if: ${{ github.repository_owner == 'nvidia' }}
+    needs: find-wheels
+    permissions:
+      contents: read
+      actions: read
+    secrets: inherit
+    uses: ./.github/workflows/test-wheel-linux.yml
+    with:
+      build-type: nightly
+      host-platform: linux-64
+      build-ctk-ver: ${{ needs.find-wheels.outputs.CUDA_BUILD_VER }}
+      run-id: ${{ needs.find-wheels.outputs.RUN_ID }}
+      sha: ${{ needs.find-wheels.outputs.HEAD_SHA }}
+      test-mode: nightly-numba-cuda-mlir
+      matrix_filter: 'map(select(.ENV.MODE == "nightly-numba-cuda-mlir"))'
+
+  test-numba-cuda-mlir-windows:
+    name: "Nightly numba-cuda-mlir (win-64)"
+    if: ${{ github.repository_owner == 'nvidia' }}
+    needs: find-wheels
+    permissions:
+      contents: read
+      actions: read
+    secrets: inherit
+    uses: ./.github/workflows/test-wheel-windows.yml
+    with:
+      build-type: nightly
+      host-platform: win-64
+      build-ctk-ver: ${{ needs.find-wheels.outputs.CUDA_BUILD_VER }}
+      run-id: ${{ needs.find-wheels.outputs.RUN_ID }}
+      sha: ${{ needs.find-wheels.outputs.HEAD_SHA }}
+      test-mode: nightly-numba-cuda-mlir
+      matrix_filter: 'map(select(.ENV.MODE == "nightly-numba-cuda-mlir"))'
+
+  # ── Released cuda-core against main pathfinder/bindings ──
+
+  test-cuda-core-linux-64:
+    name: "Nightly cuda-core (linux-64)"
+    if: ${{ github.repository_owner == 'nvidia' }}
+    needs: find-wheels
+    permissions:
+      contents: read
+      actions: read
+    secrets: inherit
+    uses: ./.github/workflows/test-wheel-linux.yml
+    with:
+      build-type: nightly
+      host-platform: linux-64
+      build-ctk-ver: ${{ needs.find-wheels.outputs.CUDA_BUILD_VER }}
+      run-id: ${{ needs.find-wheels.outputs.RUN_ID }}
+      sha: ${{ needs.find-wheels.outputs.HEAD_SHA }}
+      test-mode: nightly-cuda-core
+      matrix_filter: 'map(select(.ENV.MODE == "nightly-cuda-core"))'
+
+  test-cuda-core-windows:
+    name: "Nightly cuda-core (win-64)"
+    if: ${{ github.repository_owner == 'nvidia' }}
+    needs: find-wheels
+    permissions:
+      contents: read
+      actions: read
+    secrets: inherit
+    uses: ./.github/workflows/test-wheel-windows.yml
+    with:
+      build-type: nightly
+      host-platform: win-64
+      build-ctk-ver: ${{ needs.find-wheels.outputs.CUDA_BUILD_VER }}
+      run-id: ${{ needs.find-wheels.outputs.RUN_ID }}
+      sha: ${{ needs.find-wheels.outputs.HEAD_SHA }}
+      test-mode: nightly-cuda-core
+      matrix_filter: 'map(select(.ENV.MODE == "nightly-cuda-core"))'
+
   # ── Standard tests on nightly-only runners ──
 
   test-standard-linux-aarch64:
@@ -244,6 +320,10 @@ jobs:
       - test-numba-cuda-linux-64
       - test-numba-cuda-linux-aarch64
       - test-numba-cuda-windows
+      - test-numba-cuda-mlir-linux-64
+      - test-numba-cuda-mlir-windows
+      - test-cuda-core-linux-64
+      - test-cuda-core-windows
       - test-standard-linux-aarch64
     steps:
       - name: Exit
@@ -270,6 +350,14 @@ jobs:
                  needs.test-numba-cuda-linux-aarch64.result == 'failure' ||
                  needs.test-numba-cuda-windows.result == 'cancelled' ||
                  needs.test-numba-cuda-windows.result == 'failure' ||
+                 needs.test-numba-cuda-mlir-linux-64.result == 'cancelled' ||
+                 needs.test-numba-cuda-mlir-linux-64.result == 'failure' ||
+                 needs.test-numba-cuda-mlir-windows.result == 'cancelled' ||
+                 needs.test-numba-cuda-mlir-windows.result == 'failure' ||
+                 needs.test-cuda-core-linux-64.result == 'cancelled' ||
+                 needs.test-cuda-core-linux-64.result == 'failure' ||
+                 needs.test-cuda-core-windows.result == 'cancelled' ||
+                 needs.test-cuda-core-windows.result == 'failure' ||
                  needs.test-standard-linux-aarch64.result == 'cancelled' ||
                  needs.test-standard-linux-aarch64.result == 'failure' }}; then
             exit 1

.github/workflows/test-wheel-linux.yml

@@ -37,8 +37,9 @@ on:
         default: ''
       test-mode:
         description: >
-          Test mode: 'standard' (default), 'nightly-pytorch', or
-          'nightly-numba-cuda'.
+          Test mode: 'standard' (default), 'nightly-pytorch',
+          'nightly-numba-cuda', 'nightly-numba-cuda-mlir', or
+          'nightly-cuda-core'.
         type: string
         default: 'standard'
       sha:
@@ -409,6 +410,20 @@ jobs:
           LOCAL_CTK: ${{ matrix.LOCAL_CTK }}
         run: run-tests nightly-numba-cuda
 
+      - name: Install cuda-python wheels + numba-cuda-mlir
+        if: ${{ inputs.test-mode == 'nightly-numba-cuda-mlir' }}
+        env:
+          CUDA_VER: ${{ matrix.CUDA_VER }}
+          LOCAL_CTK: ${{ matrix.LOCAL_CTK }}
+        run: run-tests nightly-numba-cuda-mlir
+
+      - name: Install main pathfinder/bindings + released cuda-core
+        if: ${{ inputs.test-mode == 'nightly-cuda-core' }}
+        env:
+          CUDA_VER: ${{ matrix.CUDA_VER }}
+          LOCAL_CTK: ${{ matrix.LOCAL_CTK }}
+        run: run-tests nightly-cuda-core
+
       # ── Nightly: run tests ──
       - name: Run PyTorch interop tests
         if: ${{ inputs.test-mode == 'nightly-pytorch' }}
@@ -420,3 +435,98 @@ jobs:
       - name: Run numba-cuda tests
         if: ${{ inputs.test-mode == 'nightly-numba-cuda' }}
         run: python -m numba.runtests numba.cuda.tests
+
+      - name: Checkout numba-cuda-mlir tests at matching tag
+        if: ${{ inputs.test-mode == 'nightly-numba-cuda-mlir' && env.NUMBA_CUDA_MLIR_VER != '' }}
+        uses: actions/checkout@df4cb1c069e1874edd31b4311f1884172cec0e10  # v6.0.3
+        with:
+          repository: NVIDIA/numba-cuda-mlir
+          ref: v${{ env.NUMBA_CUDA_MLIR_VER }}
+          path: numba-cuda-mlir-released
+
+      - name: Run numba-cuda-mlir tests
+        if: ${{ inputs.test-mode == 'nightly-numba-cuda-mlir' && env.NUMBA_CUDA_MLIR_VER != '' }}
+        run: |
+          pushd numba-cuda-mlir-released
+          # Install this tag's test deps (pytest + plugins + ml-dtypes + ...).
+          pip install --upgrade "pip>=25.1"
+          pip install --group test
+          # Skip tests/benchmarks/ and tests/doc_examples/ — they import the
+          # numba package at collection time, which cuSIMT intentionally does
+          # not depend on. See NVIDIA/numba-cuda-mlir#136.
+          #
+          # Version-gated deselects: when a newer numba-cuda-mlir release
+          # ships with the referenced fix, the guard evaluates false and the
+          # tests get run automatically. If they still fail on the newer
+          # version we hear about it loudly (rather than silently masking).
+          DESELECTS=()
+          if python -c "from packaging.version import Version; import sys; sys.exit(0 if Version('${NUMBA_CUDA_MLIR_VER}') <= Version('0.4.0') else 1)"; then
+            # NVIDIA/numba-cuda-mlir#135: serial-pytest contamination of
+            # numba_cuda_mlir.cuda.cudadrv from an xfailed test in
+            # test_nrt_comprehensive.py contaminates any later test that
+            # touches cuda.cudadrv.driver. Upstream CI hides it via
+            # `-n auto --dist loadscope`. Which specific tests fail depends
+            # on collection order (we saw different subsets on linux-64 vs
+            # win-64 across runs), so we deselect the union of all tests
+            # #135 lists as vulnerable + test_fortran_contiguous (observed
+            # to hit the same contamination in our runs).
+            #
+            # test_nvjitlink_jit_with_linkable_code_lto_dump_assembly_warn:
+            # subprocess-invokes `cuobjdump`, not on PATH in the base
+            # ubuntu:24.04 container. (Linux-only; Windows runners ship
+            # cuobjdump with the local CTK. No upstream fix yet — pending
+            # a skip-guard bug to be filed against NVIDIA/numba-cuda-mlir.)
+            DESELECTS+=(
+              --deselect 'tests/numba_cuda_tests/cudadrv/test_cuda_array_slicing.py::CudaArraySetting::test_no_sync_default_stream'
+              --deselect 'tests/numba_cuda_tests/cudadrv/test_cuda_array_slicing.py::CudaArraySetting::test_no_sync_supplied_stream'
+              --deselect 'tests/numba_cuda_tests/cudadrv/test_cuda_array_slicing.py::CudaArraySetting::test_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_consume_no_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_consume_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_launch_no_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_launch_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_launch_sync_two_streams'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_fortran_contiguous'
+              --deselect 'tests/numba_cuda_tests/cudadrv/test_nvjitlink.py::TestLinkerDumpAssembly::test_nvjitlink_jit_with_linkable_code_lto_dump_assembly_warn'
+            )
+          fi
+          pytest -rxXs -v --durations=0 \
+            --ignore=tests/benchmarks \
+            --ignore=tests/doc_examples \
+            "${DESELECTS[@]}" \
+            tests/
+          popd
+
+      - name: Checkout released cuda-core tests at matching tag
+        if: ${{ inputs.test-mode == 'nightly-cuda-core' && env.CUDA_CORE_RELEASED_VER != '' }}
+        uses: actions/checkout@df4cb1c069e1874edd31b4311f1884172cec0e10  # v6.0.3
+        with:
+          ref: cuda-core-v${{ env.CUDA_CORE_RELEASED_VER }}
+          path: cuda-core-released
+
+      - name: Run released cuda-core tests
+        if: ${{ inputs.test-mode == 'nightly-cuda-core' && env.CUDA_CORE_RELEASED_VER != '' }}
+        run: |
+          pushd cuda-core-released/cuda_core
+          # Install the released tag's test group so we exercise the exact deps
+          # that cuda-core version shipped with.
+          pip install --upgrade "pip>=25.1"
+          pip install --group "${CUDA_CORE_TEST_GROUP}"
+          # Cap pytest below 9.1: released cuda-core <=1.0.1 has parametrize
+          # patterns that pytest 9.1 rejects; the main-side fix (#2212) has
+          # not yet shipped in a cuda-core release.
+          pip install "pytest<9.1"
+          # Version-gated deselect: drops automatically when a newer
+          # cuda-core release with the wrapper-mapping update ships.
+          DESELECTS=()
+          if python -c "from packaging.version import Version; import sys; sys.exit(0 if Version('${CUDA_CORE_RELEASED_VER}') <= Version('1.0.1') else 1)"; then
+            # NvlinkVersion: v1.0.1's wrapper mapping predates
+            # NvlinkVersion.VERSION_6_0 which main cuda-bindings adds.
+            # Expected drift on this mode until released cuda-core catches up.
+            DESELECTS+=(
+              --deselect 'tests/test_enum_coverage.py::test_wrapper_covers_all_binding_members[NvlinkVersion]'
+            )
+          fi
+          pytest -rxXs -v --durations=0 --randomly-dont-reorganize \
+            "${DESELECTS[@]}" \
+            tests/
+          popd

.github/workflows/test-wheel-windows.yml

@@ -37,8 +37,9 @@ on:
         default: ''
       test-mode:
         description: >
-          Test mode: 'standard' (default), 'nightly-pytorch', or
-          'nightly-numba-cuda'.
+          Test mode: 'standard' (default), 'nightly-pytorch',
+          'nightly-numba-cuda', 'nightly-numba-cuda-mlir', or
+          'nightly-cuda-core'.
         type: string
         default: 'standard'
       sha:
@@ -388,6 +389,22 @@ jobs:
         shell: bash --noprofile --norc -xeuo pipefail {0}
         run: run-tests nightly-numba-cuda
 
+      - name: Install cuda-python wheels + numba-cuda-mlir
+        if: ${{ inputs.test-mode == 'nightly-numba-cuda-mlir' }}
+        env:
+          CUDA_VER: ${{ matrix.CUDA_VER }}
+          LOCAL_CTK: ${{ matrix.LOCAL_CTK }}
+        shell: bash --noprofile --norc -xeuo pipefail {0}
+        run: run-tests nightly-numba-cuda-mlir
+
+      - name: Install main pathfinder/bindings + released cuda-core
+        if: ${{ inputs.test-mode == 'nightly-cuda-core' }}
+        env:
+          CUDA_VER: ${{ matrix.CUDA_VER }}
+          LOCAL_CTK: ${{ matrix.LOCAL_CTK }}
+        shell: bash --noprofile --norc -xeuo pipefail {0}
+        run: run-tests nightly-cuda-core
+
       # ── Nightly: run tests ──
       - name: Run PyTorch interop tests
         if: ${{ inputs.test-mode == 'nightly-pytorch' }}
@@ -401,3 +418,85 @@ jobs:
         if: ${{ inputs.test-mode == 'nightly-numba-cuda' }}
         shell: bash --noprofile --norc -xeuo pipefail {0}
         run: python -m numba.runtests numba.cuda.tests
+
+      - name: Checkout numba-cuda-mlir tests at matching tag
+        if: ${{ inputs.test-mode == 'nightly-numba-cuda-mlir' && env.NUMBA_CUDA_MLIR_VER != '' }}
+        uses: actions/checkout@df4cb1c069e1874edd31b4311f1884172cec0e10  # v6.0.3
+        with:
+          repository: NVIDIA/numba-cuda-mlir
+          ref: v${{ env.NUMBA_CUDA_MLIR_VER }}
+          path: numba-cuda-mlir-released
+
+      - name: Run numba-cuda-mlir tests
+        if: ${{ inputs.test-mode == 'nightly-numba-cuda-mlir' && env.NUMBA_CUDA_MLIR_VER != '' }}
+        shell: bash --noprofile --norc -xeuo pipefail {0}
+        run: |
+          pushd numba-cuda-mlir-released
+          pip install --upgrade "pip>=25.1"
+          pip install --group test
+          # Version-gated deselects — dropped automatically when newer
+          # cuSIMT release ships. See linux step for full rationale.
+          # NVIDIA/numba-cuda-mlir#135 poisons a subset of tests that
+          # varies across runs based on collection order, so we deselect
+          # the full union rather than trying to enumerate what happened
+          # to fail on the most recent nightly.
+          DESELECTS=()
+          if python -c "from packaging.version import Version; import sys; sys.exit(0 if Version('${NUMBA_CUDA_MLIR_VER}') <= Version('0.4.0') else 1)"; then
+            DESELECTS+=(
+              --deselect 'tests/numba_cuda_tests/cudadrv/test_cuda_array_slicing.py::CudaArraySetting::test_no_sync_default_stream'
+              --deselect 'tests/numba_cuda_tests/cudadrv/test_cuda_array_slicing.py::CudaArraySetting::test_no_sync_supplied_stream'
+              --deselect 'tests/numba_cuda_tests/cudadrv/test_cuda_array_slicing.py::CudaArraySetting::test_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_consume_no_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_consume_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_launch_no_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_launch_sync'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_launch_sync_two_streams'
+              --deselect 'tests/numba_cuda_tests/cudapy/test_cuda_array_interface.py::TestCudaArrayInterface::test_fortran_contiguous'
+            )
+          fi
+          pytest -rxXs -v --durations=0 \
+            --ignore=tests/benchmarks \
+            --ignore=tests/doc_examples \
+            "${DESELECTS[@]}" \
+            tests/
+          popd
+
+      - name: Checkout released cuda-core tests at matching tag
+        if: ${{ inputs.test-mode == 'nightly-cuda-core' && env.CUDA_CORE_RELEASED_VER != '' }}
+        uses: actions/checkout@df4cb1c069e1874edd31b4311f1884172cec0e10  # v6.0.3
+        with:
+          ref: cuda-core-v${{ env.CUDA_CORE_RELEASED_VER }}
+          path: cuda-core-released
+
+      - name: Run released cuda-core tests
+        if: ${{ inputs.test-mode == 'nightly-cuda-core' && env.CUDA_CORE_RELEASED_VER != '' }}
+        shell: bash --noprofile --norc -xeuo pipefail {0}
+        run: |
+          pushd cuda-core-released/cuda_core
+          pip install --upgrade "pip>=25.1"
+          pip install --group "${CUDA_CORE_TEST_GROUP}"
+          # Cap pytest below 9.1 — released cuda-core <=1.0.1 has parametrize
+          # patterns that pytest 9.1 rejects (see #2212).
+          pip install "pytest<9.1"
+          # Version-gated deselects — dropped automatically when a newer
+          # cuda-core release ships. See linux step for full rationale on
+          # NvlinkVersion. The Windows-only tests are:
+          # - test_rlcompleter_patch: env-dependent expectation that
+          #   passes on Linux, fails on Windows MCDM.
+          # - test_non_managed_resources_report_not_managed[pinned]: same
+          #   MCDM mempool OOM v1.0.1 already xfails in
+          #   test_pinned_memory_resource_initialization (TODO(#9999));
+          #   main fixed the parametrized case via #2139 but v1.0.1 lacks
+          #   the fix.
+          DESELECTS=()
+          if python -c "from packaging.version import Version; import sys; sys.exit(0 if Version('${CUDA_CORE_RELEASED_VER}') <= Version('1.0.1') else 1)"; then
+            DESELECTS+=(
+              --deselect 'tests/test_enum_coverage.py::test_wrapper_covers_all_binding_members[NvlinkVersion]'
+              --deselect 'tests/test_rlcompleter_patch.py::test_opt_out_env_var_disables_patch_even_when_interactive'
+              --deselect 'tests/test_memory.py::test_non_managed_resources_report_not_managed[pinned]'
+            )
+          fi
+          pytest -rxXs -v --durations=0 --randomly-dont-reorganize \
+            "${DESELECTS[@]}" \
+            tests/
+          popd