feat(rocthrust): CCCL 3.0.x changes by yungshengtu · Pull Request #3773 · ROCm/rocm-libraries

yungshengtu · 2026-01-12T15:36:47Z

Most of the changes here are pulling in the commits from CCCL 3.0.3.
With the fixes and workarounds attached to these commits.

In addition, this PR also includes:

upgrading the versions of dependencies
moving all libcxx wrapper codes into thrust/detail/libcxx_wrapper
C++20 onwards support
fixing the errors from thrust/unique_ptr.h
fixing the errors from test/generate_resource_spec.cpp

JIRA ID : EXSWSTRHPC-300

stanleytsang-amd · 2026-01-19T23:12:15Z

Can confirm that the error: Illegal instruction detected: Operand has incorrect register class. build errors in the codecov job are not specific to the changes in this PR - they appear even in develop branch.

We'll need to file a big with the compiler team about this, but for now we can waive this failure for approval purposes.

paulklie · 2026-01-20T14:41:36Z

Can confirm that the error: Illegal instruction detected: Operand has incorrect register class. build errors in the codecov job are not specific to the changes in this PR - they appear even in develop branch.

I built everything with ninja all, but was unable to reproduce this on the develop branch (with and without coverage), while building for gfx950 with rocm-7.1.1. Though since we do not have a gfx950 gpu I am unable to run tests.

However we were able to reproduce the error with rocm 7.2-rel. Probably an compiler regression? What compiler version are you using?

obersteiner · 2026-01-26T17:29:06Z

Can confirm that the error: Illegal instruction detected: Operand has incorrect register class. build errors in the codecov job are not specific to the changes in this PR - they appear even in develop branch.

We'll need to file a big with the compiler team about this, but for now we can waive this failure for approval purposes.

This sounds very familiar. We have ticket for this.

stanleytsang-amd · 2026-01-30T20:54:37Z

@yungshengtu The latest commits have resulted in a build failure:

[ 95%] Building HIP object test/public_headers_host_only/CMakeFiles/headertest_std_cuda_bit.dir/headers/cuda_bit.cpp.o
In file included from /var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/build/test/public_headers/headers/cuda_std_numbers.cu:27:
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:318:30: error: constexpr variable 'e_v<__half>' must be initialized by a constant expression
318 | _CCCL_GLOBAL_CONSTANT __half e_v<__half> = __numbers<__half>::__e();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:136:12: note: in call to '__half(__half_raw{.x = 16752U})'
136 | return __half_raw{.x = 0x4170u};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:318:44: note: in call to '__e()'
318 | _CCCL_GLOBAL_CONSTANT __half e_v<__half> = __numbers<__half>::__e();
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:320:30: error: constexpr variable 'log2e_v<__half>' must be initialized by a constant expression
320 | _CCCL_GLOBAL_CONSTANT __half log2e_v<__half> = __numbers<__half>::__log2e();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:143:12: note: in call to '__half(__half_raw{.x = 15813U})'
143 | return __half_raw{.x = 0x3dc5u};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:320:48: note: in call to '__log2e()'
320 | _CCCL_GLOBAL_CONSTANT __half log2e_v<__half> = __numbers<__half>::__log2e();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:322:30: error: constexpr variable 'log10e_v<__half>' must be initialized by a constant expression
322 | _CCCL_GLOBAL_CONSTANT __half log10e_v<__half> = __numbers<__half>::__log10e();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:150:12: note: in call to '__half(__half_raw{.x = 14067U})'
150 | return __half_raw{.x = 0x36f3u};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:322:49: note: in call to '__log10e()'
322 | _CCCL_GLOBAL_CONSTANT __half log10e_v<__half> = __numbers<__half>::__log10e();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:324:30: error: constexpr variable 'pi_v<__half>' must be initialized by a constant expression
324 | _CCCL_GLOBAL_CONSTANT __half pi_v<__half> = __numbers<__half>::__pi();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:157:12: note: in call to '__half(__half_raw{.x = 16968U})'
157 | return __half_raw{.x = 0x4248u};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:324:45: note: in call to '__pi()'
324 | _CCCL_GLOBAL_CONSTANT __half pi_v<__half> = __numbers<__half>::__pi();
| ^~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:326:30: error: constexpr variable 'inv_pi_v<__half>' must be initialized by a constant expression
326 | _CCCL_GLOBAL_CONSTANT __half inv_pi_v<__half> = __numbers<__half>::__inv_pi();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:164:12: note: in call to '__half(__half_raw{.x = 13592U})'
164 | return __half_raw{.x = 0x3518u};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:326:49: note: in call to '__inv_pi()'
326 | _CCCL_GLOBAL_CONSTANT __half inv_pi_v<__half> = __numbers<__half>::__inv_pi();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:328:30: error: constexpr variable 'inv_sqrtpi_v<__half>' must be initialized by a constant expression
328 | _CCCL_GLOBAL_CONSTANT __half inv_sqrtpi_v<__half> = __numbers<__half>::__inv_sqrtpi();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:171:12: note: in call to '__half(__half_raw{.x = 14467U})'
171 | return __half_raw{.x = 0x3883u};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:328:53: note: in call to '__inv_sqrtpi()'
328 | _CCCL_GLOBAL_CONSTANT __half inv_sqrtpi_v<__half> = __numbers<__half>::__inv_sqrtpi();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:330:30: error: constexpr variable 'ln2_v<__half>' must be initialized by a constant expression
330 | _CCCL_GLOBAL_CONSTANT __half ln2_v<__half> = __numbers<__half>::__ln2();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:178:12: note: in call to '__half(__half_raw{.x = 14732U})'
178 | return __half_raw{.x = 0x398cu};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:330:46: note: in call to '__ln2()'
330 | _CCCL_GLOBAL_CONSTANT __half ln2_v<__half> = __numbers<__half>::__ln2();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:332:30: error: constexpr variable 'ln10_v<__half>' must be initialized by a constant expression
332 | _CCCL_GLOBAL_CONSTANT __half ln10_v<__half> = __numbers<__half>::__ln10();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:185:12: note: in call to '__half(__half_raw{.x = 16539U})'
185 | return __half_raw{.x = 0x409bu};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:332:47: note: in call to '__ln10()'
332 | _CCCL_GLOBAL_CONSTANT __half ln10_v<__half> = __numbers<__half>::__ln10();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:334:30: error: constexpr variable 'sqrt2_v<__half>' must be initialized by a constant expression
334 | _CCCL_GLOBAL_CONSTANT __half sqrt2_v<__half> = __numbers<__half>::__sqrt2();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:192:12: note: in call to '__half(__half_raw{.x = 15784U})'
192 | return __half_raw{.x = 0x3da8u};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:334:48: note: in call to '__sqrt2()'
334 | _CCCL_GLOBAL_CONSTANT __half sqrt2_v<__half> = __numbers<__half>::__sqrt2();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:336:30: error: constexpr variable 'sqrt3_v<__half>' must be initialized by a constant expression
336 | _CCCL_GLOBAL_CONSTANT __half sqrt3_v<__half> = __numbers<__half>::__sqrt3();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:199:12: note: in call to '__half(__half_raw{.x = 16110U})'
199 | return __half_raw{.x = 0x3eeeu};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:336:48: note: in call to '__sqrt3()'
336 | _CCCL_GLOBAL_CONSTANT __half sqrt3_v<__half> = __numbers<__half>::__sqrt3();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:338:30: error: constexpr variable 'inv_sqrt3_v<__half>' must be initialized by a constant expression
338 | _CCCL_GLOBAL_CONSTANT __half inv_sqrt3_v<__half> = __numbers<__half>::__inv_sqrt3();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:206:12: note: in call to '__half(__half_raw{.x = 14494U})'
206 | return __half_raw{.x = 0x389eu};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:338:52: note: in call to '__inv_sqrt3()'
338 | _CCCL_GLOBAL_CONSTANT __half inv_sqrt3_v<__half> = __numbers<__half>::__inv_sqrt3();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:340:30: error: constexpr variable 'egamma_v<__half>' must be initialized by a constant expression
340 | _CCCL_GLOBAL_CONSTANT __half egamma_v<__half> = __numbers<__half>::__egamma();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:213:12: note: in call to '__half(__half_raw{.x = 14494U})'
213 | return __half_raw{.x = 0x389eu};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:340:49: note: in call to '__egamma()'
340 | _CCCL_GLOBAL_CONSTANT __half egamma_v<__half> = __numbers<__half>::__egamma();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:342:30: error: constexpr variable 'phi_v<__half>' must be initialized by a constant expression
342 | _CCCL_GLOBAL_CONSTANT __half phi_v<__half> = __numbers<__half>::__phi();
| ^ ~~~~~~~~~~~~~~~~~~~~~~~~~~
/opt/rocm-7.2.0/lib/llvm/bin/../../../include/hip/amd_detail/amd_hip_fp16.h:102:64: note: read of member 'data' of union with active member 'x' is not allowed in a constant expression
102 | HOST_DEVICE constexpr __half(const __half_raw& x) : data{x.data} {}
| ^
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:220:12: note: in call to '__half(__half_raw{.x = 15993U})'
220 | return __half_raw{.x = 0x3e79u};
| ^~~~~~~~~~~~~~~~~~~~~~~~
/var/jenkins_home/workspace/ies_precheckin_rocthrust_PR-3773/oBRIoS14f/rocthrust/projects/rocthrust/libhipcxx/include/cuda/std/numbers:342:46: note: in call to '__phi()'
342 | _CCCL_GLOBAL_CONSTANT __half phi_v<__half> = __numbers<__half>::__phi();
| ^~~~~~~~~~~~~~~~~~~~~~~~~~
[ 95%] Linking HIP shared library libheadertest_cuda_std_initializer_list.so
[ 95%] Built target headertest_cuda_std_initializer_list
[ 95%] Linking HIP shared library libheadertest_cuda_std_chrono.so
[ 95%] Linking HIP shared library libheadertest_cuda_std_cmath.so
[ 95%] Building HIP object test/public_headers_host_only/CMakeFiles/headertest_std_cuda_cmath.dir/headers/cuda_cmath.cpp.o
13 errors generated when compiling for gfx90a.
make[2]: *** [test/public_headers/CMakeFiles/headertest_cuda_std_numbers.dir/build.make:75: test/public_headers/CMakeFiles/headertest_cuda_std_numbers.dir/headers/cuda_std_numbers.cu.o] Error 1
make[1]: *** [CMakeFiles/Makefile2:20406: test/public_headers/CMakeFiles/headertest_cuda_std_numbers.dir/all] Error 2
make[1]: *** Waiting for unfinished jobs....

