Skip to content
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

Make traccc::alpaka executables work in pure C++ #832

Merged
merged 13 commits into from
Feb 6, 2025
1 change: 1 addition & 0 deletions device/alpaka/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -27,6 +27,7 @@ traccc_add_alpaka_library( traccc_alpaka alpaka TYPE SHARED
# Utility definitions.
"include/traccc/alpaka/utils/make_prefix_sum_buff.hpp"
"src/utils/make_prefix_sum_buff.cpp"
"src/utils/get_device_info.cpp"
# Seed finding code.
"include/traccc/alpaka/seeding/spacepoint_binning.hpp"
"include/traccc/alpaka/seeding/seed_finding.hpp"
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -8,6 +8,8 @@
#pragma once

// Project include(s)
#include <vecmem/utils/copy.hpp>

#include "traccc/edm/seed.hpp"
#include "traccc/edm/spacepoint.hpp"
#include "traccc/edm/track_parameters.hpp"
Expand Down Expand Up @@ -58,7 +60,7 @@ struct track_params_estimation
/// Memory resource used by the algorithm
traccc::memory_resource m_mr;
/// Copy object used by the algorithm
vecmem::copy& m_copy;
::vecmem::copy& m_copy;
};

} // namespace traccc::alpaka
31 changes: 31 additions & 0 deletions device/alpaka/include/traccc/alpaka/utils/device_tag.hpp
Original file line number Diff line number Diff line change
@@ -0,0 +1,31 @@
/** TRACCC library, part of the ACTS project (R&D line)
*
* (c) 2024 CERN for the benefit of the ACTS project
*
* Mozilla Public License Version 2.0
*/

#pragma once

#include <alpaka/acc/Tag.hpp>

namespace traccc::alpaka {

// Get alpaka tag for current device
#if defined(ALPAKA_ACC_GPU_CUDA_ENABLED)
using AccTag = ::alpaka::TagGpuCudaRt;
#elif defined(ALPAKA_ACC_GPU_HIP_ENABLED)
using AccTag = ::alpaka::TagGpuHipRt;
#elif defined(ALPAKA_ACC_SYCL_ENABLED)
#if defined(ALPAKA_SYCL_ONEAPI_CPU)
using AccTag = ::alpaka::TagCpuSycl;
#elif defined(ALPAKA_SYCL_ONEAPI_FPGA)
using AccTag = ::alpaka::TagFpgaSyclIntel;
#elif defined(ALPAKA_SYCL_ONEAPI_GPU)
using AccTag = ::alpaka::TagGpuSyclIntel;
#endif
#elif defined(ALPAKA_ACC_CPU_B_SEQ_T_THREADS_ENABLED)
using AccTag = ::alpaka::TagCpuThreads;
#endif

} // namespace traccc::alpaka
19 changes: 19 additions & 0 deletions device/alpaka/include/traccc/alpaka/utils/get_device_info.hpp
Original file line number Diff line number Diff line change
@@ -0,0 +1,19 @@
/** TRACCC library, part of the ACTS project (R&D line)
*
* (c) 2024 CERN for the benefit of the ACTS project
*
* Mozilla Public License Version 2.0
*/

#pragma once

#include <string>

namespace traccc::alpaka {

/// Function that prints the current device information to the console.
/// Included as part of the traccc::alpaka namespace, to avoid having to include
/// alpaka headers in any users of the library.
std::string get_device_info();

} // namespace traccc::alpaka
13 changes: 11 additions & 2 deletions device/alpaka/include/traccc/alpaka/utils/vecmem_types.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -32,7 +32,7 @@
#include <vecmem/utils/copy.hpp>
#endif

#include <alpaka/alpaka.hpp>
#include "traccc/alpaka/utils/device_tag.hpp"

