-
Notifications
You must be signed in to change notification settings - Fork 190
New issue
Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.
By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.
Already on GitHub? Sign in to your account
Add thrust::universal_host_pinned_vector
#2653
Add thrust::universal_host_pinned_vector
#2653
Conversation
// TODO: This should be upstreamed and then removed. | ||
namespace thrust | ||
{ | ||
|
||
using universal_raw_memory_resource = | ||
thrust::system::cuda::detail::cuda_memory_resource<thrust::system::cuda::detail::cudaMallocManaged, cudaFree, void*>; | ||
|
||
template <typename T> | ||
using universal_allocator = thrust::mr::stateless_resource_allocator<T, universal_raw_memory_resource>; | ||
|
||
template <typename T> | ||
using universal_vector = thrust::device_vector<T, universal_allocator<T>>; | ||
|
||
} // namespace thrust | ||
|
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
This is a drive-by fix, since this functionality has been upstreamed into Thrust IIUC.
VTEST<thrust::device_vector< \ | ||
int, \ | ||
thrust::mr::stateless_resource_allocator<int, thrust::universal_host_pinned_memory_resource>>>(); \ |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Here, our unit tests already cover cost pinned memory.
1ce4a2e
to
e9c32c6
Compare
using universal_memory_resource = detail::universal_native_resource; | ||
// FIXME(bgruber): comment below is wrong or alias should be to universal_memory_resource | ||
/*! An alias for \p cpp::universal_memory_resource. */ | ||
using universal_host_pinned_memory_resource = detail::native_resource; |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
We have these aliases in each system. IMO, universal_host_pinned_memory_resource
should be an alias to universal_memory_resource
except for the CUDA system. This caused a test to fail because the pointer type of universal_host_pinned_memory_resource
is a native pointer and a not a universial_ptr<T>
. I commented out that test for now. Fixing this is technically a breaking change.
If we want to be conservative, we should postpone fixing those aliases to the next major release.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Tracking here: #2659
universal_host_pinned_memory_resource
's pointer type should probably be a universal_ptr
#2659
🟨 CI finished in 1h 39m: Pass: 92%/394 | Total: 4d 14h | Avg: 16m 51s | Max: 1h 08m | Hits: 66%/25781
|
Project | |
---|---|
CCCL Infrastructure | |
+/- | libcu++ |
CUB | |
+/- | Thrust |
CUDA Experimental | |
python | |
CCCL C Parallel Library | |
Catch2Helper |
Modifications in project or dependencies?
Project | |
---|---|
CCCL Infrastructure | |
+/- | libcu++ |
+/- | CUB |
+/- | Thrust |
+/- | CUDA Experimental |
+/- | python |
+/- | CCCL C Parallel Library |
+/- | Catch2Helper |
🏃 Runner counts (total jobs: 394)
# | Runner |
---|---|
326 | linux-amd64-cpu16 |
28 | linux-arm64-cpu16 |
25 | linux-amd64-gpu-v100-latest-1 |
15 | windows-amd64-cpu16 |
e9c32c6
to
02905b3
Compare
🟩 CI finished in 1h 31m: Pass: 100%/394 | Total: 3d 12h | Avg: 12m 50s | Max: 1h 12m | Hits: 60%/25793
|
Project | |
---|---|
CCCL Infrastructure | |
+/- | libcu++ |
CUB | |
+/- | Thrust |
CUDA Experimental | |
python | |
CCCL C Parallel Library | |
Catch2Helper |
Modifications in project or dependencies?
Project | |
---|---|
CCCL Infrastructure | |
+/- | libcu++ |
+/- | CUB |
+/- | Thrust |
+/- | CUDA Experimental |
+/- | python |
+/- | CCCL C Parallel Library |
+/- | Catch2Helper |
🏃 Runner counts (total jobs: 394)
# | Runner |
---|---|
326 | linux-amd64-cpu16 |
28 | linux-arm64-cpu16 |
25 | linux-amd64-gpu-v100-latest-1 |
15 | windows-amd64-cpu16 |
Error snippets: /root/cccl/libcudacxx/examples/concurrent_hash_table.cu(28): error: namespace "thrust" has no member "universal_raw_memory_resource" typename MemoryResource = thrust::universal_raw_memory_resource> ^ and: /root/cccl/libcudacxx/examples/concurrent_hash_table.cu(188): error: namespace "thrust" has no member "universal_allocator" auto freq = thrust::allocate_unique<table>(thrust::universal_allocator<table>{}, 8); ^ This issue is introduced in PR NVIDIA#2653 and commit d893269.
We have all the pieces in Thrust to provide a vector backed by pinned host memory. Only a few aliases were missing. Furthermore, the unit tests were already covering a vector with pinned memory.
This PR adds the missing aliases and extends a few tests. The main new alias is
thrust::universal_host_pinned_vector<T>
which uses CUDA pinned host memory (cudaMallocHost
), and the native memory resource on other systems (as dictated by the existinguniversal_host_pinned_memory_resource
on each system).Fixes: #2485