[BUG]: Memory access error for thrust::sort
- Dominant language
- C++
- Stars
- 2.5k
- Forks
- 486
- Avg merge
- 2d 6h
- Merged PRs (30d)
- 295
Description
### Is this a duplicate?
- [X] I confirmed there appear to be no [duplicate issues](https://github.com/NVIDIA/cccl/issues) for this bug and that I agree to the [Code of Conduct](CODE_OF_CONDUCT.md)
### Type of Bug
Runtime Error
### Component
Thrust
### Describe the bug
When compiling the following code https://gist.github.com/PhilipDeegan/ac3931cccb48e0e61e742fc2b731c4c7
on ubuntu 22.04
with command
```
clang++-17 --std=c++17 -fPIC -fsized-deallocation -Wno-unknown-cuda-version -Xclang -fcuda-allow-variadic-functions -x cuda --cuda-gpu-arch=sm_80 -O3 -o sorting_thrust sorting_thrust.cpp
```
I see
```
terminate called after throwing an instance of 'thrust::system::system_error'
what(): merge_sort: failed to synchronize: cudaErrorIllegalAddress: an illegal memory access was encountered
```
However:
I do not see this issue with thrust via AMD.
I do not see the issue with `-O0`
I do not see the issue if I compile with the define `-DDO_NOT_ADD_EB`
I continue to see no issues with the define `-DDO_NOT_ADD_EB` even with 2e7 items
And I do not see issues with only 1e3 items.
### How to Reproduce
Compile with previous instructions
### Expected behavior
No memory access error
----
godbolt doesn't support cuda clang 17`
### Reproduction link
_No response_
### Operating System
Ubuntu 22.04
### nvidia-smi output
Mon Nov 13 22:44:28 2023
+---------------------------------------------------------------------------------------+
| NVIDIA-SMI 545.23.06 Driver Version: 545.23.06 CUDA Version: 12.3 |
|-----------------------------------------+----------------------+----------------------+
| GPU Name Persistence-M | Bus-Id Disp.A | Volatile Uncorr. ECC |
| Fan Temp Perf Pwr:Usage/Cap | Memory-Usage | GPU-Util Compute M. |
| | | MIG M. |
|=========================================+======================+======================|
| 0 NVIDIA A100-PCIE-40GB Off | 00000000:21:00.0 Off | 0 |
| N/A 69C P0 246W / 250W | 1171MiB / 40960MiB | 99% Default |
| | | Disabled |
+-----------------------------------------+----------------------+----------------------+
+---------------------------------------------------------------------------------------+
| Processes: |
| GPU GI CI PID Type Process name GPU Memory |
| ID ID Usage |
|=======================================================================================|
+---------------------------------------------------------------------------------------+
### NVCC version
```
nvcc --version
nvcc: NVIDIA (R) Cuda compiler driver
Copyright (c) 2005-2023 NVIDIA Corporation
Built on Fri_Sep__8_19:17:24_PDT_2023
Cuda compilation tools, release 12.3, V12.3.52
Build cuda_12.3.r12.3/compiler.33281558_0
```
```
clang++-17 -v
Ubuntu clang version 17.0.5 (++20231113073440+094cfd14c813-1~exp1~20231113073538.63)
Target: x86_64-pc-linux-gnu
Thread model: posix
InstalledDir: /usr/bin
Found candidate GCC installation: /usr/bin/../lib/gcc/x86_64-linux-gnu/11
Selected GCC installation: /usr/bin/../lib/gcc/x86_64-linux-gnu/11
Candidate multilib: .;@m64
Selected multilib: .;@m64
Found CUDA installation: /usr/local/cuda-12.3, version
```
Contributor guide
Assessment
This issue has not been assessed yet.