// Forward declarations so we can compile the types below
namespace vecmem {
Expand Down Expand Up @@ -69,7 +69,7 @@ struct host_device_types {
};
template <>
struct host_device_types<::alpaka::TagGpuCudaRt> {
using device_memory_resource = ::vecmem::cuda::host_memory_resource;
using device_memory_resource = ::vecmem::cuda::device_memory_resource;
using host_memory_resource = ::vecmem::cuda::host_memory_resource;
using managed_memory_resource = ::vecmem::cuda::managed_memory_resource;
using device_copy = ::vecmem::cuda::copy;
Expand Down Expand Up @@ -102,4 +102,13 @@ struct host_device_types<::alpaka::TagGpuSyclIntel> {
using managed_memory_resource = ::vecmem::sycl::shared_memory_resource;
using device_copy = ::vecmem::sycl::copy;
};

using device_memory_resource =
typename host_device_types<AccTag>::device_memory_resource;
using host_memory_resource =
typename host_device_types<AccTag>::host_memory_resource;
using managed_memory_resource =
typename host_device_types<AccTag>::managed_memory_resource;
using device_copy = typename host_device_types<AccTag>::device_copy;

} // namespace traccc::alpaka::vecmem
4 changes: 3 additions & 1 deletion device/alpaka/src/seeding/seed_finding.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -343,7 +343,9 @@ seed_finding::output_type seed_finding::operator()(

// Calculate the number of threads and thread blocks to run the weight
// updating kernel for.
threadsPerBlock = warpSize * 2 < maxThreads ? warpSize * 2 : maxThreads;
threadsPerBlock = getWarpSize<Acc>() * 2 < maxThreads
? getWarpSize<Acc>() * 2
: maxThreads;
blocksPerGrid =
(pBufHost_counter->m_nTriplets + threadsPerBlock - 1) / threadsPerBlock;
workDiv = makeWorkDiv<Acc>(blocksPerGrid, threadsPerBlock);
Expand Down
2 changes: 1 addition & 1 deletion device/alpaka/src/seeding/track_params_estimation.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -63,7 +63,7 @@ track_params_estimation::output_type track_params_estimation::operator()(
// Run the kernel
::alpaka::exec<Acc>(queue, workDiv, EstimateTrackParamsKernel{},
spacepoints_view, seeds_view, bfield, stddev,
vecmem::get_data(params_buffer));
::vecmem::get_data(params_buffer));
::alpaka::wait(queue);

return params_buffer;
Expand Down
23 changes: 23 additions & 0 deletions device/alpaka/src/utils/get_device_info.cpp
Original file line number Diff line number Diff line change
@@ -0,0 +1,23 @@
/** TRACCC library, part of the ACTS project (R&D line)
*
* (c) 2025 CERN for the benefit of the ACTS project
*
* Mozilla Public License Version 2.0
*/

// Local include(s).
#include "utils.hpp"

// Project include(s).
#include "traccc/alpaka/utils/get_device_info.hpp"

namespace traccc::alpaka {

std::string get_device_info() {
int device = 0;
auto devAcc = ::alpaka::getDevByIdx(::alpaka::Platform<Acc>{}, 0u);
return std::string("Using Alpaka device: " + ::alpaka::getName(devAcc) +
" [id: " + std::to_string(device) + "] ");
}

} // namespace traccc::alpaka
49 changes: 31 additions & 18 deletions device/alpaka/src/utils/utils.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -8,34 +8,47 @@
#pragma once

#include <alpaka/alpaka.hpp>
#include <alpaka/example/ExampleDefaultAcc.hpp>

#ifdef ALPAKA_ACC_GPU_CUDA_ENABLED
#include <vecmem/utils/cuda/copy.hpp>
#endif

#ifdef ALPAKA_ACC_GPU_HIP_ENABLED
#include <vecmem/utils/hip/copy.hpp>
#endif

#include <vecmem/utils/copy.hpp>

namespace traccc::alpaka {

using Dim = ::alpaka::DimInt<1>;
using Idx = uint32_t;
using WorkDiv = ::alpaka::WorkDivMembers<Dim, Idx>;

using Acc = ::alpaka::ExampleDefaultAcc<Dim, Idx>;
// Get alpaka accelerator - based on alpaka/examples/ExampleDefaultAcc.hpp
#if defined(ALPAKA_ACC_GPU_CUDA_ENABLED)
using Acc = ::alpaka::AccGpuCudaRt<Dim, Idx>;
#elif defined(ALPAKA_ACC_GPU_HIP_ENABLED)
using Acc = ::alpaka::AccGpuHipRt<Dim, Idx>;
#elif defined(ALPAKA_ACC_SYCL_ENABLED)
#if defined(ALPAKA_SYCL_ONEAPI_CPU)
using Acc = ::alpaka::AccCpuSycl<Dim, Idx>;
#elif defined(ALPAKA_SYCL_ONEAPI_FPGA)
using Acc = ::alpaka::AccFpgaSyclIntel<Dim, Idx>;
#elif defined(ALPAKA_SYCL_ONEAPI_GPU)
using Acc = ::alpaka::AccGpuSyclIntel<Dim, Idx>;
#endif
#elif defined(ALPAKA_ACC_CPU_B_SEQ_T_THREADS_ENABLED)
using Acc = ::alpaka::AccCpuThreads<Dim, Idx>;
#else
#error "No supported backend selected." //we definitely want to fail the build if no matching accelerator is found
#endif

using Host = ::alpaka::DevCpu;
using Queue = ::alpaka::Queue<Acc, ::alpaka::Blocking>;

static constexpr std::size_t warpSize =
#if defined(ALPAKA_ACC_GPU_CUDA_ENABLED) || defined(ALPAKA_ACC_GPU_HIP_ENABLED)
32;
#else
4;
#endif
template <typename TAcc>
consteval std::size_t getWarpSize() {
if constexpr (::alpaka::accMatchesTags<TAcc, ::alpaka::TagGpuCudaRt,
::alpaka::TagGpuSyclIntel>) {
return 32;
}
if constexpr (::alpaka::accMatchesTags<TAcc, ::alpaka::TagGpuHipRt>) {
return 64;
} else {
return 4;
}
}

template <typename TAcc>
inline WorkDiv makeWorkDiv(Idx blocks, Idx threadsOrElements) {
Expand Down
12 changes: 0 additions & 12 deletions examples/run/alpaka/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -15,16 +15,12 @@ include(traccc-alpaka-functions)
traccc_enable_language_alpaka()

if(alpaka_ACC_GPU_CUDA_ENABLE)
set_source_files_properties(${TRACCC_ALPAKA_EXAMPLE_SOURCES} PROPERTIES LANGUAGE CUDA)

list (APPEND EXTRA_LIBS vecmem::cuda)
elseif(alpaka_ACC_GPU_HIP_ENABLE)
find_package( HIPToolkit REQUIRED )
set_source_files_properties(${TRACCC_ALPAKA_EXAMPLE_SOURCES} PROPERTIES LANGUAGE HIP)
list(APPEND EXTRA_LIBS HIP::hiprt vecmem::hip)
elseif(alpaka_ACC_SYCL_ENABLE)
list(APPEND EXTRA_LIBS vecmem::sycl)
set_source_files_properties(${TRACCC_ALPAKA_EXAMPLE_SOURCES} PROPERTIES LANGUAGE SYCL)
endif()

set(LIBRARIES vecmem::core traccc::io traccc::performance
Expand All @@ -35,11 +31,3 @@ traccc_add_executable( seq_example_alpaka "seq_example_alpaka.cpp"
LINK_LIBRARIES ${LIBRARIES} )
traccc_add_executable( seeding_example_alpaka "seeding_example_alpaka.cpp"
LINK_LIBRARIES ${LIBRARIES} )

#Can only do this once target is defined, so need another if here
if(alpaka_ACC_GPU_HIP_ENABLE)
set_target_properties( traccc_seq_example_alpaka PROPERTIES
POSITION_INDEPENDENT_CODE TRUE )
set_target_properties( traccc_seeding_example_alpaka PROPERTIES
POSITION_INDEPENDENT_CODE TRUE )
endif()
31 changes: 8 additions & 23 deletions examples/run/alpaka/seeding_example_alpaka.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -35,7 +35,6 @@
#include "traccc/seeding/track_params_estimation.hpp"

// Detray include(s).
#include "alpaka/example/ExampleDefaultAcc.hpp"
#include "detray/core/detector.hpp"
#include "detray/detectors/bfield.hpp"
#include "detray/io/frontend/detector_reader.hpp"
Expand Down Expand Up @@ -63,33 +62,19 @@ int seq_run(const traccc::opts::track_seeding& seeding_opts,
const traccc::opts::performance& performance_opts,
const traccc::opts::accelerator& accelerator_opts) {

using Dim = ::alpaka::DimInt<1>;
using Idx = uint32_t;

using Acc = ::alpaka::ExampleDefaultAcc<Dim, Idx>;
#ifdef ALPAKA_ACC_SYCL_ENABLED
::sycl::queue q;
vecmem::sycl::queue_wrapper qw{&q};
traccc::alpaka::vecmem::host_device_types<
::alpaka::trait::AccToTag<Acc>::type>::device_copy copy(qw);
traccc::alpaka::vecmem::host_device_types<
::alpaka::trait::AccToTag<Acc>::type>::host_memory_resource host_mr(qw);
traccc::alpaka::vecmem::host_device_types<
::alpaka::trait::AccToTag<Acc>::type>::device_memory_resource
device_mr(qw);
traccc::alpaka::vecmem::host_device_types<
::alpaka::trait::AccToTag<Acc>::type>::managed_memory_resource
mng_mr(qw);
traccc::alpaka::vecmem::device_copy copy(qw);
traccc::alpaka::vecmem::host_memory_resource host_mr(qw);
traccc::alpaka::vecmem::device_memory_resource device_mr(qw);
traccc::alpaka::vecmem::managed_memory_resource mng_mr(qw);
traccc::memory_resource mr{device_mr, &host_mr};
#else
traccc::alpaka::vecmem::host_device_types<
::alpaka::trait::AccToTag<Acc>::type>::device_copy copy;
traccc::alpaka::vecmem::host_device_types<
::alpaka::trait::AccToTag<Acc>::type>::host_memory_resource host_mr;
traccc::alpaka::vecmem::host_device_types<
::alpaka::trait::AccToTag<Acc>::type>::device_memory_resource device_mr;
traccc::alpaka::vecmem::host_device_types<
::alpaka::trait::AccToTag<Acc>::type>::managed_memory_resource mng_mr;
traccc::alpaka::vecmem::device_copy copy;
traccc::alpaka::vecmem::host_memory_resource host_mr;
traccc::alpaka::vecmem::device_memory_resource device_mr;
traccc::alpaka::vecmem::managed_memory_resource mng_mr;
traccc::memory_resource mr{device_mr, &host_mr};
#endif

Expand Down
24 changes: 6 additions & 18 deletions examples/run/alpaka/seq_example_alpaka.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -6,7 +6,6 @@
*/

// Project include(s).
#include "alpaka/example/ExampleDefaultAcc.hpp"
#include "traccc/alpaka/clusterization/clusterization_algorithm.hpp"
#include "traccc/alpaka/clusterization/measurement_sorting_algorithm.hpp"
#include "traccc/alpaka/seeding/seeding_algorithm.hpp"
Expand Down Expand Up @@ -66,28 +65,17 @@ int seq_run(const traccc::opts::detector& detector_opts,
const traccc::vector3 field_vec = {0.f, 0.f,
seeding_opts.seedfinder.bFieldInZ};

using Dim = ::alpaka::DimInt<1>;
using Idx = uint32_t;

using Acc = ::alpaka::ExampleDefaultAcc<Dim, Idx>;
// Memory resources used by the application.
#ifdef ALPAKA_ACC_SYCL_ENABLED
::sycl::queue q;
vecmem::sycl::queue_wrapper qw{&q};
traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::device_copy copy(qw);
traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::host_memory_resource host_mr(qw);
traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::device_memory_resource
device_mr(qw);
traccc::alpaka::vecmem::device_copy copy(qw);
traccc::alpaka::vecmem::host_memory_resource host_mr(qw);
traccc::alpaka::vecmem::device_memory_resource device_mr(qw);
#else
traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::device_copy copy;
traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::host_memory_resource host_mr;
traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::device_memory_resource device_mr;
traccc::alpaka::vecmem::device_copy copy;
traccc::alpaka::vecmem::host_memory_resource host_mr;
traccc::alpaka::vecmem::device_memory_resource device_mr;
#endif
traccc::memory_resource mr{device_mr, &host_mr};

Expand Down
15 changes: 5 additions & 10 deletions tests/alpaka/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -4,28 +4,23 @@
#
# Mozilla Public License Version 2.0

set(TRACCC_ALPAKA_TEST_SOURCES
alpaka_basic.cpp
test_cca.cpp
)


include(traccc-alpaka-functions)
traccc_enable_language_alpaka()

if(alpaka_ACC_GPU_CUDA_ENABLE)
set_source_files_properties(${TRACCC_ALPAKA_TEST_SOURCES} PROPERTIES LANGUAGE CUDA)
set_source_files_properties(alpaka_basic.cpp PROPERTIES LANGUAGE CUDA)
list(APPEND DEVICE_LIBRARIES vecmem::cuda)
elseif(alpaka_ACC_GPU_HIP_ENABLE)
set_source_files_properties(${TRACCC_ALPAKA_TEST_SOURCES} PROPERTIES LANGUAGE HIP)
set_source_files_properties(alpaka_basic.cpp PROPERTIES LANGUAGE HIP)
list(APPEND DEVICE_LIBRARIES vecmem::hip)
elseif(alpaka_ACC_SYCL_ENABLE)
list(APPEND DEVICE_LIBRARIES vecmem::sycl)
set_source_files_properties(${TRACCC_ALPAKA_TEST_SOURCES} PROPERTIES LANGUAGE SYCL)
set_source_files_properties(alpaka_basic.cpp PROPERTIES LANGUAGE SYCL)
endif()

traccc_add_test( alpaka
${TRACCC_ALPAKA_TEST_SOURCES}
alpaka_basic.cpp
test_cca.cpp
LINK_LIBRARIES
GTest::gtest_main
traccc_tests_common
Expand Down
12 changes: 4 additions & 8 deletions tests/alpaka/alpaka_basic.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -139,17 +139,13 @@ GTEST_TEST(AlpakaBasic, VecMemOp) {
#ifdef ALPAKA_ACC_SYCL_ENABLED
::sycl::queue q;
vecmem::sycl::queue_wrapper qw{&q};
traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::device_copy vm_copy(qw);
traccc::alpaka::vecmem::device_copy vm_copy(qw);
#else
traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::device_copy vm_copy;
traccc::alpaka::vecmem::device_copy vm_copy;
#endif

traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::host_memory_resource host_mr;
traccc::alpaka::vecmem::host_device_types<
alpaka::trait::AccToTag<Acc>::type>::device_memory_resource device_mr;
traccc::alpaka::vecmem::host_memory_resource host_mr;
traccc::alpaka::vecmem::device_memory_resource device_mr;

vecmem::vector<float> host_vector{n, &host_mr};

Expand Down
Loading
Loading