Naraenda · 2026-02-02T11:04:16Z

@stanleytsang-amd This error might be related to ROCm/rocm-systems#2240. Does the image that Jenkins CI is building on also include this change to CLR?

stanleytsang-amd · 2026-02-06T21:57:03Z

@stanleytsang-amd This error might be related to ROCm/rocm-systems#2240. Does the image that Jenkins CI is building on also include this change to CLR?

Sorry for the delay - it does not seem to be part of the ROCm build used in math-ci (7.2 RC2). I will try RC3 or 7.2.1

codecov-commenter · 2026-02-18T13:04:37Z

Codecov Report

❌ Patch coverage is 97.89157% with 14 lines in your changes missing coverage. Please review.

Files with missing lines	Patch %	Lines
projects/rocthrust/test/test_scan.cpp	0.00%	10 Missing ⚠️
projects/rocthrust/test/test_device_delete.cpp	0.00%	2 Missing ⚠️
projects/rocthrust/test/test_counting_iterator.cpp	91.67%	1 Missing ⚠️
projects/rocthrust/test/test_memory.cpp	97.83%	0 Missing and 1 partial ⚠️

Additional details and impacted files

@@             Coverage Diff              @@
##           develop    #3773       +/-   ##
============================================
+ Coverage    69.59%   91.60%   +22.01%     
============================================
  Files         2771      469     -2302     
  Lines       452699    40018   -412681     
  Branches     66658     5085    -61573     
============================================
- Hits        315022    36655   -278367     
+ Misses      117337     3034   -114303     
+ Partials     20340      329    -20011

Flag	Coverage Δ
TensileLite	`?`
TensileLite-CPP	`?`
TensileLite-Unit	`?`
hipBLAS	`?`
hipBLASLt	`?`
hipCUB	`?`
hipDNN	`?`
hipFFT	`?`
hipRAND	`?`
hipSOLVER	`?`
hipSPARSE	`?`
rocBLAS	`?`
rocFFT	`?`
rocRAND	`?`
rocSOLVER	`?`
rocSPARSE	`?`
rocThrust	`91.60% <97.89%> (+0.24%)`	⬆️

Files with missing lines	Coverage Δ
projects/rocthrust/test/test_address_stability.cpp	`94.59% <100.00%> (ø)`
projects/rocthrust/test/test_allocator.cpp	`94.64% <100.00%> (ø)`
...s/rocthrust/test/test_allocator_aware_policies.cpp	`100.00% <ø> (ø)`
projects/rocthrust/test/test_binary_search.cpp	`99.41% <100.00%> (ø)`
projects/rocthrust/test/test_complex.cpp	`98.45% <100.00%> (ø)`
projects/rocthrust/test/test_constant_iterator.cpp	`100.00% <100.00%> (ø)`
projects/rocthrust/test/test_copy.cpp	`96.85% <ø> (ø)`
projects/rocthrust/test/test_dereference.cpp	`100.00% <ø> (ø)`
projects/rocthrust/test/test_discard_iterator.cpp	`100.00% <100.00%> (ø)`
projects/rocthrust/test/test_find.cpp	`98.02% <100.00%> (-0.01%)`	⬇️
... and 74 more

