onnxruntime
6e19374b - Fix CUDA 13 build error in gqa_unfused_attention.cu (#28309)

Commit
113 days ago
Fix CUDA 13 build error in gqa_unfused_attention.cu (#28309) ### Description Fix CUDA 13 build failure introduced by PR #28198 (commit 997c4798). **Root cause:** `gqa_unfused_attention.cu` directly includes `<cub/cub.cuh>`, which on CUDA 13.x transitively pulls in CCCL's `tcgen05_ld.h`. That header uses `__out` as a parameter name in inline PTX assembly, but on Windows MSVC the SAL annotation macro `#define __out` expands it, turning `__out[0]` into `[0]` — causing a parse error. **Fix:** Changed `#include <cub/cub.cuh>` to `#include "core/providers/cuda/cu_inc/cub.cuh"` — the existing ORT wrapper that `#undef __out` before including CUB. ### Motivation and Context The CUDA 13 packaging pipeline (`py-cuda13-packaging-pipeline.yml`) has been failing since PR #28198 was merged, with errors like: ``` E:/_work/_temp/v13.0/include/cccl/cuda/__ptx/instructions/generated/tcgen05_ld.h(20): error : expected an identifier asm("tcgen05.ld.sync.aligned.16x64b.x1.b32 {%0}, [%1];" : "=r"( [0]) : "r"(__taddr) : "memory"); ```
Author
Parents
Loading