Skip to content

Commit 6ead071

Browse files
hkaiserhkaiser
authored
Merge pull request #6744 from adityacodes30/master
Integrating thrust into HPX
2 parents 4dacbac + cb1b869 commit 6ead071

20 files changed

Lines changed: 2088 additions & 1 deletion

File tree

‎CMakeLists.txt‎

Lines changed: 15 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -606,6 +606,21 @@ if(HPX_WITH_CUDA AND HPX_WITH_HIP)
606606
)
607607
endif()
608608

609+
# HPX THRUST configuration
610+
set(HPX_WITH_THRUST_DEFAULT OFF)
611+
if(HPX_WITH_CUDA)
612+
set(HPX_WITH_THRUST_DEFAULT ON)
613+
endif()
614+
615+
hpx_option(
616+
HPX_WITH_THRUST
617+
BOOL
618+
"Enable support for NVIDIA Thrust integration (default: ON when CUDA is enabled, OFF otherwise)"
619+
${HPX_WITH_THRUST_DEFAULT}
620+
ADVANCED
621+
CATEGORY "Build Targets"
622+
)
623+
609624
# ## HPX STDEXEC configuration ##
610625

611626
hpx_option(

‎libs/core/CMakeLists.txt‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -79,6 +79,7 @@ set(_hpx_core_modules
7979
threading
8080
threading_base
8181
threadmanager
82+
thrust
8283
timed_execution
8384
timing
8485
topology

‎libs/core/execution_base/include/hpx/execution_base/any_sender.hpp‎

Lines changed: 3 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -34,7 +34,9 @@ namespace hpx::detail {
3434
template <typename T>
3535
using empty_vtable_t = typename empty_vtable_type<T>::type;
3636

37-
#if !defined(HPX_MSVC) && defined(HPX_HAVE_CXX20_TRIVIAL_VIRTUAL_DESTRUCTOR)
37+
#if !defined(HPX_MSVC) && \
38+
defined(HPX_HAVE_CXX20_TRIVIAL_VIRTUAL_DESTRUCTOR) && \
39+
!defined(__CUDACC__) && !defined(HPX_COMPUTE_DEVICE_CODE)
3840
template <typename T>
3941
inline constexpr empty_vtable_t<T> empty_vtable{};
4042

‎libs/core/thrust/CMakeLists.txt‎

Lines changed: 39 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,39 @@
1+
# Copyright (c) 2025 The STE||AR-Group
2+
#
3+
# SPDX-License-Identifier: BSL-1.0
4+
# Distributed under the Boost Software License, Version 1.0. (See accompanying
5+
# file LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt)
6+
7+
if(NOT (HPX_WITH_CUDA AND HPX_WITH_THRUST))
8+
return()
9+
endif()
10+
11+
list(APPEND CMAKE_MODULE_PATH "${CMAKE_CURRENT_SOURCE_DIR}/cmake")
12+
13+
set(thrust_headers hpx/thrust/policy.hpp hpx/thrust/algorithms.hpp
14+
hpx/thrust/detail/algorithm_map.hpp
15+
)
16+
17+
# cmake-format: off
18+
set(thrust_compat_headers)
19+
# cmake-format: on
20+
21+
set(thrust_sources)
22+
23+
include(HPX_AddModule)
24+
add_hpx_module(
25+
core thrust
26+
GLOBAL_HEADER_GEN ON
27+
SOURCES ${thrust_sources}
28+
HEADERS ${thrust_headers}
29+
COMPAT_HEADERS ${thrust_compat_headers}
30+
MODULE_DEPENDENCIES
31+
hpx_algorithms
32+
hpx_concepts
33+
hpx_execution
34+
hpx_execution_base
35+
hpx_functional
36+
hpx_futures
37+
hpx_async_cuda
38+
CMAKE_SUBDIRS examples tests
39+
)

‎libs/core/thrust/docs/index.rst‎

Lines changed: 90 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,90 @@
1+
..
2+
Copyright (c) 2025 The STE||AR-Group
3+
4+
SPDX-License-Identifier: BSL-1.0
5+
Distributed under the Boost Software License, Version 1.0. (See accompanying
6+
file LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt)
7+
8+
.. _modules_thrust:
9+
10+
======
11+
thrust
12+
======
13+
14+
The thrust module integrates |hpx| parallel algorithms with NVIDIA Thrust,
15+
enabling GPU acceleration using familiar |hpx| algorithm syntax.
16+
17+
Execution Policies
18+
==================
19+
20+
``hpx::thrust::thrust_host_policy``
21+
CPU execution using optimized parallel implementations
22+
23+
``hpx::thrust::thrust_device_policy``
24+
Synchronous GPU execution
25+
26+
``hpx::thrust::thrust_task_policy``
27+
Asynchronous GPU execution returning |hpx| futures
28+
29+
Policy Mappings
30+
===============
31+
32+
The |hpx| thrust policies map directly to Thrust execution policies:
33+
34+
.. code-block:: c++
35+
36+
// Policy mappings
37+
hpx::thrust::thrust_host_policy -> thrust::host
38+
hpx::thrust::thrust_device_policy -> thrust::device
39+
hpx::thrust::thrust_task_policy -> thrust::par_nosync
40+
41+
Usage
42+
=====
43+
44+
Synchronous Device Execution
45+
-----------------------------
46+
47+
.. code-block:: c++
48+
49+
#include <hpx/thrust/policy.hpp>
50+
#include <hpx/thrust/algorithms.hpp>
51+
#include <thrust/device_vector.h>
52+
53+
hpx::thrust::thrust_device_policy device{};
54+
thrust::device_vector<int> d_vec(1000, 0);
55+
56+
hpx::fill(device, d_vec.begin(), d_vec.end(), 42);
57+
int sum = hpx::reduce(device, d_vec.begin(), d_vec.end(), 0);
58+
59+
Asynchronous Execution with Futures
60+
------------------------------------
61+
62+
.. code-block:: c++
63+
64+
#include <hpx/thrust/policy.hpp>
65+
#include <hpx/async_cuda/cuda_polling_helper.hpp>
66+
67+
int hpx_main() {
68+
// Required for async operations
69+
hpx::cuda::experimental::enable_user_polling polling_guard("default");
70+
71+
hpx::thrust::thrust_task_policy task{};
72+
thrust::device_vector<int> d_vec(1000, 1);
73+
74+
auto fill_future = hpx::fill(task, d_vec.begin(), d_vec.end(), 42);
75+
fill_future.get(); // Standard HPX future operations
76+
77+
}
78+
79+
Build Requirements
80+
==================
81+
82+
* CUDA Toolkit 12.4.0+
83+
* |hpx| with ``HPX_WITH_CUDA=ON``
84+
* Enable with: ``cmake -DHPX_WITH_CUDA=ON ...``
85+
86+
See Also
87+
========
88+
89+
* :ref:`modules_async_cuda` - CUDA integration
90+
* :ref:`modules_execution` - |hpx| execution policies

‎libs/core/thrust/examples/CMakeLists.txt‎

Lines changed: 36 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,36 @@
1+
# Copyright (c) 2025 Aditya Sapra
2+
#
3+
# SPDX-License-Identifier: BSL-1.0
4+
# Distributed under the Boost Software License, Version 1.0. (See accompanying
5+
# file LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt)
6+
7+
if(HPX_WITH_EXAMPLES)
8+
add_hpx_pseudo_target(examples.modules.thrust)
9+
add_hpx_pseudo_dependencies(examples.modules examples.modules.thrust)
10+
if(HPX_WITH_CUDA AND HPX_WITH_THRUST)
11+
set(thrust_examples device_fill host_fill device_copy device_transform
12+
async_fill_n reverse_reduce
13+
)
14+
15+
foreach(example ${thrust_examples})
16+
set(source ${example}.cu)
17+
source_group("Source Files" FILES ${source})
18+
19+
add_hpx_executable(
20+
${example} INTERNAL_FLAGS
21+
SOURCES ${source}
22+
EXCLUDE_FROM_ALL
23+
HPX_PREFIX ${HPX_BUILD_PREFIX}
24+
FOLDER "Examples/Modules/Core/Thrust"
25+
)
26+
27+
add_hpx_pseudo_dependencies(examples.modules.thrust ${example})
28+
endforeach()
29+
endif()
30+
if(HPX_WITH_TESTS AND HPX_WITH_TESTS_EXAMPLES)
31+
add_hpx_pseudo_target(tests.examples.modules.thrust)
32+
add_hpx_pseudo_dependencies(
33+
tests.examples.modules tests.examples.modules.thrust
34+
)
35+
endif()
36+
endif()

‎libs/core/thrust/examples/async_fill_n.cu‎

Lines changed: 35 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,35 @@
1+
// Copyright (c) 2025 Aditya Sapra
2+
//
3+
// SPDX-License-Identifier: BSL-1.0
4+
// Distributed under the Boost Software License, Version 1.0. (See accompanying
5+
// file LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt)
6+
//
7+
8+
#include <hpx/algorithm.hpp>
9+
#include <hpx/execution.hpp>
10+
#include <hpx/init.hpp>
11+
12+
#include <hpx/thrust/algorithms.hpp>
13+
#include <hpx/thrust/policy.hpp>
14+
15+
#include <hpx/async_cuda/cuda_polling_helper.hpp>
16+
#include <hpx/modules/async_cuda.hpp>
17+
#include <thrust/device_vector.h>
18+
19+
int hpx_main(int, char**)
20+
{
21+
// Enable CUDA event polling on the "default" pool for HPX <-> CUDA future bridging
22+
hpx::cuda::experimental::enable_user_polling polling_guard("default");
23+
24+
hpx::thrust::thrust_task_policy task{};
25+
thrust::device_vector<int> v(1024, 0);
26+
auto f = hpx::fill_n(task, v.begin(), 512, 9);
27+
f.get();
28+
return hpx::local::finalize();
29+
}
30+
31+
int main(int argc, char** argv)
32+
{
33+
hpx::local::init_params init_args;
34+
return hpx::local::init(hpx_main, argc, argv, init_args);
35+
}

‎libs/core/thrust/examples/device_copy.cu‎

Lines changed: 30 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,30 @@
1+
// Copyright (c) 2025 Aditya Sapra
2+
//
3+
// SPDX-License-Identifier: BSL-1.0
4+
// Distributed under the Boost Software License, Version 1.0. (See accompanying
5+
// file LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt)
6+
//
7+
8+
#include <hpx/algorithm.hpp>
9+
#include <hpx/execution.hpp>
10+
#include <hpx/init.hpp>
11+
12+
#include <hpx/thrust/algorithms.hpp>
13+
#include <hpx/thrust/policy.hpp>
14+
15+
#include <thrust/device_vector.h>
16+
17+
int hpx_main(int, char**)
18+
{
19+
hpx::thrust::thrust_device_policy device{};
20+
thrust::device_vector<int> src(1024, 5);
21+
thrust::device_vector<int> dst(1024, 0);
22+
hpx::copy(device, src.begin(), src.end(), dst.begin());
23+
return hpx::local::finalize();
24+
}
25+
26+
int main(int argc, char** argv)
27+
{
28+
hpx::local::init_params init_args;
29+
return hpx::local::init(hpx_main, argc, argv, init_args);
30+
}

‎libs/core/thrust/examples/device_fill.cu‎

Lines changed: 29 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,29 @@
1+
// Copyright (c) 2025 Aditya Sapra
2+
//
3+
// SPDX-License-Identifier: BSL-1.0
4+
// Distributed under the Boost Software License, Version 1.0. (See accompanying
5+
// file LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt)
6+
//
7+
8+
#include <hpx/algorithm.hpp>
9+
#include <hpx/execution.hpp>
10+
#include <hpx/init.hpp>
11+
12+
#include <hpx/thrust/algorithms.hpp>
13+
#include <hpx/thrust/policy.hpp>
14+
15+
#include <thrust/device_vector.h>
16+
17+
int hpx_main(int, char**)
18+
{
19+
hpx::thrust::thrust_device_policy device{};
20+
thrust::device_vector<int> v(1024, 0);
21+
hpx::fill(device, v.begin(), v.end(), 1);
22+
return hpx::local::finalize();
23+
}
24+
25+
int main(int argc, char** argv)
26+
{
27+
hpx::local::init_params init_args;
28+
return hpx::local::init(hpx_main, argc, argv, init_args);
29+
}

‎libs/core/thrust/examples/device_transform.cu‎

Lines changed: 33 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,33 @@
1+
// Copyright (c) 2025 Aditya Sapra
2+
//
3+
// SPDX-License-Identifier: BSL-1.0
4+
// Distributed under the Boost Software License, Version 1.0. (See accompanying
5+
// file LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt)
6+
//
7+
8+
#include <hpx/algorithm.hpp>
9+
#include <hpx/execution.hpp>
10+
#include <hpx/init.hpp>
11+
12+
#include <hpx/thrust/algorithms.hpp>
13+
#include <hpx/thrust/policy.hpp>
14+
15+
#include <thrust/device_vector.h>
16+
17+
int hpx_main(int, char**)
18+
{
19+
hpx::thrust::thrust_device_policy device{};
20+
thrust::device_vector<int> a(1024, 2);
21+
thrust::device_vector<int> b(1024, 3);
22+
thrust::device_vector<int> out(1024, 0);
23+
24+
hpx::transform(device, a.begin(), a.end(), b.begin(), out.begin(),
25+
[] __device__(int x, int y) { return x + y; });
26+
return hpx::local::finalize();
27+
}
28+
29+
int main(int argc, char** argv)
30+
{
31+
hpx::local::init_params init_args;
32+
return hpx::local::init(hpx_main, argc, argv, init_args);
33+
}

0 commit comments

Comments
 (0)