... and 2524 files with indirect coverage changes

🚀 New features to boost your workflow:

❄️ Test Analytics: Detect flaky tests, report on failures, and find test suite problems.
📦 JS Bundle Analysis: Save yourself from yourself by tracking and limiting bundle sizes in JS merges.

stanleytsang-amd · 2026-02-20T16:50:45Z

@yungshengtu When installing this rocThrust either via make install or via package, the __cccl_config file appears to be omitted from the install.

yungshengtu · 2026-02-20T19:55:33Z

@yungshengtu When installing this rocThrust either via make install or via package, the __cccl_config file appears to be omitted from the install.

@stanleytsang-amd Sorry about that. I missed it.
I’ve realised that when installing rocThrust, it only picks up files matching the patterns *.h, *.cuh, *.hpp, and *.inl. Some files under libcxx_wrapper didn’t have an extension, so they weren’t being installed. I’ve now renamed them all to .h (modified in this commit 8416434), so they’re picked up properly.

…anges

Temporarily skip these tests, which are running out of memory. We can re-enable them once we have a robust solution.

feat(rocthrust): CCCL 3.0.x changes MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit Most of the changes here are pulling in the commits from [CCCL 3.0.3](https://github.com/NVIDIA/cccl/tree/v3.0.3). With the fixes and workarounds attached to these commits. In addition, this PR also includes: - upgrading the versions of dependencies - moving all libcxx wrapper codes into thrust/detail/libcxx_wrapper - C++20 onwards support - fixing the errors from thrust/unique_ptr.h - fixing the errors from test/generate_resource_spec.cpp JIRA ID : EXSWSTRHPC-300

(cherry picked from commit 607b365a6a408a6a72fd1a0753f33796e475a50e)

Most of the changes here are pulling in the commits from [CCCL 3.0.3](https://github.com/NVIDIA/cccl/tree/v3.0.3). With the fixes and workarounds attached to these commits. In addition, this PR also includes: - upgrading the versions of dependencies - moving all libcxx wrapper codes into thrust/detail/libcxx_wrapper - C++20 onwards support - fixing the errors from thrust/unique_ptr.h - fixing the errors from test/generate_resource_spec.cpp JIRA ID : EXSWSTRHPC-300 --------- Co-authored-by: Nara Prasetya <nara@streamhpc.com> Co-authored-by: Mátyás Aradi <matyas@streamhpc.com> Co-authored-by: Wayne Franz <wayfranz@amd.com>

yungshengtu self-assigned this Jan 12, 2026

yungshengtu requested review from a team as code owners January 12, 2026 15:36

yungshengtu added project: rocthrust organization: streamhpc contributors from streamhpc labels Jan 12, 2026

github-actions Bot added the documentation label Jan 12, 2026

umfranzw requested changes Jan 14, 2026

View reviewed changes

Comment thread projects/rocthrust/CHANGELOG.md Resolved

yungshengtu force-pushed the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch 4 times, most recently from 5133c96 to 44d7902 Compare January 15, 2026 08:20

stanleytsang-amd reviewed Jan 28, 2026

View reviewed changes

Comment thread projects/rocthrust/extra/CMakeLists.txt Resolved

Comment thread projects/rocthrust/cmake/FindROCMCmake.cmake Resolved

yungshengtu force-pushed the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch from 44d7902 to 998118a Compare January 29, 2026 22:02

yungshengtu force-pushed the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch 2 times, most recently from 11629e4 to c1d1094 Compare February 18, 2026 12:10

yungshengtu force-pushed the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch 2 times, most recently from 63afda6 to 2f148ff Compare February 20, 2026 09:13

yungshengtu force-pushed the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch from 2f148ff to e26f228 Compare February 20, 2026 19:46

yungshengtu force-pushed the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch from e26f228 to b9de8c7 Compare March 3, 2026 06:36

yungshengtu force-pushed the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch 2 times, most recently from 0b7d7a2 to d95e575 Compare April 10, 2026 07:26

yungshengtu and others added 15 commits July 24, 2026 16:12

Bump branch/3.0.x to 3.0.1.

cee9fa4

Bump branch/3.0.x to 3.0.2.

3fd9883

Add a macro to disable PDL

9b35db7

Bump branch/3.0.x to 3.0.3.

651187b

v3.0.3

f5993c8

fix compatability with libhipcxx

21c174e

allow C++20, C++23 and C++26

1fd6786

update CHANGELOG.md

0f0e683

update versions of dependencies

6f90aef

work with gnu++17

ff7d3f2

fix(rocthrust): various libhipcxx fixes

39c9816

update copyright date

fd91d4a

fix benchmark error caused by #5953

98242ac

reflect the idea of #6460

f33fcce

[rocThrust] Fix libcxx version check

b5e0a8c

yungshengtu force-pushed the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch from 8429739 to b5e0a8c Compare July 24, 2026 16:18

JeniferC99 mentioned this pull request Jul 24, 2026

[Issue] Windows gfx1151 CI: Driver / GPU sanity check hangs on hipInfo.exe (rocprim shard 1) ROCm/TheRock#6857

Open

Merge branch 'develop' into users/yungshengtu/rocthrust-cccl-3-0-x-ch…

27db7c0

…anges

github-actions Bot added the project: rocprim label Jul 26, 2026

Temporarily skip misbehaving tests on gfx1151/Windows

2eb1f88

Temporarily skip these tests, which are running out of memory. We can re-enable them once we have a robust solution.

umfranzw force-pushed the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch from 05926d1 to 2eb1f88 Compare July 26, 2026 02:14

umfranzw merged commit 1eb022d into develop Jul 26, 2026
68 of 70 checks passed

umfranzw deleted the users/yungshengtu/rocthrust-cccl-3-0-x-changes branch July 26, 2026 12:17

TriveniTadapaneni mentioned this pull request Jul 27, 2026

Revert "feat(rocthrust): CCCL 3.0.x changes" #10001

Closed

Naraenda mentioned this pull request Sep 14, 2026

DO NOT MERGE: feat(hipcub, rocprim, rocthrust): cccl 3.x parity #10568

Closed

1 task done

stanleytsang-amd pushed a commit that referenced this pull request Sep 18, 2026

refactor cuda cmath docs (#3773)

f94e09f

(cherry picked from commit 607b365a6a408a6a72fd1a0753f33796e475a50e)

Naraenda mentioned this pull request Oct 2, 2026

feat(hipccl): cccl 3.x parity #12952

Merged

1 task done

Navigation Menu

Repository navigation

feat(rocthrust): CCCL 3.0.x changes - #3773

feat(rocthrust): CCCL 3.0.x changes#3773
umfranzw merged 154 commits into
developfrom
users/yungshengtu/rocthrust-cccl-3-0-x-changes

yungshengtu commented Jan 12, 2026 •

edited

Loading

Uh oh!

Uh oh!

stanleytsang-amd commented Jan 19, 2026

Uh oh!

paulklie commented Jan 20, 2026 •

edited

Loading

Uh oh!

obersteiner commented Jan 26, 2026

Uh oh!

Uh oh!

Uh oh!

stanleytsang-amd commented Jan 30, 2026

Uh oh!

Naraenda commented Feb 2, 2026

Uh oh!

stanleytsang-amd commented Feb 6, 2026

Uh oh!

codecov-commenter commented Feb 18, 2026 •

edited by codecov Bot

Loading

Uh oh!

stanleytsang-amd commented Feb 20, 2026

Uh oh!

yungshengtu commented Feb 20, 2026 •

edited

Loading

Uh oh!

Uh oh!

Reviewers

Uh oh!

Assignees

Labels

Projects

Milestone

Development

Uh oh!

8 participants

Conversation

yungshengtu commented Jan 12, 2026 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

Uh oh!

stanleytsang-amd commented Jan 19, 2026

Uh oh!

paulklie commented Jan 20, 2026 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

obersteiner commented Jan 26, 2026

Uh oh!

Uh oh!

Uh oh!

stanleytsang-amd commented Jan 30, 2026

Uh oh!

Naraenda commented Feb 2, 2026

Uh oh!

stanleytsang-amd commented Feb 6, 2026

Uh oh!

codecov-commenter commented Feb 18, 2026 • edited by codecov Bot Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Codecov Report

Uh oh!

stanleytsang-amd commented Feb 20, 2026

Uh oh!

yungshengtu commented Feb 20, 2026 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

Uh oh!

Reviewers

Uh oh!

Assignees

Labels

Projects

Milestone

Development

Uh oh!

8 participants

yungshengtu commented Jan 12, 2026 •

edited

Loading

paulklie commented Jan 20, 2026 •

edited

Loading

codecov-commenter commented Feb 18, 2026 •

edited by codecov Bot

Loading

yungshengtu commented Feb 20, 2026 •

edited

Loading