Skip to content

SCALE's cub wrapper header breaks thrust against CUDA 13.1 CCCL #52

Description

@diskdog

SCALE 1.7.1 and the 2026.07.02 nightly (clang 20.1.8), NVIDIA target, CUDA 13.1 toolkit, sm_75, compile-only.

Any file that includes a thrust container header fails under SCALE, with the error inside SCALE's own compatibility shim:

#include <thrust/device_vector.h>
int main() { thrust::device_vector<int> v(4); return 0; }
  • NVIDIA nvcc -arch=sm_75 -std=c++17 -c repro.cu: ok
  • SCALE llvm/bin/nvcc -arch=sm_75 -std=c++17 -c repro.cu (1.7.1 and nightly, same error):
llvm/lib/clang/20/include/cuda_wrappers/cub/util_device.cuh:7:26: error: explicit
instantiation of 'EmptyKernel' does not refer to a function template, ...

Cause: the wrapper does

CUB_NAMESPACE_BEGIN
template __global__ void EmptyKernel<void>();
CUB_NAMESPACE_END

but in the CCCL that ships with CUDA 13.1, EmptyKernel lives in cub::detail (include/cccl/cub/util_device.cuh:83), not at cub:: scope, so the explicit instantiation names a template that no longer exists there. The wrapper predates the CCCL move; detail::EmptyKernel<void> is what it wants now.

Footprint in the validation sweep: every repo that touches thrust via CUDA 13.1 dies on this (8 of the first 122 swept repos: Alenka, cs344, multem, NKSR, nvParse, PyBindToGPUs, SIGA, and a Multigrid solver). In larger translation units the error is often followed by the issue #43 assertion crash, so this wrapper error is also the gateway to a chunk of the #43 noise.

Metadata

Metadata

Assignees

Labels

VerifiedInternal QA verifies the bug reproduces on the current dev build.bugSomething isn't working

Type

No type

Projects

No projects

Milestone

No milestone

Relationships

None yet

Development

No branches or pull requests

Issue actions