diff --git a/.github/workflows/traccc.yml b/.github/workflows/traccc.yml new file mode 100644 index 00000000000..38b749fef1a --- /dev/null +++ b/.github/workflows/traccc.yml @@ -0,0 +1,235 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2021-2026 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +name: Traccc + +on: + push: + paths: + - "Traccc/**" + - "Detray/**" + pull_request: + branches: + - main + paths: + - "Traccc/**" + - "Detray/**" + +concurrency: + group: ${{ github.head_ref || github.run_id }} + cancel-in-progress: true + +jobs: + containers: + name: ${{ matrix.platform.name }}-${{ matrix.build }} + runs-on: ubuntu-latest + container: ${{ matrix.platform.container }} + strategy: + matrix: + platform: + - name: CPU + container: ghcr.io/acts-project/ubuntu2404:69 + options: --preset host-fp32 + run_tests: true + enable_cuda: false + enable_sycl: false + - name: CPU + container: ghcr.io/acts-project/ubuntu2404:69 + options: --preset host-fp64 + run_tests: false + enable_cuda: false + enable_sycl: false + - name: CUDA + container: ghcr.io/acts-project/ubuntu2404_cuda:69 + options: --preset cuda-fp32 -DTRACCC_BUILD_TESTING=FALSE + run_tests: false + enable_cuda: true + enable_sycl: false + - name: "SYCL Intel" + container: ghcr.io/acts-project/ubuntu2404_oneapi:69 + options: --preset sycl-fp32 + run_tests: true + enable_cuda: false + enable_sycl: true + build: + - Release + - Debug + include: + - platform: + name: CUDA + container: ghcr.io/acts-project/ubuntu2404_cuda:69 + options: --preset cuda-fp64 + run_tests: false + enable_cuda: true + enable_hip: false + enable_sycl: false + build: Release + - platform: + name: "SYCL NVIDIA" + container: ghcr.io/acts-project/ubuntu2404_cuda_oneapi:69 + options: --preset sycl-fp32 + run_tests: false + enable_cuda: false + enable_hip: false + enable_sycl: true + build: Release + - platform: + name: "SYCL AMD" + container: ghcr.io/acts-project/ubuntu2404_rocm_oneapi:69 + options: --preset sycl-fp32 + run_tests: false + enable_cuda: false + enable_hip: true + enable_sycl: true + build: Release + - platform: + name: "ALPAKA_CPU" + container: ghcr.io/acts-project/ubuntu2404:69 + options: --preset alpaka-fp32 + run_tests: true + enable_cuda: false + enable_hip: false + enable_sycl: false + build: Release + - platform: + name: "ALPAKA_CUDA" + container: ghcr.io/acts-project/ubuntu2404_cuda:69 + options: --preset alpaka-fp32 -Dalpaka_ACC_GPU_CUDA_ENABLE=ON -DTRACCC_FAIL_ON_WARNINGS=OFF -DTRACCC_BUILD_CUDA_UTILS=ON + run_tests: false + enable_cuda: true + enable_hip: false + enable_sycl: false + build: Release + - platform: + name: "ALPAKA_HIP_SYCL" + container: ghcr.io/acts-project/ubuntu2404_rocm_oneapi:69 + options: --preset alpaka-fp32 -Dalpaka_ACC_GPU_HIP_ENABLE=ON -Dalpaka_DISABLE_VENDOR_RNG=ON -DTRACCC_USE_ROOT=OFF -DTRACCC_SETUP_ROCTHRUST=ON -DTRACCC_BUILD_HIP=ON -DCMAKE_HIP_COMPILE_FEATURES=hip_std_20 + run_tests: false + enable_cuda: false + enable_hip: true + enable_sycl: true + build: Release + - platform: + name: "ALPAKA_SYCL" + container: ghcr.io/acts-project/ubuntu2404_oneapi:69 + options: --preset alpaka-fp32 -Dalpaka_ACC_CPU_B_SEQ_T_THREADS_ENABLE=OFF -Dalpaka_ACC_SYCL_ENABLE=ON -Dalpaka_SYCL_ONEAPI_GPU=ON -Dalpaka_SYCL_ONEAPI_GPU_DEVICES=spir64 -DTRACCC_BUILD_SYCL_UTILS=ON + run_tests: false + enable_cuda: false + enable_hip: false + enable_sycl: true + build: Release + # Use BASH as the shell from the images. + defaults: + run: + shell: bash + steps: + - uses: actions/checkout@v4 + - name: Install dependencies + run: | + apt install -y zstd wget + curl --retry 5 --retry-delay 10 --output deps.tar.zst https://acts.web.cern.ch/ACTS/ci/ubuntu-24.04/deps.v6.tar.zst + tar -xf deps.tar.zst -C /usr/local --strip-components=1 + rm deps.tar.zst + - name: Build nlohmann + run: | + wget https://github.com/nlohmann/json/archive/refs/tags/v3.11.3.tar.gz + tar -xzvf v3.11.3.tar.gz + export CXX=g++ + cmake \ + -S json-3.11.3 \ + -B /json_build/ \ + -DJSON_BuildTests=OFF \ + -DCMAKE_INSTALL_PREFIX=/json_install/ \ + -DCMAKE_BUILD_TYPE=Release + cmake --build /json_build/ -- -j $(nproc) + cmake --install /json_build/ + - name: Build covfie + run: | + wget https://github.com/acts-project/covfie/archive/refs/tags/v0.15.2.tar.gz + tar -xzvf v0.15.2.tar.gz + export CXX=g++ + cmake \ + -S covfie-0.15.2 \ + -B /covfie_build/ \ + -DCMAKE_INSTALL_PREFIX=/covfie_install/ + cmake --build /covfie_build/ -- -j $(nproc) + cmake --install /covfie_build/ + - name: Build vecmem + run: | + source ${GITHUB_WORKSPACE}/Traccc/.github/ci_setup.sh ${{ matrix.platform.name }} + wget https://github.com/acts-project/vecmem/archive/refs/tags/v1.25.0.tar.gz + tar -xzvf v1.25.0.tar.gz + export CXX=${SYCLCXX:-g++} + cmake \ + -S vecmem-1.25.0/ \ + -B /vecmem_build/ \ + -DVECMEM_BUILD_CUDA_LIBRARY=${{ matrix.platform.enable_cuda && 'ON' || 'OFF' }} \ + -DVECMEM_BUILD_SYCL_LIBRARY=${{ matrix.platform.enable_sycl && 'ON' || 'OFF' }} \ + -DVECMEM_BUILD_HIP_LIBRARY=${{ matrix.platform.enable_hip && 'ON' || 'OFF' }} \ + -DCMAKE_INSTALL_PREFIX=/vecmem_install/ + cmake --build /vecmem_build/ -- -j $(nproc) + cmake --install /vecmem_build/ + # vecmem ships FindHIPToolkit.cmake but does not install it, so + # downstream find_package(HIPToolkit) (e.g. traccc's alpaka HIP + # backend) cannot locate it. Copy it into the installed cmake dir, + # which vecmem-config.cmake adds to CMAKE_MODULE_PATH. + cp vecmem-1.25.0/cmake/FindHIPToolkit.cmake /vecmem_install/lib/cmake/vecmem-*/ + - name: Build ACTS + run: | + source ${GITHUB_WORKSPACE}/Traccc/.github/ci_setup.sh ${{ matrix.platform.name }} + export CMAKE_PREFIX_PATH=/json_install/:/vecmem_install/:/covfie_install/:${CMAKE_PREFIX_PATH} + export CXX=g++ + cmake \ + -S ${GITHUB_WORKSPACE} \ + -B /acts_build/ \ + -DACTS_BUILD_PLUGIN_JSON=ON \ + -DACTS_USE_SYSTEM_NLOHMANN_JSON=ON \ + -DACTS_SETUP_DETRAY=ON \ + -DACTS_USE_SYSTEM_COVFIE=ON \ + -DACTS_USE_SYSTEM_VECMEM=ON \ + -DDETRAY_USE_SYSTEM_GOOGLETEST=ON \ + -DDETRAY_BUILD_TEST_UTILS=ON \ + -DCMAKE_INSTALL_PREFIX=/acts_install/ \ + -DCMAKE_BUILD_TYPE=Release \ + -DDETRAY_GENERATE_METADATA=itk_metadata \ + -DDETRAY_SET_LOGGING=NONE \ + -DACTS_ENABLE_CUDA=${{ matrix.platform.enable_cuda && 'ON' || 'OFF' }} + cmake --build /acts_build/ -- -j $(nproc) + cmake --install /acts_build/ + - name: Configure + run: | + source ${GITHUB_WORKSPACE}/Traccc/.github/ci_setup.sh ${{ matrix.platform.name }} + export CMAKE_PREFIX_PATH=/json_install/:/vecmem_install/:/covfie_install/:/acts_install/:${CMAKE_PREFIX_PATH} + cmake \ + -DCMAKE_BUILD_TYPE=${{ matrix.build }} \ + ${{ matrix.platform.options }} \ + -DTRACCC_USE_SYSTEM_ACTS=ON \ + -DTRACCC_USE_SYSTEM_VECMEM=ON \ + -DDETRAY_USE_SYSTEM_NLOHMANN=ON \ + -DTRACCC_USE_SYSTEM_COVFIE=ON \ + -DTRACCC_USE_SYSTEM_GOOGLETEST_DEFAULT=ON \ + -DTRACCC_SETUP_GOOGLETEST=ON \ + -DTRACCC_SUPPORTED_DETECTORS="default_detector;odd_detector;telescope_detector;toy_detector;itk_detector" \ + -DDETRAY_GENERATE_METADATA="itk_metadata" \ + -S ${GITHUB_WORKSPACE}/Traccc \ + -B build + - name: Build + run: | + source ${GITHUB_WORKSPACE}/Traccc/.github/ci_setup.sh ${{ matrix.platform.name }} + cmake --build build -- -j $(nproc) + - name: Download data files + if: "matrix.platform.run_tests" + run: ${GITHUB_WORKSPACE}/Traccc/data/traccc_data_get_files.sh + - name: Test + if: "matrix.platform.run_tests && matrix.build == 'Release'" + run: | + cd build + source ${GITHUB_WORKSPACE}/Traccc/.github/ci_setup.sh ${{ matrix.platform.name }} + ctest --output-on-failure + - name: FP64 Compliance + if: "matrix.platform.name == 'CUDA' && matrix.build == 'Debug'" + continue-on-error: true + run: ${GITHUB_WORKSPACE}/Traccc/.github/find_f64_ptx.py --source ${GITHUB_WORKSPACE}/Traccc --build build $(find build -name "*.ptx") diff --git a/.gitignore b/.gitignore index bfecf0a66a7..5d408f8c3c4 100644 --- a/.gitignore +++ b/.gitignore @@ -40,3 +40,6 @@ codegen/src/codegen.egg-info/ !.clang-tidy !.oclint !.devcontainer + +# don't ignore CMake preset files +!CMakePresets.json diff --git a/.gitlab-ci.yml b/.gitlab-ci.yml index ed69e1f71a2..1f19ce27785 100644 --- a/.gitlab-ci.yml +++ b/.gitlab-ci.yml @@ -592,3 +592,98 @@ detray_test_sycl: - sycl-ls - ctest --output-on-failure -R ".*sycl.*" - find bin -type f -name "*sycl" -not -name "*text*" -exec {} \; + +# traccc GitLab CI workflows +.traccc_base_template: &traccc_base_job + before_script: + - apt install -y zstd + - curl --retry 5 --retry-delay 10 --output deps.tar.zst https://acts.web.cern.ch/ACTS/ci/ubuntu-24.04/deps.v6.tar.zst + - tar -xf deps.tar.zst -C /usr/local --strip-components=1 + - rm deps.tar.zst + - git clone $CLONE_URL src + - git -C src checkout $HEAD_SHA + - source ./src/Traccc/.github/ci_setup.sh ${TRACCC_BUILD_TYPE} + - export CTEST_PARALLEL_LEVEL=2 + +.traccc_build_template: &traccc_build_job + <<: *traccc_base_job + tags: [docker] + stage: build + artifacts: + paths: + - build + script: + - SYCLFLAGS="${TRACCC_SYCL_FLAGS}" CXXFLAGS="${TRACCC_CXX_FLAGS}" + cmake -G Ninja --preset ${TRACCC_BUILD_PRESET} -DFETCHCONTENT_SOURCE_DIR_ACTS=src/ -DTRACCC_SETUP_VECMEM=ON -DTRACCC_SETUP_COVFIE=ON -DACTS_SETUP_VECMEM=OFF -DACTS_SETUP_COVFIE=OFF -DDETRAY_SET_LOGGING=NONE -DCMAKE_BUILD_TYPE=Release -DTRACCC_BUILD_EXAMPLES=FALSE ${TRACCC_CMAKE_ARGS} + -S src/Traccc -B build + - cmake --build build --parallel 2 + +.traccc_nvidia_test_template: &traccc_nvidia_test_job + <<: *traccc_base_job + tags: [docker-gpu-nvidia] + stage: test + script: + - nvidia-smi + - ./src/Traccc/data/traccc_data_get_files.sh + - ctest --output-on-failure --test-dir build/ -R "${TRACCC_BUILD_TYPE}" + +.traccc_intel_test_template: &traccc_intel_test_job + <<: *traccc_base_job + tags: [docker-gpu-nvidia] + stage: test + script: + - ./src/Traccc/data/traccc_data_get_files.sh + - ctest --output-on-failure --test-dir build/ -R "${TRACCC_BUILD_TYPE}" + +traccc_build_cuda: + <<: *traccc_build_job + image: ghcr.io/acts-project/ubuntu2404_cuda:69 + variables: + TRACCC_BUILD_TYPE: CUDA + TRACCC_BUILD_PRESET: cuda-fp32 + +traccc_test_cuda: + <<: *traccc_nvidia_test_job + image: ghcr.io/acts-project/ubuntu2404_cuda:69 + variables: + TRACCC_BUILD_TYPE: CUDA + TRACCC_CMAKE_ARGS: -DACTS_ENABLE_CUDA=ON -DVECMEM_BUILD_CUDA_LIBRARY=ON + dependencies: + - traccc_build_cuda + +traccc_build_sycl_intel: + <<: *traccc_build_job + image: ghcr.io/acts-project/ubuntu2404_oneapi:69 + variables: + TRACCC_BUILD_TYPE: SYCL + TRACCC_BUILD_PRESET: sycl-fp32 + TRACCC_SYCL_FLAGS: -fsycl -fsycl-targets=spir64 + TRACCC_CMAKE_ARGS: -DTRACCC_BUILD_CUDA=FALSE -DACTS_ENABLE_CUDA=OFF -DVECMEM_BUILD_SYCL_LIBRARY=ON + +traccc_test_sycl_intel: + <<: *traccc_intel_test_job + image: ghcr.io/acts-project/ubuntu2404_oneapi:69 + variables: + TRACCC_BUILD_TYPE: SYCL + ONEAPI_DEVICE_SELECTOR: opencl:* + dependencies: + - traccc_build_sycl_intel + +traccc_build_sycl_nvidia: + <<: *traccc_build_job + image: ghcr.io/acts-project/ubuntu2404_cuda_oneapi:69 + variables: + TRACCC_BUILD_TYPE: SYCL + TRACCC_BUILD_PRESET: sycl-fp32 + TRACCC_SYCL_FLAGS: -fsycl -fsycl-targets=nvidia_gpu_sm_75 -Wno-unknown-cuda-version -Wno-deprecated-declarations + TRACCC_CXX_FLAGS: -Wno-deprecated-declarations + TRACCC_CMAKE_ARGS: -DTRACCC_BUILD_CUDA=FALSE -DACTS_ENABLE_CUDA=OFF -DVECMEM_BUILD_SYCL_LIBRARY=ON + +traccc_test_sycl_nvidia: + <<: *traccc_nvidia_test_job + image: ghcr.io/acts-project/ubuntu2404_cuda_oneapi:69 + variables: + TRACCC_BUILD_TYPE: SYCL + ONEAPI_DEVICE_SELECTOR: cuda:* + dependencies: + - traccc_build_sycl_nvidia diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index bc85abea158..2c7948d2721 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -5,6 +5,7 @@ repos: - id: clang-format types_or: [file] files: \.(cpp|hpp|ipp|cu|cuh)$ + exclude: ^Traccc/ - repo: https://github.com/pre-commit/pre-commit-hooks rev: v6.0.0 @@ -23,12 +24,14 @@ repos: rev: 26.5.1 hooks: - id: black-jupyter + exclude: ^Traccc/ - repo: https://github.com/BlankSpruce/gersemi-pre-commit rev: 0.27.6 hooks: - id: gersemi args: ["-i", "--no-warn-about-unknown-commands"] + exclude: ^Traccc/ - repo: https://github.com/codespell-project/codespell rev: v2.4.2 @@ -39,7 +42,7 @@ repos: "-I", "./CI/codespell_ignore.txt", "-w" ] - exclude: ^CI/.*|^docs/old|docs/tex-mml-chtml.js$ + exclude: ^CI/.*|^docs/old|docs/tex-mml-chtml.js$|^Traccc/ - repo: local hooks: @@ -48,26 +51,28 @@ repos: language: system entry: CI/check_license.py --fix files: \.(cpp|hpp|ipp|cu|cuh)$ + exclude: ^Traccc/ - id: include_guards name: include_guards language: system entry: CI/check_include_guards.py --fail-global files: \.hpp$ - exclude: Detray/core/include/detray/utils/logging.hpp + exclude: ^Detray/core/include/detray/utils/logging.hpp$|^Traccc/ - id: pragma_once name: pragma_once language: system entry: CI/check_pragma_once.sh files: \.(hpp|ipp)$ - exclude: Detray/core/include/detray/utils/quiet_log_start.hpp|Detray/core/include/detray/utils/quiet_log_end.hpp + exclude: "Detray/core/include/detray/utils/quiet_log_start.hpp|Detray/core/include/detray/utils/quiet_log_end.hpp|^Traccc/" - id: type_t name: type_t language: system entry: CI/check_type_t.py --fix files: \.(cpp|hpp|ipp|cu|cuh)$ + exclude: ^Traccc/ - id: boost_test name: boost_test @@ -92,6 +97,7 @@ repos: language: system entry: CI/check_math_macros.py files: \.(cpp|hpp|ipp|cu|cuh)$ + exclude: ^Traccc/ - id: codegen_dependencies name: codegen_dependencies @@ -107,7 +113,7 @@ repos: types: [json] require_serial: true args: [--indent=4] - exclude: ^.zenodo.json$ + exclude: ^.zenodo.json$|^Traccc/ - id: sync_citation_metadata name: sync citation metadata diff --git a/CI/check_unused_files.py b/CI/check_unused_files.py index 8c138630a3c..e4cbcafb0df 100755 --- a/CI/check_unused_files.py +++ b/CI/check_unused_files.py @@ -16,6 +16,8 @@ "Detray/detectors", # CLI tools "Detray/tests/tools", + # TODO: Remove the traccc part. + "Traccc", "git", "Python", "Scripts", diff --git a/CI/clang_tidy/filter.yml b/CI/clang_tidy/filter.yml index 78675c1a35b..7d79834ba32 100644 --- a/CI/clang_tidy/filter.yml +++ b/CI/clang_tidy/filter.yml @@ -23,5 +23,9 @@ exclude_path_regexes: - ^/opt/ - ^/cvmfs/ + # traccc source files + # TODO: Remove this at a future date. + - Traccc/ + # Severity for GitHub annotations ("error" or "warning") severity: error diff --git a/CMakeLists.txt b/CMakeLists.txt index 377f5a3227a..65373e0c3e4 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -222,6 +222,7 @@ set_option_if(ACTS_BUILD_PLUGIN_GNN ACTS_BUILD_EXAMPLES_GNN) set_option_if(ACTS_BUILD_PLUGIN_FPEMON ACTS_BUILD_EXAMPLES) set_option_if(ACTS_BUILD_PLUGIN_JSON ACTS_BUILD_PLUGIN_TRACCC) set_option_if(ACTS_BUILD_PLUGIN_ACTSVG ACTS_BUILD_PLUGIN_TRACCC) +set_option_if(ACTS_BUILD_PLUGIN_ACTSVG ACTS_SETUP_DETRAY) set_option_if( ACTS_ENABLE_CUDA ACTS_GNN_ENABLE_MODULEMAP @@ -586,8 +587,10 @@ endif() if(ACTS_SETUP_VECMEM) if(ACTS_USE_SYSTEM_VECMEM) find_package(vecmem ${_acts_vecmem_version} REQUIRED) - else() + elseif(NOT TARGET vecmem::core) add_subdirectory(thirdparty/vecmem) + list(PREPEND CMAKE_MODULE_PATH "${VECMEM_LANGUAGE_DIR}") + list(PREPEND CMAKE_MODULE_PATH "${VECMEM_LANGUAGE_DIR}/sycl") # Make the "VecMem language code" available for the whole project. include("${VECMEM_LANGUAGE_DIR}/vecmem-check-language.cmake") endif() @@ -628,6 +631,19 @@ if(ACTS_BUILD_TRACCC) option(TRACCC_USE_ROOT "" FALSE) option(TRACCC_SETUP_ACTS "" FALSE) option(TRACCC_SETUP_VECMEM "" FALSE) + option(TRACCC_SETUP_EIGEN3 "" FALSE) + option(TRACCC_SETUP_THRUST "" TRUE) + option(TRACCC_SETUP_COVFIE "" FALSE) + option(TRACCC_SETUP_DETRAY "" FALSE) + option(TRACCC_BUILD_CUDA "" ${ACTS_ENABLE_CUDA}) + option(TRACCC_SETUP_TBB "" FALSE) + option(TRACCC_BUILD_IO "" TRUE) + option(TRACCC_BUILD_PERFORMANCE "" FALSE) + option(TRACCC_BUILD_SIMULATION "" FALSE) + option(TRACCC_BUILD_TESTING "" FALSE) + option(TRACCC_BUILD_EXAMPLES "" FALSE) + option(TRACCC_BUILD_BENCHMARKS "" FALSE) + find_package(TBB ${_acts_tbb_version} CONFIG) if(NOT TBB_FOUND) set(TRACCC_SETUP_TBB TRUE CACHE BOOL "") @@ -635,13 +651,7 @@ if(ACTS_BUILD_TRACCC) set(TRACCC_SETUP_TBB FALSE CACHE BOOL "") endif() - # traccc also depends on vecmem and covfie, but those plugins should always - # be enabled if traccc is. - if(ACTS_USE_SYSTEM_TRACCC) - find_package(traccc ${_acts_traccc_version} REQUIRED CONFIG) - else() - add_subdirectory(thirdparty/traccc) - endif() + add_subdirectory(Traccc) endif() # examples dependencies diff --git a/Detray/extern/googletest/CMakeLists.txt b/Detray/extern/googletest/CMakeLists.txt index 29354419c97..9f5c0e03dd0 100644 --- a/Detray/extern/googletest/CMakeLists.txt +++ b/Detray/extern/googletest/CMakeLists.txt @@ -34,7 +34,7 @@ else() endif() mark_as_advanced(DETRAY_GOOGLETEST_SOURCE_FULL) -FetchContent_Declare(GoogleTest ${DETRAY_GOOGLETEST_SOURCE_FULL}) +FetchContent_Declare(GoogleTest SYSTEM ${DETRAY_GOOGLETEST_SOURCE_FULL}) # Options used in the build of GoogleTest. set(BUILD_GMOCK TRUE CACHE BOOL "Turn off the build of GMock") diff --git a/Traccc/.github/check_duplicate_cu_files.sh b/Traccc/.github/check_duplicate_cu_files.sh new file mode 100755 index 00000000000..bba4ca19998 --- /dev/null +++ b/Traccc/.github/check_duplicate_cu_files.sh @@ -0,0 +1,4 @@ +#!/bin/bash + +! [[ $(find device tests -name "*\.cu" | xargs -I {} basename {} | sort | uniq -d) ]] +exit $? diff --git a/Traccc/.github/check_quote_includes.sh b/Traccc/.github/check_quote_includes.sh new file mode 100755 index 00000000000..6cc1709cfed --- /dev/null +++ b/Traccc/.github/check_quote_includes.sh @@ -0,0 +1,13 @@ +#!/bin/bash + +if ! [ -x "$(command -v rg)" ]; then + GREP="grep -R" +else + GREP="rg -j 1 --no-heading -N" +fi + +${GREP} "#include \"vecmem" "$@" ; test $? -eq 1 || exit 1 +${GREP} "#include \"detray" "$@" ; test $? -eq 1 || exit 1 +${GREP} "#include \"Acts" "$@" ; test $? -eq 1 || exit 1 +${GREP} "#include \"covfie" "$@" ; test $? -eq 1 || exit 1 +${GREP} "#include \"algebra" "$@" ; test $? -eq 1 || exit 1 diff --git a/Traccc/.github/check_taboos.sh b/Traccc/.github/check_taboos.sh new file mode 100755 index 00000000000..0f03e0becc4 --- /dev/null +++ b/Traccc/.github/check_taboos.sh @@ -0,0 +1,27 @@ +#!/bin/bash + +if ! [ -x "$(command -v rg)" ]; then + GREP="grep -R" +else + GREP="rg -j 1 --no-heading -N" +fi + +INPUT="$@" + +INPUT_EX_THIS_FILE=${INPUT[@]/".github/check_taboos.sh"} + +${GREP} "barcode" ${INPUT_EX_THIS_FILE[@]} ; test $? -eq 1 || exit 1 + +UNIT_CONSTANT_EXCUDE_FILE="core/include/traccc/definitions/common.hpp" +INPUT_EX_UNIT_CONSTANT=${INPUT_EX_THIS_FILE[@]/$UNIT_CONSTANT_EXCUDE_FILE} + +${GREP} "detray::unit" ${INPUT_EX_UNIT_CONSTANT[@]} ; test $? -eq 1 || exit 1 +${GREP} "detray::constant" ${INPUT_EX_UNIT_CONSTANT[@]} ; test $? -eq 1 || exit 1 + +TRACK_PARAMS_EXCLUDE_FILE="core/include/traccc/edm/track_parameters.hpp" +INPUT_EX_TRACK_PARAMS=${INPUT_EX_THIS_FILE[@]/$TRACK_PARAMS_EXCLUDE_FILE} + +${GREP} "detray::free_track_parameters" ${INPUT_EX_TRACK_PARAMS[@]} ; test $? -eq 1 || exit 1 +${GREP} "detray::bound_track_parameters" ${INPUT_EX_TRACK_PARAMS[@]} ; test $? -eq 1 || exit 1 +${GREP} "detray::bound_vector" ${INPUT_EX_TRACK_PARAMS[@]} ; test $? -eq 1 || exit 1 +${GREP} "detray::bound_matrix" ${INPUT_EX_TRACK_PARAMS[@]} ; test $? -eq 1 || exit 1 diff --git a/Traccc/.github/ci_setup.sh b/Traccc/.github/ci_setup.sh new file mode 100644 index 00000000000..37d15f72d6d --- /dev/null +++ b/Traccc/.github/ci_setup.sh @@ -0,0 +1,28 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2022-2024 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 +# +# This script is meant to configure the build/runtime environment of the +# Docker contaners that are used in the project's CI configuration. +# +# Usage: source .github/ci_setup.sh +# + +# The platform name. +PLATFORM_NAME=$1 + +# Make sure that the build and CTest would use all available cores. +export CMAKE_BUILD_PARALLEL_LEVEL=`nproc` +export CTEST_PARALLEL_LEVEL=${CMAKE_BUILD_PARALLEL_LEVEL} +export MAKEFLAGS="-j${CMAKE_BUILD_PARALLEL_LEVEL}" + +# Set up the correct environment for the SYCL tests. +if [[ "${PLATFORM_NAME}" == *"SYCL"* ]]; then + if [ -f "/opt/intel/oneapi/setvars.sh" ]; then + OLD_CPATH=${CPATH} + source /opt/intel/oneapi/setvars.sh --include-intel-llvm + export CPATH=${OLD_CPATH} + fi +fi diff --git a/Traccc/.github/find_f64_ptx.py b/Traccc/.github/find_f64_ptx.py new file mode 100755 index 00000000000..15f0593ea4b --- /dev/null +++ b/Traccc/.github/find_f64_ptx.py @@ -0,0 +1,168 @@ +#!/bin/python3 + + +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2023 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + + +import argparse +import re +import collections +import pathlib + + +class InstructionCounter: + """ + Class for counting the use of certain instructions in translation units. + + The instructions and translation units are counted, not linked. So this + class cannot reconstruct that a particular instruction was emitted in a + particular translation unit. But I don't think that's necessary at this + time.""" + + def __init__(self): + """ + Initialize the counter. + + This creates some empty integer dicts with default value zero.""" + self.instructions = collections.defaultdict(int) + self.translations = collections.defaultdict(int) + + def add(self, instr, trans): + """Register the occurance of an instruction in a translation unit.""" + self.instructions[instr] += 1 + self.translations[trans] += 1 + + +def oxford_join(lst): + """ + Format a list of strings in a human-readable way using an Oxford comma. + + This function takes ["a", "b", "c"] to the string "a, b, and c".""" + if not lst: + return "" + elif len(lst) == 1: + return str(lst[0]) + elif len(lst) == 2: + return f"{str(lst[0])} and {str(lst[1])}" + return f"{', '.join(lst[:-1])}, and {lst[-1]}" + + +def run(files, source, build): + """ + Perform a search for FP64 instructions in a list of files. + + This function takes a list of file paths as well as the root path of the + source code and the build path. These are necessary because the paths + reported in the PTX emitted by NVCC are relative to the build directory. In + order to get the paths relative to the GitHub root directory (which GitHub + Action Commands require), we need to do some path magic.""" + # Create a dictionary of counters. The keys in this dictionary are line + # information tuples (source file name and line) and the values are counter + # objects which count how many times that line generates each instruction, + # and how many times it generates instructions in a given translation unit. + counter = collections.defaultdict(InstructionCounter) + + # Resolve the source and build paths if they are relevant. Since these are + # constant, we can move this operation out of the loop. + source_path = source.resolve() + build_path = build.resolve() + + # Iterate over the list of files that we are given by the user. We do this + # multi-file analysis so we can analyse the mapping of shared source code + # to multiple translation units. + for n in files: + # Read the PTX file and split it into multiple lines. This is NOT a + # proper parsing of PTX and could break, but works for now. + with open(n, "r") as f: + lines = f.read().split("\n") + + # At the beginning of the file, the line data is unknown. + linedata = None + + # Iterate over the source lines in the PTX. + for l in lines: + if m := re.match(r"^//(?P[/\w\-. ]*):(?P\d+)", l): + # If the line of the form "//[filename]:[line] [code]", we + # parse the file name and line number, then update the line + # data. Any subsequent instructions will be mapped onto this + # source line. + linedata = (m.group("file"), int(m.group("line"))) + elif m := re.match( + r"^\s*(?P(?:[a-z][a-z0-9]*)(?:\.[a-z][a-z0-9]+)*)", l + ): + # If the line is of the form " [instruction] [operands]", we + # parse the instruction. The operands are irrelevant. + if "f64" in m.group("instruction"): + # PTX has the pleasant property that all instructions + # explicitly specify their operand types (Intel x86 syntax + # could learn from this), so if "f64" is contained in the + # instruction it will be a double-precision operator. We + # now proceed to compute the real path of the line that + # produced this instruction. + real_path = (build_path / linedata[0]).resolve() + if linedata is not None: + # If the line data is not none, we have a line to link + # this instruction to. We compute the relative path of + # the source file to the root of the source directory, + # and add the result to the counting dictionary. + try: + counter[ + (real_path.relative_to(source_path), linedata[1]) + ].add(m.group("instruction"), n) + except ValueError: + pass + else: + # If we do not have line data, we register an FP64 + # instruction of unknown origin. + counter[None].add(m.group("instruction"), n) + + # After we complete our analysis, we print some output to stdout which will + # be parsed by GitHub Actions. For the syntax, please refer to + # https://docs.github.com/en/actions/using-workflows/workflow-commands-for-github-actions + for dt in counter: + instrs = oxford_join( + [f"{counter[dt].instructions[i]} × `{i}`" for i in counter[dt].instructions] + ) + units = oxford_join( + [f"`{pathlib.Path(f).name}`" for f in counter[dt].translations] + ) + details = ( + f"Instruction(s) generated are {instrs} in translation unit(s) {units}." + ) + + # Handle the cases where the source line information is unknown and + # known, respectively. + if dt is None: + print( + f"::warning title=FP64 instructions emitted in unknown locations::{details}" + ) + else: + print( + f"::warning file={dt[0]},line={dt[1]},title=FP64 instructions emitted::{details}" + ) + + +if __name__ == "__main__": + # Construct an argument parser, asking the user for a set of files, as well + # as their source and build directories, in a fashion similar to what CMake + # does. + parser = argparse.ArgumentParser( + description="Find unwanted 64-bit float operations in annotated PTX." + ) + + parser.add_argument("files", type=str, help="PTX file to use", nargs="+") + parser.add_argument( + "--source", "-S", type=pathlib.Path, help="source directory", required=True + ) + parser.add_argument( + "--build", "-B", type=pathlib.Path, help="build directory", required=True + ) + + args = parser.parse_args() + + # Finally, run the analysis! + run(args.files, args.source, args.build) diff --git a/Traccc/.gitignore b/Traccc/.gitignore new file mode 100644 index 00000000000..5b36d7fc9c7 --- /dev/null +++ b/Traccc/.gitignore @@ -0,0 +1,42 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2021-2025 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# Prerequisites +*.d + +# Compiled Object files +*.slo +*.lo +*.o +*.obj + +# Precompiled Headers +*.gch +*.pch + +# Compiled Dynamic libraries +*.so +*.dylib +*.dll + +# Fortran module files +*.mod +*.smod + +# Compiled Static libraries +*.lai +*.la +*.a +*.lib + +# Executables +*.exe +*.out +*.app + +# Build system related files. +CMakeUserPresets.json +out/ diff --git a/Traccc/CMakeLists.txt b/Traccc/CMakeLists.txt new file mode 100755 index 00000000000..7871f7f5d66 --- /dev/null +++ b/Traccc/CMakeLists.txt @@ -0,0 +1,415 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2021-2026 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# Set up the project. +cmake_minimum_required( VERSION 3.25 ) +project( traccc VERSION 1.1.0 LANGUAGES CXX ) + +# Set up the used C++ standard(s). +set( CMAKE_CXX_STANDARD 20 CACHE STRING "The (host) C++ standard to use" ) +set( CMAKE_CXX_EXTENSIONS FALSE CACHE BOOL "Disable (host) C++ extensions" ) +set( CMAKE_CUDA_STANDARD 20 CACHE STRING "The (CUDA) C++ standard to use" ) +set( CMAKE_CUDA_EXTENSIONS FALSE CACHE BOOL "Disable (CUDA) C++ extensions" ) +set( CMAKE_SYCL_STANDARD 20 CACHE STRING "The (SYCL) C++ standard to use" ) +set( CMAKE_HIP_STANDARD 20 CACHE STRING "The (HIP) C++ standard to use" ) + +if(${CMAKE_CXX_STANDARD} LESS 20) + message(SEND_ERROR "CMAKE_CXX_STANDARD=${CMAKE_CXX_STANDARD}, but traccc requires C++>=20") +endif() + +# Set the CUDA architecture to build code for. +set( CMAKE_CUDA_ARCHITECTURES "75" CACHE STRING + "CUDA architectures to build device code for" ) + +# Set the HIP architecture to build code for. +set( CMAKE_HIP_ARCHITECTURES "gfx1101;gfx90a" CACHE STRING + "HIP architectures to build device code for" ) + +# Flag controlling whether warnings should make the build fail. +option( TRACCC_FAIL_ON_WARNINGS + "Make the build fail on compiler/linker warnings" FALSE ) + +# Standard CMake include(s). +include( GNUInstallDirs ) + +# Explicitly set the output directory for the binaries. Such that if this +# project is included by another project, the main project's configuration would +# win out. +set( CMAKE_RUNTIME_OUTPUT_DIRECTORY + "${CMAKE_BINARY_DIR}/${CMAKE_INSTALL_BINDIR}" CACHE PATH + "Directory for the built binaries" ) +set( CMAKE_LIBRARY_OUTPUT_DIRECTORY + "${CMAKE_BINARY_DIR}/${CMAKE_INSTALL_LIBDIR}" CACHE PATH + "Directory for the built libraries" ) +set( CMAKE_ARCHIVE_OUTPUT_DIRECTORY + "${CMAKE_BINARY_DIR}/${CMAKE_INSTALL_LIBDIR}" CACHE PATH + "Directory for the built static libraries" ) + +# Include the traccc CMake code. +list( APPEND CMAKE_MODULE_PATH "${CMAKE_CURRENT_SOURCE_DIR}/cmake" ) +include( traccc-functions ) + +# Temporary setting for the traccc::scalar type, until it can be removed. +set( TRACCC_CUSTOM_SCALARTYPE "float" CACHE STRING + "Scalar type to use in the TRACCC code" ) + +# Temporary setting for the traccc device log level, until it can be removed. +set( TRACCC_DEVICE_LOG_LVL "NONE" CACHE STRING + "Log level for traccc and detray device code" ) + +# Flags controlling which parts of traccc to build. +option( TRACCC_BUILD_CUDA + "Build the traccc::cuda library" FALSE ) +option( TRACCC_BUILD_CUDA_UTILS + "Build the traccc::cuda_utils library" ${TRACCC_BUILD_CUDA} ) +option( TRACCC_BUILD_HIP "Build the HIP sources included in traccc" FALSE) +option( TRACCC_BUILD_SYCL + "Build the traccc::sycl library" FALSE ) +option( TRACCC_BUILD_SYCL_UTILS + "Build the traccc::sycl_utils library" ${TRACCC_BUILD_SYCL} ) +option( TRACCC_BUILD_ALPAKA "Build the Alpaka sources included in traccc" + FALSE ) +option( TRACCC_BUILD_IO "Build the IO module (needed by examples, performance, testing)" TRUE ) +option( TRACCC_BUILD_PERFORMANCE "Build the performance module" TRUE ) +option( TRACCC_BUILD_SIMULATION "Build the simulation module" TRUE ) +option( TRACCC_BUILD_TESTING "Build the (unit) tests of traccc" TRUE ) +option( TRACCC_BUILD_EXAMPLES "Build the examples of traccc" TRUE ) + +# Flags controlling what traccc should use. +option( TRACCC_USE_SYSTEM_LIBS "Use system libraries be default" FALSE ) +option( TRACCC_USE_SPACK_LIBS "Use Spack libraries by default" FALSE ) +option( TRACCC_USE_ROOT "Use ROOT in the build (if needed)" TRUE ) + +# Check CUDA and SYCL C++ standards +if(${TRACCC_BUILD_CUDA} AND ${CMAKE_CUDA_STANDARD} LESS 20) + message(SEND_ERROR "CMAKE_CUDA_STANDARD=${CMAKE_CUDA_STANDARD}, but traccc requires C++>=20") +endif() + +if(${TRACCC_BUILD_SYCL} AND ${CMAKE_SYCL_STANDARD} LESS 20) + message(SEND_ERROR "CMAKE_SYCL_STANDARD=${CMAKE_SYCL_STANDARD}, but traccc requires C++>=20") +endif() + +if(${TRACCC_BUILD_HIP} AND ${CMAKE_HIP_STANDARD} LESS 20) + message(SEND_ERROR "CMAKE_HIP_STANDARD=${CMAKE_HIP_STANDARD}, but traccc requires C++>=20") +endif() + +# In order to generate specializations for our kernels, we need to have both +# a working Python interpreter, and a list of supported detectors. +find_package (Python COMPONENTS Interpreter REQUIRED) +# The detector types supported by this build of traccc. +set( TRACCC_SUPPORTED_DETECTORS + "default_detector;odd_detector;telescope_detector" + CACHE STRING + "Semicolon-separated list of detector types to support in this build" ) + +# Set up build profiling for the project. +if( CTEST_USE_LAUNCHERS ) + + # Find the bash and time executables. + find_program( BASH_EXECUTABLE bash REQUIRED ) + find_program( TIME_EXECUTABLE time REQUIRED ) + + # Decide what flag to use with the time executable to make it print verbose + # information. + if( "${CMAKE_HOST_SYSTEM_NAME}" MATCHES "Darwin" ) + set( TIME_VERBOSE_FLAG "-l" ) + elseif( "${CMAKE_HOST_SYSTEM_NAME}" MATCHES "Linux" ) + set( TIME_VERBOSE_FLAG "-v" ) + else() + message( WARNING "Build profiling is only supported on Linux and macOS." + "This build will likely fail." ) + set( TIME_VERBOSE_FLAG "" ) + endif() + + # Configure the script that would intercept the build commands and save + # profile logs for them. + configure_file( "${CMAKE_CURRENT_SOURCE_DIR}/cmake/traccc-ctest.sh.in" + "${CMAKE_BINARY_DIR}${CMAKE_FILES_DIRECTORY}/traccc-ctest.sh" + @ONLY ) + set( CMAKE_CTEST_COMMAND + "${CMAKE_BINARY_DIR}${CMAKE_FILES_DIRECTORY}/traccc-ctest.sh" ) + + # Clean up. + unset( TIME_VERBOSE_FLAG ) + + # Remove the performance log during a cleaning step. + set_property( DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}" APPEND PROPERTY + ADDITIONAL_MAKE_CLEAN_FILES + "${CMAKE_CURRENT_BINARY_DIR}/traccc_build_performance.log" ) + + # Let the user know what happened. + message( STATUS + "Saving traccc build performance logs using: ${CMAKE_CTEST_COMMAND}" ) +endif() + +# Set up VecMem. +option( TRACCC_SETUP_VECMEM + "Set up the VecMem target(s) explicitly" TRUE ) +option( TRACCC_USE_SYSTEM_VECMEM + "Pick up an existing installation of VecMem from the build environment" + ${TRACCC_USE_SYSTEM_LIBS} ) +if( TRACCC_SETUP_VECMEM ) + if( TRACCC_USE_SYSTEM_VECMEM ) + find_package( vecmem 1.9.0 REQUIRED ) + else() + add_subdirectory( extern/vecmem ) + endif() + # Make the "VecMem language code" available for the whole project. + list( PREPEND CMAKE_MODULE_PATH "${VECMEM_LANGUAGE_DIR}" ) + list( PREPEND CMAKE_MODULE_PATH "${VECMEM_LANGUAGE_DIR}/sycl" ) + # Including vecmem-check-language makes sure that the HIP and SYCL languages + # would be picked up from VecMem. + include( vecmem-check-language ) +endif() + +# Set up TBB. +option( TRACCC_SETUP_TBB + "Set up the TBB target(s) explicitly" TRUE ) +if (TRACCC_USE_SYSTEM_LIBS OR TRACCC_USE_SPACK_LIBS) + set(TRACCC_USE_SYSTEM_TBB_DEFAULT ON) +else() + set(TRACCC_USE_SYSTEM_TBB_DEFAULT OFF) +endif() +option( TRACCC_USE_SYSTEM_TBB + "Pick up an existing installation of TBB from the build environment" + ${TRACCC_USE_SYSTEM_TBB_DEFAULT} ) +unset(TRACCC_USE_SYSTEM_TBB_DEFAULT) +if( TRACCC_SETUP_TBB ) + if( TRACCC_USE_SYSTEM_TBB ) + find_package( TBB REQUIRED ) + else() + add_subdirectory( extern/tbb ) + endif() +endif() + +# Set up CCCL. +option( TRACCC_SETUP_THRUST + "Set up the Thrust target(s) explicitly" TRUE ) +if (TRACCC_USE_SYSTEM_LIBS OR TRACCC_USE_SPACK_LIBS) + set(TRACCC_USE_SYSTEM_THRUST_DEFAULT ON) +else() + set(TRACCC_USE_SYSTEM_THRUST_DEFAULT OFF) +endif() +option( TRACCC_USE_SYSTEM_THRUST + "Pick up an existing installation of Thrust from the build environment" + ${TRACCC_USE_SYSTEM_THRUST_DEFAULT} ) +unset(TRACCC_USE_SYSTEM_THRUST_DEFAULT) +if( TRACCC_SETUP_THRUST ) + if( TRACCC_USE_SYSTEM_THRUST ) + find_package( Thrust REQUIRED ) + else() + add_subdirectory( extern/cccl ) + endif() +endif() + +# Set up rocThrust. +option( TRACCC_SETUP_ROCTHRUST + "Set up the rocThrust target(s) explicitly" FALSE ) +option( TRACCC_USE_SYSTEM_ROCTHRUST + "Pick up an existing installation of rocThrust from the build environment" + ${TRACCC_USE_SYSTEM_LIBS} ) +if( TRACCC_SETUP_ROCTHRUST ) + if( TRACCC_USE_SYSTEM_ROCTHRUST ) + find_package( rocthrust REQUIRED ) + # Dress up the rocthrust target a little. + target_compile_definitions( roc::rocthrust INTERFACE + THRUST_IGNORE_CUB_VERSION_CHECK ) + else() + add_subdirectory( extern/rocThrust ) + # Dress up the rocthrust target a little. + target_compile_definitions( rocthrust INTERFACE + THRUST_IGNORE_CUB_VERSION_CHECK ) + endif() +endif() + +# Set up DPL if SYCL is built. +option( TRACCC_SETUP_DPL + "Set up the DPL target(s) explicitly" ${TRACCC_BUILD_SYCL} ) +if (TRACCC_USE_SYSTEM_LIBS OR TRACCC_USE_SPACK_LIBS) + set(TRACCC_USE_SYSTEM_DPL_DEFAULT ON) +else() + set(TRACCC_USE_SYSTEM_DPL_DEFAULT OFF) +endif() +option( TRACCC_USE_SYSTEM_DPL + "Pick up an existing installation of DPL from the build environment" + ${TRACCC_USE_SYSTEM_DPL_DEFAULT} ) +unset(TRACCC_USE_SYSTEM_DPL_DEFAULT) +if( TRACCC_SETUP_DPL ) + if( TRACCC_USE_SYSTEM_DPL ) + # OneDPL determines whether SYCL is supported by asking the C++ compiler + # rather than the SYCL compiler, as a dedicated SYCL compiler is a non- + # standard traccc idea. To override the default behaviour (i.e. OneDPL + # testing the C++ compiler and finding that it does _not_ support SYCL) + # we simply override the support flags. This is fragile, as the flags + # are internal to OneDPL, but it's the best we can do. Note thr flag was + # renamed in https://github.com/uxlfoundation/oneDPL/pull/1949. + set(_sycl_support ON) + set(SYCL_SUPPORT ON) + find_package( oneDPL REQUIRED ) + else() + add_subdirectory( extern/dpl ) + endif() + # OneDPL attaches the `-fsycl` flag to the C++ compiler, which causes + # it to incorrectly trigger some preprocessor definitions, thereby + # loading SYCL-specific header files which do not exist for e.g. a + # generic clang installation. Therefore, we have to manually wipe the + # compile flags that OneDPL sets. + set_target_properties(oneDPL PROPERTIES + INTERFACE_COMPILE_OPTIONS "" + INTERFACE_LINK_LIBRARIES "" ) +endif() + +# Set up Alpaka. +option( TRACCC_SETUP_ALPAKA + "Set up the Alpaka library" ${TRACCC_BUILD_ALPAKA}) +if (TRACCC_USE_SYSTEM_LIBS OR TRACCC_USE_SPACK_LIBS) + set(TRACCC_USE_SYSTEM_ALPAKA_DEFAULT ON) +else() + set(TRACCC_USE_SYSTEM_ALPAKA_DEFAULT OFF) +endif() +option( TRACCC_USE_SYSTEM_ALPAKA + "Pick up an existing installation of Alpaka from the build environment" + ${TRACCC_USE_SYSTEM_ALPAKA_DEFAULT} ) +unset(TRACCC_USE_SYSTEM_ALPAKA_DEFAULT) +if( TRACCC_SETUP_ALPAKA ) + # Default options for the Alpaka build. + set( alpaka_ACC_CPU_B_SEQ_T_THREADS_ENABLE TRUE CACHE BOOL + "Enable the serial backend of Alpaka" ) + if( TRACCC_USE_SYSTEM_ALPAKA ) + find_package( alpaka REQUIRED ) + else() + add_subdirectory( extern/alpaka ) + endif() +endif() + +# Set up covfie. +option( TRACCC_SETUP_COVFIE + "Set up the covfie target(s) explicitly" TRUE ) +option( TRACCC_USE_SYSTEM_COVFIE + "Pick up an existing installation of covfie from the build environment" + ${TRACCC_USE_SYSTEM_LIBS} ) +if( TRACCC_SETUP_COVFIE ) + if( TRACCC_USE_SYSTEM_COVFIE ) + find_package( covfie REQUIRED ) + else() + add_subdirectory( extern/covfie ) + endif() +endif() + +# Set up Acts. +option( TRACCC_SETUP_ACTS + "Set up the Acts target(s) explicitly" TRUE ) +if (TRACCC_USE_SYSTEM_LIBS OR TRACCC_USE_SPACK_LIBS) + set(TRACCC_USE_SYSTEM_ACTS_DEFAULT ON) +else() + set(TRACCC_USE_SYSTEM_ACTS_DEFAULT OFF) +endif() +option( TRACCC_USE_SYSTEM_ACTS + "Pick up an existing installation of Acts from the build environment" + ${TRACCC_USE_SYSTEM_ACTS_DEFAULT} ) +unset(TRACCC_USE_SYSTEM_ACTS_DEFAULT) +if( TRACCC_SETUP_ACTS ) + if( TRACCC_USE_SYSTEM_ACTS ) + find_package( Acts REQUIRED COMPONENTS PluginJson ) + find_package( Detray REQUIRED ) + else() + add_subdirectory( extern/acts ) + endif() +endif() + +# Set up GoogleTest. +include( CTest ) +if (BUILD_TESTING AND TRACCC_BUILD_TESTING) +set (TRACCC_DEFAULT_SETUP_GOOGLETEST TRUE) +endif () +option( TRACCC_SETUP_GOOGLETEST + "Set up the GoogleTest target(s) explicitly" ${TRACCC_DEFAULT_SETUP_GOOGLETEST} ) +if (TRACCC_USE_SYSTEM_LIBS OR TRACCC_USE_SPACK_LIBS) + set(TRACCC_USE_SYSTEM_GOOGLETEST_DEFAULT ON) +else() + set(TRACCC_USE_SYSTEM_GOOGLETEST_DEFAULT OFF) +endif() +option( TRACCC_USE_SYSTEM_GOOGLETEST + "Pick up an existing installation of GoogleTest from the build environment" + ${TRACCC_USE_SYSTEM_GOOGLETEST_DEFAULT} ) +unset(TRACCC_USE_SYSTEM_GOOGLETEST_DEFAULT) +if( TRACCC_SETUP_GOOGLETEST ) + if( TRACCC_USE_SYSTEM_GOOGLETEST ) + find_package( GTest REQUIRED ) + else() + add_subdirectory( extern/googletest ) + endif() +endif() + + +option( TRACCC_ENABLE_NVTX_PROFILING + "Use instrument functions to enable fine grained profiling" FALSE ) + +# Build the traccc code. +add_subdirectory( core ) +add_subdirectory( device/common ) +if( TRACCC_BUILD_CUDA_UTILS ) + add_subdirectory( device/cuda_utils ) +endif() +if( TRACCC_BUILD_CUDA ) + add_subdirectory( device/cuda ) +endif() +if( TRACCC_BUILD_SYCL_UTILS ) + add_subdirectory( device/sycl_utils ) +endif() +if( TRACCC_BUILD_SYCL ) + add_subdirectory( device/sycl ) +endif() +if( TRACCC_BUILD_ALPAKA ) + add_subdirectory( device/alpaka ) +endif() +if ( TRACCC_BUILD_IO ) + add_subdirectory( io ) +else() + message(STATUS "traccc::io not built, traccc::performance and traccc::simulation are forcefully switched off.") +endif() + +if ( TRACCC_BUILD_PERFORMANCE ) + if ( NOT TRACCC_BUILD_IO ) + message(FATAL_ERROR "traccc::io is disabled, but it is required to build the performance tests.") + endif() + add_subdirectory( performance ) +endif() +if ( TRACCC_BUILD_SIMULATION ) + if ( NOT TRACCC_BUILD_IO ) + message(FATAL_ERROR "traccc::io is disabled, but it is required to build the simulation.") + endif() + add_subdirectory( simulation ) +endif() + +if ( TRACCC_BUILD_EXAMPLES ) + # Find Boost. + find_package( Boost CONFIG REQUIRED COMPONENTS program_options filesystem ) + if ( NOT TRACCC_BUILD_IO ) + message(FATAL_ERROR "traccc::io is disabled, but it is required to build the examples.") + endif() + if ( NOT TRACCC_BUILD_PERFORMANCE ) + message(FATAL_ERROR "traccc::performance is disabled, but it is required to build the examples.") + endif() + add_subdirectory( examples ) +endif() + +# Set up the test(s). +if( BUILD_TESTING AND TRACCC_BUILD_TESTING ) + if ( NOT TRACCC_BUILD_IO ) + message(FATAL_ERROR "traccc::io is disabled, but it is required to build the tests.") + endif() + if ( NOT TRACCC_BUILD_SIMULATION ) + message(FATAL_ERROR "traccc::simulation is disabled, but it is required to build the tests.") + endif() + if ( NOT TRACCC_BUILD_PERFORMANCE ) + message(FATAL_ERROR "traccc::performance is disabled, but it is required to build the tests.") + endif() + add_subdirectory( tests ) +endif() + +# Set up the packaging of the project. +include( traccc-packaging ) diff --git a/Traccc/CMakePresets.json b/Traccc/CMakePresets.json new file mode 100644 index 00000000000..3acfc04b68a --- /dev/null +++ b/Traccc/CMakePresets.json @@ -0,0 +1,157 @@ +{ + "version" : 3, + "configurePresets": [ + { + "name" : "base-fp32", + "displayName" : "Base FP32 Configuration", + "warnings": { + "deprecated": true + }, + "cacheVariables": { + "CMAKE_BUILD_TYPE" : "RelWithDebInfo", + "TRACCC_BUILD_TESTING" : "TRUE", + "TRACCC_BUILD_EXAMPLES" : "TRUE", + "TRACCC_FAIL_ON_WARNINGS" : "TRUE", + "TRACCC_USE_ROOT" : "FALSE", + "TRACCC_CUSTOM_SCALARTYPE" : "float", + "DETRAY_GENERATE_METADATA" : "itk_metadata" + } + }, + { + "name" : "base-fp64", + "displayName" : "Base FP64 Configuration", + "warnings": { + "deprecated": true + }, + "cacheVariables": { + "CMAKE_BUILD_TYPE" : "RelWithDebInfo", + "TRACCC_BUILD_TESTING" : "TRUE", + "TRACCC_BUILD_EXAMPLES" : "TRUE", + "TRACCC_FAIL_ON_WARNINGS" : "TRUE", + "TRACCC_USE_ROOT" : "FALSE", + "TRACCC_CUSTOM_SCALARTYPE" : "double", + "DETRAY_GENERATE_METADATA" : "itk_metadata" + } + }, + { + "name" : "cuda-fp32", + "displayName" : "CUDA FP32 Code Development", + "inherits": ["base-fp32"], + "cacheVariables": { + "TRACCC_BUILD_CUDA_UTILS" : "TRUE", + "TRACCC_BUILD_CUDA" : "TRUE", + "VECMEM_BUILD_CUDA_LIBRARY" : "TRUE" + } + }, + { + "name" : "cuda-fp64", + "displayName" : "CUDA FP64 Code Development", + "inherits": ["base-fp64"], + "cacheVariables": { + "TRACCC_BUILD_CUDA_UTILS" : "TRUE", + "TRACCC_BUILD_CUDA" : "TRUE", + "VECMEM_BUILD_CUDA_LIBRARY" : "TRUE" + } + }, + { + "name" : "sycl-fp32", + "displayName" : "SYCL FP32 Code Development", + "inherits": ["base-fp32"], + "cacheVariables": { + "TRACCC_BUILD_SYCL_UTILS" : "TRUE", + "TRACCC_BUILD_SYCL" : "TRUE", + "VECMEM_BUILD_SYCL_LIBRARY" : "TRUE" + } + }, + { + "name" : "sycl-fp64", + "displayName" : "SYCL FP64 Code Development", + "inherits": ["base-fp64"], + "cacheVariables": { + "TRACCC_BUILD_SYCL_UTILS" : "TRUE", + "TRACCC_BUILD_SYCL" : "TRUE", + "VECMEM_BUILD_SYCL_LIBRARY" : "TRUE" + } + }, + { + "name" : "alpaka-fp32", + "displayName" : "Alpaka FP32 Code Development", + "inherits": ["base-fp32"], + "cacheVariables": { + "TRACCC_BUILD_ALPAKA" : "TRUE" + } + }, + { + "name" : "alpaka-fp64", + "displayName" : "Alpaka FP64 Code Development", + "inherits": ["base-fp64"], + "cacheVariables": { + "TRACCC_BUILD_ALPAKA" : "TRUE" + } + }, + { + "name" : "alpaka-fp32-cuda", + "displayName" : "Alpaka CUDA FP32 Code Development", + "inherits": ["alpaka-fp32", "cuda-fp32"], + "cacheVariables": { + "alpaka_ACC_GPU_CUDA_ENABLE" : "TRUE" + } + }, + { + "name" : "alpaka-fp64-cuda", + "displayName" : "Alpaka CUDA FP64 Code Development", + "inherits": ["alpaka-fp64", "cuda-fp64"], + "cacheVariables": { + "alpaka_ACC_GPU_CUDA_ENABLE" : "TRUE" + } + }, + { + "name": "alpaka-fp32-hip", + "displayName": "Alpaka HIP FP32 Code Development", + "inherits": ["alpaka-fp32"], + "cacheVariables": { + "alpaka_ACC_GPU_HIP_ENABLE" : "TRUE", + "TRACCC_BUILD_HIP": "ON", + "TRACCC_SETUP_ROCTHRUST": "ON" + } + }, + { + "name": "alpaka-fp64-hip", + "displayName": "Alpaka HIP FP64 Code Development", + "inherits": ["alpaka-fp64"], + "cacheVariables": { + "alpaka_ACC_GPU_HIP_ENABLE" : "TRUE", + "TRACCC_BUILD_HIP": "ON", + "TRACCC_SETUP_ROCTHRUST": "ON" + } + }, + { + "name" : "host-fp32", + "displayName": "Host FP32 Code Development", + "inherits": ["base-fp32"], + "cacheVariables": { + "TRACCC_USE_ROOT" : "TRUE", + "TRACCC_USE_SYSTEM_TBB" : "TRUE" + } + }, + { + "name" : "host-fp64", + "displayName": "Host FP64 Code Development", + "inherits": ["base-fp64"], + "cacheVariables": { + "TRACCC_USE_ROOT" : "TRUE", + "TRACCC_USE_SYSTEM_TBB" : "TRUE" + } + }, + { + "name" : "full-fp32", + "displayName": "Full FP32 Code Development", + "inherits": ["host-fp32", "cuda-fp32", "sycl-fp32", "alpaka-fp32"] + }, + { + "name" : "full-fp64", + "displayName": "Full FP64 Code Development", + "inherits": ["host-fp64", "cuda-fp64", "sycl-fp64", "alpaka-fp64"] + } + ] +} diff --git a/Traccc/README.md b/Traccc/README.md new file mode 100644 index 00000000000..30afa5f6715 --- /dev/null +++ b/Traccc/README.md @@ -0,0 +1,392 @@ +# traccc + +Demonstrator tracking chain for accelerators. + +## Features + +| Category | Algorithms | CPU | CUDA | SYCL | Alpaka | +| ------------------------- | ---------------------- | --- | ---- | ---- | ------ | +| **Clusterization** | CCL / FastSv / etc. | ✅ | ✅ | ✅ | ✅ | +| | Measurement creation | ✅ | ✅ | ✅ | ✅ | +| **Seeding** | Spacepoint formation | ✅ | ✅ | ✅ | ✅ | +| | Spacepoint binning | ✅ | ✅ | ✅ | ✅ | +| | Seed finding | ✅ | ✅ | ✅ | ✅ | +| | Track param estimation | ✅ | ✅ | ✅ | ✅ | +| **Track finding** | Combinatorial KF | ✅ | ✅ | ✅ | ✅ | +| **Ambiguity resolution** | Greedy resolver | ✅ | 🟡 | ⚪ | ⚪ | +| **Track fitting** | KF | ✅ | ✅ | 🟡 | ✅ | + +✅: exists, 🟡: work started, ⚪: work not started yet + +The relations between datatypes and algorithms is given in the (approximately +commutative) diagram shown below. Black lines indicate CPU algorithms, green +lines indicate CUDA algorithms, blue lines indicate SYCL algorithms. Solid algorithms are ready for use, dashed +algorithms are in development or future goals. Data types for different +heterogeneous platforms are contracted for legibility, and identities are +hidden. + +```mermaid +flowchart LR + subgraph clusterization [Clusterization] + direction TB + cell(Cells); + cluster(Clusters); + meas(Measurements); + end + + subgraph trkfinding [Track Finding] + sp(Spacepoints); + bin(Spacepoint Grid); + seed(Seeds); + ptrack(Prototracks); + end + + subgraph trkfitting [Track Fitting] + trackc(Track Candidates) + tracks(Track States); + end + + click cell href "https://github.com/acts-project/traccc/blob/main/core/include/traccc/edm/silicon_cell_collection.hpp"; + click cluster href "https://github.com/acts-project/traccc/blob/main/core/include/traccc/edm/silicon_cluster_collection.hpp"; + click meas href "https://github.com/acts-project/traccc/blob/main/core/include/traccc/edm/measurement.hpp"; + click sp href "https://github.com/acts-project/traccc/blob/main/core/include/traccc/edm/spacepoint.hpp"; + click bin href "https://github.com/acts-project/traccc/blob/main/core/include/traccc/seeding/detail/spacepoint_grid.hpp"; + click seed href "https://github.com/acts-project/traccc/blob/main/core/include/traccc/edm/seed.hpp"; + click ptrack href "https://github.com/acts-project/traccc/blob/main/core/include/traccc/edm/track_parameters.hpp"; + click trackc href "https://github.com/acts-project/traccc/blob/main/core/include/traccc/edm/track_candidate.hpp"; + click tracks href "https://github.com/acts-project/traccc/blob/main/core/include/traccc/edm/track_state.hpp"; + + %% Host CCL algorithm + cell -->|Sparse CCL| cluster; + linkStyle 0 stroke: black; + + %% Host measurement creation + cluster -->|Meas. Creat.| meas; + linkStyle 1 stroke: black; + + %% SYCL clusterization + cell -->|Clustering| meas; + linkStyle 2 stroke: blue; + + %% CUDA clusterization + cell -->|Clustering| meas; + linkStyle 3 stroke: green; + + %% Alpaka clusterization + cell -->|Clustering| meas; + linkStyle 4 stroke: orange; + + %% Host spacepoint formation + meas -->|SP Form.| sp; + linkStyle 5 stroke: black; + + %% SYCL spacepoint formation + meas -->|SP Form.| sp; + linkStyle 6 stroke: blue; + + %% CUDA spacepoint formation + meas -->|SP Form.| sp; + linkStyle 7 stroke: green; + + %% Alpaka spacepoint formation + meas -->|SP Form.| sp; + linkStyle 8 stroke: orange; + + %% Host spacepoint binning + sp -->|SP Binning| bin; + linkStyle 9 stroke: black; + + %% SYCL triplet seeding + sp -->|Triplet Seeding| seed; + linkStyle 10 stroke: blue; + + %% CUDA triplet seeding + sp -->|Triplet Seeding| seed; + linkStyle 11 stroke: green; + + %% Alpaka triplet seeding + sp -->|Triplet Seeding| seed; + linkStyle 12 stroke: orange; + + %% Host seeding + bin -->|Seeding| seed; + linkStyle 13 stroke: black; + + %% Host param est. + seed -->|Param. Est.| ptrack; + linkStyle 14 stroke: black; + + %% SYCL param est. + seed -->|Param. Est.| ptrack; + linkStyle 15 stroke: blue; + + %% CUDA param est. + seed -->|Param. Est.| ptrack; + linkStyle 16 stroke: green; + + %% Alpaka param est. + seed -->|Param. Est.| ptrack; + linkStyle 17 stroke: orange; + + %% Host CKF + ptrack -->|CKF| trackc; + linkStyle 18 stroke: black; + + %% Host Kalman filter + trackc -->|Kalman filter| tracks; + linkStyle 19 stroke: black; + + %% SYCL CKF + ptrack -->|CKF| trackc; + linkStyle 20 stroke: blue; + + %% SYCL Kalman filter + trackc -->|Kalman filter| tracks; + linkStyle 21 stroke: blue; + + %% CUDA CKF + ptrack -->|CKF| trackc; + linkStyle 22 stroke: green; + + %% CUDA Kalman filter + trackc -->|Kalman filter| tracks; + linkStyle 23 stroke: green; +``` + +## Requirements and dependencies + +### OS & compilers: + +Please note that due to the complexity of this software and its build system, +it may be somewhat fragile in the face of compiler version changes. The +following are general guidelines for getting _traccc_ to compile: + +- The C++ compiler must support C++20 + +In addition, the following requirements hold when CUDA is enabled: + +- The CUDA Toolkit version must be greater than major version 12 +- Ensure that the CUDA host compiler supports C++20 and is compatible with the + `nvcc` compiler driver + +The following table lists currently combinations of builds, compilers, +and toolchains that are currently known to work (last updated 2022/01/24): + +| Build | OS | gcc | CUDA | comment | +| --- | --- | --- | --- | --- | +| CUDA | Ubuntu 24.04 | 13.3.0 | 12.6 | runs on CI | + +### Dependencies + +- [Boost](https://www.boost.org/): program_options +- [CMake](https://cmake.org/) +- (Optional) [ROOT](https://root.cern/): RIO, Hist, Tree + +### Dependency management with Spack + +The [Spack](https://spack.io/) project provides a particularly easy way to +install the dependencies that you need to use traccc. In order to use Spack to +manage your dependencies, simply create a new Spack environment using the +provided environment file: + +```sh +spack env create traccc spack.yaml +spack -e traccc concretize -f +spack -e traccc install +spack env activate traccc +``` + +This way, Spack will automatically download and install all dependencies +necessary to use traccc with the CUDA, SYCL, and Alpaka programming +models. When using Spack to manage your dependencies, make sure to compile +traccc with the `-DTRACCC_USE_SPACK_LIBS=ON` flag. + +## Getting started + +### Clone the repository + +Clone the repository and setup the data directory. + +```sh +git clone https://github.com/acts-project/traccc.git +./traccc/data/traccc_data_get_files.sh +``` + +### Build the project + +```sh +cmake --preset [options] -S traccc/ -B build +cmake --build build/ +``` + +### Build presets + +| Name | Description | +| --- | --- | +| host-fp32 | FP32 build of the host code with ROOT enabled | +| host-fp64 | FP64 build of the host code with ROOT enabled | +| cuda-fp32 | FP32 build of the CUDA code | +| cuda-fp64 | FP64 build of the CUDA code | +| sycl-fp32 | FP32 build of the SYCL code | +| sycl-fp64 | FP64 build of the SYCL code | +| alpaka-fp32 | FP32 build of the Alpaka code | +| alpaka-fp64 | FP64 build of the Alpaka code | + +### Detailed build options + +| Option | Description | +| --- | --- | +| TRACCC_BUILD_CUDA | Build the CUDA sources included in traccc | +| TRACCC_BUILD_SYCL | Build the SYCL sources included in traccc | +| TRACCC_BUILD_ALPAKA | Build the Alpaka sources included in traccc | +| TRACCC_BUILD_TESTING | Build the (unit) tests of traccc | +| TRACCC_BUILD_EXAMPLES | Build the examples of traccc | +| TRACCC_USE_SYSTEM_VECMEM | Pick up an existing installation of VecMem from the build environment | +| TRACCC_USE_SYSTEM_ACTS | Pick up an existing installation of Acts from the build environment | +| TRACCC_USE_SYSTEM_GOOGLETEST | Pick up an existing installation of GoogleTest from the build environment | +| TRACCC_USE_ROOT | Build physics performance analysis code using an existing installation of ROOT from the build environment | +| TRACCC_DEVICE_LOG_LVL | Set device code log level (NONE, WARN, INFO, VERBOSE, DEBUG) + +## Examples + +### Full reconstruction chain + +```console +/bin/traccc_seq_example \ + --detector-file=geometries/odd/odd-detray_geometry_detray.json \ + --material-file=geometries/odd/odd-detray_material_detray.json \ + --grid-file=geometries/odd/odd-detray_surface_grids_detray.json \ + --digitization-file=geometries/odd/odd-digi-geometric-config.json \ + --conditions-file=geomteries/odd/odd-digi-geometric-config.json \ + --use-acts-geom-source --input-directory=odd/geant4_10muon_10GeV/ \ + --input-events=10 + +/bin/traccc_throughput_mt \ + --detector-file=geometries/odd/odd-detray_geometry_detray.json \ + --material-file=geometries/odd/odd-detray_material_detray.json \ + --grid-file=geometries/odd/odd-detray_surface_grids_detray.json \ + --digitization-file=geometries/odd/odd-digi-geometric-config.json \ + --conditions-file=geometries/odd/odd-digi-geometric-config.json \ + --use-acts-geom-source --input-directory=odd/geant4_10muon_10GeV/ \ + --input-events=10 --processed-events=1000 --threads=1 +``` + +Depending on the build options, can also use variants of the executables +postfixed by `_cuda`, `_sycl` and `_alpaka`, with the same options. + +### Running a partial chain with simplified simulation data + +Users can generate muon-like particle simulation data with the pre-built detray geometries: + +```sh +# Generate telescope geometry data +/bin/traccc_simulate_telescope --gen-vertex-xyz-mm=0:0:0 --gen-vertex-xyz-std-mm=0:0:0 --gen-mom-gev=100:100 --gen-phi-degree=0:0 --gen-events=10 --gen-nparticles=2000 --output-directory=detray_simulation/telescope_detector/n_particles_2000/ --gen-eta=1:3 + +# Generate toy geometry data +/bin/traccc_simulate_toy_detector --gen-vertex-xyz-mm=0:0:0 --gen-vertex-xyz-std-mm=0:0:0 --gen-mom-gev=100:100 --gen-phi-degree=0:360 --gen-events=10 --gen-nparticles=2000 --output-directory=detray_simulation/toy_detector/n_particles_2000/ --gen-eta=-3:3 --constraint-step-size-mm=1 --search-window 3:3 + +# Generate drift chamber data +/bin/traccc_simulate_wire_chamber --gen-vertex-xyz-mm=0:0:0 --gen-vertex-xyz-std-mm=0:0:0 --gen-mom-gev=2:2 --gen-phi-degree=0:360 --gen-events=10 --gen-nparticles=100 --output-directory=detray_simulation/wire_chamber/n_particles_100/ --gen-eta=-1:1 --constraint-step-size-mm=1 --search-window 3:3 +``` + +The simulation will also generate the detector json files (geometry, material and surface_grid) in the current directory. It is user's responsibility to move them to an appropriate place (e.g. ``) and match them to the input file arguments of reconstruction chains. + +If users have a geometry json file already, it is also possible to run simulation with `traccc_simulate` application + +```sh +# Given that users have a geometry json file +/bin/traccc_simulate --output-directory= --detector-file= --material-file= --grid-file= --event=10 --constraint-step-size-mm=1 +``` + +There are three types of partial reconstruction chain users can operate: `seeding_example`, `truth_finding_example`, and `truth_fitting_example` where their algorithm coverages are shown in the table below. Each of them starts from truth measurements, truth seeds, and truth tracks, respectively. + +| Category | Clusterization | Seeding | Track finding | Track fitting | +| ----------------------- | -------------- | ------- | ------------- | ------------- | +| `seeding_example` | | ✅ | ✅ | ✅ | +| `truth_finding_example` | | | ✅ | ✅ | +| `truth_fitting_example` | | | | ✅ | + +The dirft chamber will not produce meaningful results with `seeding_example` as the current seeding algorithm is only designed for 2D measurement objects. Truth finding works OK in general but the combinatoric explosion can occur for a few unlucky events, leading to poor pull value distributions. The followings are example commands: + +```sh +# Run cuda seeding example for toy geometry +/bin/traccc_seeding_example_cuda --input-directory=detray_simulation/toy_detector/n_particles_2000/ --check-performance --detector-file=/toy_detector_geometry.json --material-file=/toy_detector_homogeneous_material.json --grid-file=/toy_detector_surface_grids.json --input-events=1 --track-candidates-range=3:30 --constraint-step-size-mm=1000 --run-cpu=1 --search-window 3:3 +``` + +```sh +# Run cuda truth finding example for toy geometry +/bin/traccc_truth_finding_example_cuda --input-directory=detray_simulation/toy_detector/n_particles_2000/ --check-performance --detector-file=/toy_detector_geometry.json --material-file=/toy_detector_homogeneous_material.json --grid-file=/toy_detector_surface_grids.json --input-events=1 --track-candidates-range=3:30 --constraint-step-size-mm=1000 --run-cpu=1 --search-window 3:3 +``` + +```sh +# Run cuda truth finding example for drift chamber +/bin/traccc_truth_finding_example_cuda --input-directory=detray_simulation/wire_chamber/n_particles_100/ --check-performance --detector-file=/wire_chamber_geometry.json --material-file=/wire_chamber_homogeneous_material.json --grid-file=/wire_chamber_surface_grids.json --input-events=10 --track-candidates-range=6:30 --constraint-step-size-mm=1 --run-cpu=1 --search-window 3:3 +``` + +```sh +# Run cpu truth fitting example for drift chamber +/bin/traccc_truth_fitting_example --input-directory=detray_simulation/wire_chamber/n_particles_2000_100GeV/ --check-performance --detector-file=/wire_chamber_geometry.json --material-file=/wire_chamber_homogeneous_material.json --grid-file=/wire_chamber_surface_grids.json --input-events=10 --constraint-step-size-mm=1 --search-window 3:3 +``` + +Users can open the performance root files (with `--check-performance=true`) and draw the histograms. + +```sh +$ root -l performance_track_finding.root +root [0] +Attaching file performance_track_finding.root as _file0... +(TFile *) 0x3871910 +root [1] finding_trackeff_vs_eta->Draw() +``` + +## Contributing + +### Code formatting + +The traccc code is formatted using clang-format; the recommended way to ensure that your code is properly formatted is to use [pre-commit](https://pre-commit.com/). The pre-commit webpage has a useful guide for using the tool, but the simplest way of using it (without installing it as a pre-commit hook) is as follows. First, install the tool with your favourite Python package manager: + +```console +# With pip +$ pip install pre-commit +# With pipx +$ pip install pre-commit +``` + +The install step needs to be executed only once. After that, the code can be easily formatted as follows: + +```console +$ pre-commit run --all-files +``` + +## Troubleshooting + +The following are potentially useful instructions for troubleshooting various +problems with your build: + +### CUDA + +#### Incompatible host compiler + +You may experience errors being issued about standard library features, for example: + +``` +/usr/include/c++/11/bits/std_function.h:435:145: note: ‘_ArgTypes’ +/usr/include/c++/11/bits/std_function.h:530:146: error: parameter packs not expanded with ‘...’: + 530 | operator=(_Functor&& __f) +``` + +In this case, your `nvcc` host compiler is most likely incompatible with your +CUDA toolkit. Consider installing a supported version and selecting it through +the `CUDAHOSTCXX` environment variable at build-time. + +#### `ptxas` failures due to non-ASCII characters + +You may encounter errors that look like the following: + +``` +ptxas fatal : Unexpected non-ASCII character encountered on line 30 +ptxas fatal : Ptx assembly aborted due to errors +``` + +This means that you are compiling with a version later than CUDA 12.8 which has +not been fixed and you are running in debug mode. This error is due to a bug in +the CUDA toolkit. Either downgrade to CUDA 12.6 or disable the debug build. diff --git a/Traccc/cmake/FindROOT.cmake b/Traccc/cmake/FindROOT.cmake new file mode 100644 index 00000000000..e826dc14441 --- /dev/null +++ b/Traccc/cmake/FindROOT.cmake @@ -0,0 +1,22 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2023 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# This is a simple wrapper around the ROOTConfig.cmake file that would come +# with a ROOT installation. It serves two purposes: +# - Provides a nice printout about ROOT having been found, which +# ROOTConfig.cmake doesn't do for some reason; +# - Avoids printing scary-looking warnings from CMake when ROOT is +# not found. + +# Look for ROOTConfig.cmake. +find_package( ROOT CONFIG QUIET ) + +# Print a standard output about ROOT (not) having been found. +include( FindPackageHandleStandardArgs ) +find_package_handle_standard_args( ROOT + FOUND_VAR ROOT_FOUND + REQUIRED_VARS ROOT_INCLUDE_DIRS ROOT_LIBRARY_DIR ROOT_BINDIR + VERSION_VAR ROOT_VERSION ) diff --git a/Traccc/cmake/traccc-alpaka-functions.cmake b/Traccc/cmake/traccc-alpaka-functions.cmake new file mode 100644 index 00000000000..abf134bb89c --- /dev/null +++ b/Traccc/cmake/traccc-alpaka-functions.cmake @@ -0,0 +1,108 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2024-2026 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +cmake_minimum_required( VERSION 3.16 ) + +# Guard against multiple includes. +include_guard( GLOBAL ) + +# Function for declaring the libraries of the project. +# This version calls the alpaka_add_library() function to create the library, +# which is setup to use the correct compiler flags depending on the build type. +# +# Usage: traccc_add_alpaka_library( traccc_core core +# [TYPE SHARED/INTERFACE/STATIC] +# include/source1.hpp source2.cpp ) +# +function( traccc_add_alpaka_library fullname basename ) + + # Parse the function's options. + cmake_parse_arguments( ARG "" "TYPE" "" ${ARGN} ) + + # Decide what sources to give to the library. + set( _sources ${ARG_UNPARSED_ARGUMENTS} ) + if( "${ARG_TYPE}" STREQUAL "INTERFACE" ) + set( _sources ) + endif() + + # Create the library. + alpaka_add_library( ${fullname} ${ARG_TYPE} ${_sources} ) + + # Set up how clients should find its headers. + if( IS_DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/include/" ) + set( _depType PUBLIC ) + if( "${ARG_TYPE}" STREQUAL "INTERFACE" ) + set( _depType INTERFACE ) + endif() + target_include_directories( ${fullname} ${_depType} + $ + $ ) + unset( _depType ) + endif() + + # Make sure that the library is available as "traccc::${basename}" in every + # situation. + set_target_properties( ${fullname} PROPERTIES EXPORT_NAME ${basename} ) + add_library( traccc::${basename} ALIAS ${fullname} ) + + # Specify the (SO)VERSION of the library. + if( NOT "${ARG_TYPE}" STREQUAL "INTERFACE" ) + set_target_properties( ${fullname} PROPERTIES + VERSION ${PROJECT_VERSION} + SOVERSION ${PROJECT_VERSION_MAJOR} ) + endif() + + # Set up the installation of the library and its headers. + install( TARGETS ${fullname} + EXPORT traccc-exports + LIBRARY DESTINATION "${CMAKE_INSTALL_LIBDIR}" + ARCHIVE DESTINATION "${CMAKE_INSTALL_LIBDIR}" + RUNTIME DESTINATION "${CMAKE_INSTALL_BINDIR}" ) + if( IS_DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/include/" ) + install( DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/include/" + DESTINATION "${CMAKE_INSTALL_INCLUDEDIR}" ) + endif() + +endfunction( traccc_add_alpaka_library ) + +# Helper function for setting up the Alpaka traccc test(s). +# +# Usage: traccc_add_alpaka_test( core_containers source1.cpp source2.cpp +# LINK_LIBRARIES traccc::core ) +# +function( traccc_add_alpaka_test name ) + + # Parse the function's options. + cmake_parse_arguments( ARG "" "" "LINK_LIBRARIES" ${ARGN} ) + + # Create the test executable. + set( test_exe_name "traccc_test_${name}" ) + alpaka_add_executable( ${test_exe_name} ${ARG_UNPARSED_ARGUMENTS} ) + if( ARG_LINK_LIBRARIES ) + target_link_libraries( ${test_exe_name} PRIVATE ${ARG_LINK_LIBRARIES} ) + endif() + + # Discover all of the tests from the execuable, and set them up as individual + # CTest tests. All the while ensuring that they would find their data files. + gtest_discover_tests( ${test_exe_name} + PROPERTIES ENVIRONMENT + TRACCC_TEST_DATA_DIR=${PROJECT_SOURCE_DIR}/data + DISCOVERY_TIMEOUT 20 ) + +endfunction( traccc_add_alpaka_test ) + +macro (traccc_enable_language_alpaka) + #enable_language cannot be called by a function: put it in a macro + if(alpaka_ACC_GPU_CUDA_ENABLE) + enable_language(CUDA) + include( traccc-compiler-options-cuda ) + elseif(alpaka_ACC_GPU_HIP_ENABLE) + enable_language(HIP) + elseif(alpaka_ACC_SYCL_ENABLE) + enable_language(SYCL) + include( traccc-compiler-options-sycl ) + endif() +endmacro(traccc_enable_language_alpaka) diff --git a/Traccc/cmake/traccc-compiler-options-cpp.cmake b/Traccc/cmake/traccc-compiler-options-cpp.cmake new file mode 100644 index 00000000000..032f3d192a4 --- /dev/null +++ b/Traccc/cmake/traccc-compiler-options-cpp.cmake @@ -0,0 +1,64 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2021-2024 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# Include the helper function(s). +include( traccc-functions ) +include( CheckCXXCompilerFlag ) + +# Turn on the correct setting for the __cplusplus macro with MSVC. +if( "${CMAKE_CXX_COMPILER_ID}" MATCHES "MSVC" ) + traccc_add_flag( CMAKE_CXX_FLAGS "/Zc:__cplusplus" ) +endif() + +# Turn on a number of warnings for the "known compilers". +if( ( "${CMAKE_CXX_COMPILER_ID}" MATCHES "GNU" ) OR + ( "${CMAKE_CXX_COMPILER_ID}" MATCHES "Clang" ) OR + ( "${CMAKE_CXX_COMPILER_ID}" STREQUAL "IntelLLVM" ) ) + + # Basic flags for all build modes. + traccc_add_flag( CMAKE_CXX_FLAGS "-Wall" ) + traccc_add_flag( CMAKE_CXX_FLAGS "-Wextra" ) + traccc_add_flag( CMAKE_CXX_FLAGS "-Wshadow" ) + traccc_add_flag( CMAKE_CXX_FLAGS "-Wunused-local-typedefs" ) + traccc_add_flag( CMAKE_CXX_FLAGS "-Wpedantic" ) + traccc_add_flag( CMAKE_CXX_FLAGS "-Wold-style-cast" ) + traccc_add_flag( CMAKE_CXX_FLAGS "-Wzero-as-null-pointer-constant" ) + traccc_add_flag( CMAKE_CXX_FLAGS "-Woverloaded-virtual" ) + if(PROJECT_IS_TOP_LEVEL) + traccc_add_flag( CMAKE_CXX_FLAGS "-Wconversion" ) + endif() + + # Fail on warnings, if asked for that behaviour. + if( TRACCC_FAIL_ON_WARNINGS ) + traccc_add_flag( CMAKE_CXX_FLAGS "-Werror" ) + endif() + +elseif( "${CMAKE_CXX_COMPILER_ID}" MATCHES "MSVC" ) + + # Basic flags for all build modes. + string( REGEX REPLACE "/W[0-9]" "" CMAKE_CXX_FLAGS + "${CMAKE_CXX_FLAGS}" ) + traccc_add_flag( CMAKE_CXX_FLAGS "/W4" ) + + # Fail on warnings, if asked for that behaviour. + if( TRACCC_FAIL_ON_WARNINGS ) + traccc_add_flag( CMAKE_CXX_FLAGS "/WX" ) + endif() + +endif() + +# Set architecture level to v2 - enable generation of code using SSE4.2. +# But only use the flag if it's available from the compiler. And the user +# didn't specify some other -march flag yet. +if( "${CMAKE_SYSTEM_PROCESSOR}" STREQUAL "x86_64" ) + set( X86_FLAG "-march=x86-64-v2" ) + check_cxx_compiler_flag( "${X86_FLAG}" TRACCC_SSE42_SUPPORTED ) + if( TRACCC_SSE42_SUPPORTED AND + ( NOT "${CMAKE_CXX_FLAGS}" MATCHES "-march=" ) ) + traccc_add_flag( CMAKE_CXX_FLAGS "${X86_FLAG}" ) + endif() + unset( X86_FLAG ) +endif() diff --git a/Traccc/cmake/traccc-compiler-options-cuda.cmake b/Traccc/cmake/traccc-compiler-options-cuda.cmake new file mode 100644 index 00000000000..69ded835c2b --- /dev/null +++ b/Traccc/cmake/traccc-compiler-options-cuda.cmake @@ -0,0 +1,62 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2021-2025 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# FindCUDAToolkit needs at least CMake 3.17, and C++17 support +# (set in the project's main CMakeLists.txt file) needs CMake 3.18. +cmake_minimum_required( VERSION 3.18 ) + +# Include the helper function(s). +include( traccc-functions ) + +# Figure out the properties of CUDA being used. +find_package( CUDAToolkit REQUIRED ) + +# Turn on the correct setting for the __cplusplus macro with MSVC. +if( "${CMAKE_CXX_COMPILER_ID}" MATCHES "MSVC" ) + traccc_add_flag( CMAKE_CUDA_FLAGS "-Xcompiler /Zc:__cplusplus" ) +endif() + +if( "${CMAKE_CUDA_COMPILER_ID}" MATCHES "NVIDIA" ) + traccc_add_flag( CMAKE_CUDA_FLAGS "-Wall" ) + traccc_add_flag( CMAKE_CUDA_FLAGS "-Wextra" ) + traccc_add_flag( CMAKE_CUDA_FLAGS "-Wconversion" ) +endif() + +# Allow to use functions in device code that are constexpr, even if they are +# not marked with __device__. +traccc_add_flag( CMAKE_CUDA_FLAGS "--expt-relaxed-constexpr --use_fast_math" ) + +# Make CUDA generate debug symbols for the device code as well in a debug +# build. +traccc_add_flag( CMAKE_CUDA_FLAGS_DEBUG "-G --keep" ) + +# Work around a bug in CUDA 12.8. Enabling the embedding of C++ source code in +# generated PTX code causes a ptxas error. A solution was promised for +# CUDA 13.1, but this has not yet surfaced. +# +# TODO: Add an upper bound to this statement when a fix in CUDA is presented. +if(CUDAToolkit_VERSION VERSION_GREATER_EQUAL "12.8") + message( + STATUS + "Disabling C++ source in PTX in order to work around a bug in CUDA 12.8:" + ) +else() + traccc_add_flag( CMAKE_CUDA_FLAGS_DEBUG "-src-in-ptx" ) +endif() + +# Ensure that line information is embedded in debugging builds so that +# profilers have access to line data. +traccc_add_flag( CMAKE_CUDA_FLAGS_RELWITHDEBINFO "-lineinfo" ) + +# Fail on warnings, if asked for that behaviour. +if( TRACCC_FAIL_ON_WARNINGS ) + if( ( "${CUDAToolkit_VERSION}" VERSION_GREATER_EQUAL "10.2" ) AND + ( "${CMAKE_CUDA_COMPILER_ID}" MATCHES "NVIDIA" ) ) + traccc_add_flag( CMAKE_CUDA_FLAGS "-Werror all-warnings" ) + elseif( "${CMAKE_CUDA_COMPILER_ID}" MATCHES "Clang" ) + traccc_add_flag( CMAKE_CUDA_FLAGS "-Werror" ) + endif() +endif() diff --git a/Traccc/cmake/traccc-compiler-options-sycl.cmake b/Traccc/cmake/traccc-compiler-options-sycl.cmake new file mode 100644 index 00000000000..91b60755525 --- /dev/null +++ b/Traccc/cmake/traccc-compiler-options-sycl.cmake @@ -0,0 +1,45 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2021-2023 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# Include the helper function(s). +include( traccc-functions ) + +# Only tweak the flags for the Intel compiler. +if( NOT ( ( "${CMAKE_SYCL_COMPILER_ID}" STREQUAL "IntelLLVM" ) OR + ( "${CMAKE_SYCL_COMPILER_ID}" MATCHES "Clang" ) ) ) + return() +endif() + +# Basic flags for all build modes. +foreach( mode RELEASE RELWITHDEBINFO MINSIZEREL DEBUG ) + traccc_add_flag( CMAKE_SYCL_FLAGS_${mode} "-Wall" ) + traccc_add_flag( CMAKE_SYCL_FLAGS_${mode} "-Wextra" ) + traccc_add_flag( CMAKE_SYCL_FLAGS_${mode} "-Wno-unknown-cuda-version" ) + traccc_add_flag( CMAKE_SYCL_FLAGS_${mode} "-Wshadow" ) + traccc_add_flag( CMAKE_SYCL_FLAGS_${mode} "-Wunused-local-typedefs" ) + traccc_add_flag( CMAKE_SYCL_FLAGS_${mode} "-Wconversion" ) +endforeach() + +if( NOT WIN32 ) + foreach( mode RELEASE RELWITHDEBINFO MINSIZEREL DEBUG ) + traccc_add_flag( CMAKE_SYCL_FLAGS_${mode} "-pedantic" ) + endforeach() +endif() + +# Fail on warnings, if asked for that behaviour. +if( TRACCC_FAIL_ON_WARNINGS ) + foreach( mode RELEASE RELWITHDEBINFO MINSIZEREL DEBUG ) + traccc_add_flag( CMAKE_SYCL_FLAGS_${mode} "-Werror" ) + endforeach() +endif() + +# Avoid issues coming from MSVC<->DPC++ argument differences. +if( "${CMAKE_CXX_COMPILER_ID}" MATCHES "MSVC" ) + foreach( mode RELEASE RELWITHDEBINFO MINSIZEREL DEBUG ) + traccc_add_flag( CMAKE_SYCL_FLAGS_${mode} + "-Wno-unused-command-line-argument" ) + endforeach() +endif() diff --git a/Traccc/cmake/traccc-config.cmake.in b/Traccc/cmake/traccc-config.cmake.in new file mode 100644 index 00000000000..83ccef4047e --- /dev/null +++ b/Traccc/cmake/traccc-config.cmake.in @@ -0,0 +1,53 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2022-2024 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# Set up the helper functions/macros. +@PACKAGE_INIT@ + +# Remember the options that traccc was built with. +set( TRACCC_BUILD_CUDA @TRACCC_BUILD_CUDA@ ) +set( TRACCC_BUILD_HIP @TRACCC_BUILD_HIP@ ) +set( TRACCC_BUILD_SYCL @TRACCC_BUILD_SYCL@ ) +set( TRACCC_BUILD_ALPAKA @TRACCC_BUILD_ALPAKA@ ) +set( TRACCC_BUILD_EXAMPLES @TRACCC_BUILD_EXAMPLES@ ) +set( TRACCC_USE_ROOT @TRACCC_USE_ROOT@ ) +set( TRACCC_BUILD_IO @TRACCC_BUILD_IO@ ) + +# Set up some simple variables for using the package. +set( traccc_VERSION "@PROJECT_VERSION@" ) +set_and_check( traccc_INCLUDE_DIR "@PACKAGE_CMAKE_INSTALL_INCLUDEDIR@" ) +set_and_check( traccc_LIBRARY_DIR "@PACKAGE_CMAKE_INSTALL_LIBDIR@" ) +set_and_check( traccc_CMAKE_DIR "@PACKAGE_CMAKE_INSTALL_CMAKEDIR@" ) + +# Make the "traccc modules" visible to CMake. +list( APPEND CMAKE_MODULE_PATH "${CMAKE_CURRENT_LIST_DIR}" ) + +# Find all packages that traccc needs to function. +include( CMakeFindDependencyMacro ) +find_dependency( Thrust ) +if( TRACCC_BUILD_ALPAKA ) + find_dependency( alpaka ) +endif() +if ( TRACCC_BUILD_IO ) + find_dependency( Acts ) +endif() +find_dependency( vecmem ) +find_dependency( detray ) +if( TRACCC_BUILD_EXAMPLES ) + find_dependency( Boost COMPONENTS program_options ) + if( TRACCC_USE_ROOT ) + find_dependency( ROOT COMPONENTS Core RIO Tree Hist ) + endif() +endif() + +# Include the file listing all the imported targets and options. +include( "${traccc_CMAKE_DIR}/traccc-config-targets.cmake" ) + +# Print a standard information message about the package being found. +include( FindPackageHandleStandardArgs ) +find_package_handle_standard_args( traccc REQUIRED_VARS + CMAKE_CURRENT_LIST_FILE + VERSION_VAR traccc_VERSION ) diff --git a/Traccc/cmake/traccc-ctest.sh.in b/Traccc/cmake/traccc-ctest.sh.in new file mode 100755 index 00000000000..95e508554f4 --- /dev/null +++ b/Traccc/cmake/traccc-ctest.sh.in @@ -0,0 +1,17 @@ +#!@BASH_EXECUTABLE@ +# +# 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 + +# Propagate errors. +set -e +set -o pipefail + +# Run every command through the time command. Recording the time, memory, etc. +# used by each and every build command. +command @TIME_EXECUTABLE@ @TIME_VERBOSE_FLAG@ \ + -ao @CMAKE_CURRENT_BINARY_DIR@/traccc_build_performance.log \ + @CMAKE_CTEST_COMMAND@ $* diff --git a/Traccc/cmake/traccc-functions.cmake b/Traccc/cmake/traccc-functions.cmake new file mode 100644 index 00000000000..d8453d6faf1 --- /dev/null +++ b/Traccc/cmake/traccc-functions.cmake @@ -0,0 +1,148 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2021-2023 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# DISCOVERY_TIMEOUT in gtest_discover_tests(...) requires at least CMake 3.10. +cmake_minimum_required( VERSION 3.10 ) + +# Guard against multiple includes. +include_guard( GLOBAL ) + +# CMake include(s). +include( CMakeParseArguments ) +include( GoogleTest ) + +# Function for declaring the libraries of the project +# +# Usage: traccc_add_library( traccc_core core +# [TYPE SHARED/INTERFACE/STATIC] +# include/source1.hpp source2.cpp ) +# +function( traccc_add_library fullname basename ) + + # Parse the function's options. + cmake_parse_arguments( ARG "" "TYPE" "" ${ARGN} ) + + # Decide what sources to give to the library. + set( _sources ${ARG_UNPARSED_ARGUMENTS} ) + if( "${ARG_TYPE}" STREQUAL "INTERFACE" ) + set( _sources ) + endif() + + # Create the library. + add_library( ${fullname} ${ARG_TYPE} ${_sources} ) + + # Set up how clients should find its headers. + if( IS_DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/include/" ) + set( _depType PUBLIC ) + if( "${ARG_TYPE}" STREQUAL "INTERFACE" ) + set( _depType INTERFACE ) + endif() + target_include_directories( ${fullname} ${_depType} + $ + $ ) + unset( _depType ) + endif() + + # Make sure that the library is available as "traccc::${basename}" in every + # situation. + set_target_properties( ${fullname} PROPERTIES EXPORT_NAME ${basename} ) + add_library( traccc::${basename} ALIAS ${fullname} ) + + # Specify the (SO)VERSION of the library. + if( NOT "${ARG_TYPE}" STREQUAL "INTERFACE" ) + set_target_properties( ${fullname} PROPERTIES + VERSION ${PROJECT_VERSION} + SOVERSION ${PROJECT_VERSION_MAJOR} ) + endif() + + # Set up the installation of the library and its headers. + install( TARGETS ${fullname} + EXPORT traccc-exports + LIBRARY DESTINATION "${CMAKE_INSTALL_LIBDIR}" + ARCHIVE DESTINATION "${CMAKE_INSTALL_LIBDIR}" + RUNTIME DESTINATION "${CMAKE_INSTALL_BINDIR}" ) + if( IS_DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/include/" ) + install( DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/include/" + DESTINATION "${CMAKE_INSTALL_INCLUDEDIR}" ) + endif() + +endfunction( traccc_add_library ) + +# Function for declaring the (installable) executables of the project +# +# Usage: traccc_add_executable( super_exe super_source.cpp ) +# +function( traccc_add_executable name ) + + # Parse the function's options. + cmake_parse_arguments( ARG "" "" "LINK_LIBRARIES" ${ARGN} ) + + # Set up the executable. + set( exe_name "traccc_${name}" ) + add_executable( ${exe_name} ${ARG_UNPARSED_ARGUMENTS} ) + if( ARG_LINK_LIBRARIES ) + target_link_libraries( ${exe_name} PRIVATE ${ARG_LINK_LIBRARIES} ) + endif() + + # Make sure that the executable is available as "traccc::${name}" in every + # situation. + set_target_properties( ${exe_name} PROPERTIES EXPORT_NAME ${name} ) + add_executable( traccc::${name} ALIAS ${exe_name} ) + + # Set up the installation of the executable. + install( TARGETS ${exe_name} + EXPORT traccc-exports + RUNTIME DESTINATION "${CMAKE_INSTALL_BINDIR}" ) + +endfunction( traccc_add_executable ) + +# Helper function for setting up the traccc tests. +# +# Usage: traccc_add_test( core_containers source1.cpp source2.cpp +# LINK_LIBRARIES traccc::core ) +# +function( traccc_add_test name ) + + # Parse the function's options. + cmake_parse_arguments( ARG "" "" "LINK_LIBRARIES" ${ARGN} ) + + # Create the test executable. + set( test_exe_name "traccc_test_${name}" ) + add_executable( ${test_exe_name} ${ARG_UNPARSED_ARGUMENTS} ) + if( ARG_LINK_LIBRARIES ) + target_link_libraries( ${test_exe_name} PRIVATE ${ARG_LINK_LIBRARIES} ) + endif() + + # Discover all of the tests from the execuable, and set them up as individual + # CTest tests. All the while ensuring that they would find their data files. + gtest_discover_tests( ${test_exe_name} + PROPERTIES ENVIRONMENT + TRACCC_TEST_DATA_DIR=${PROJECT_SOURCE_DIR}/data + DISCOVERY_TIMEOUT 20 ) + +endfunction( traccc_add_test ) + +# Helper function for adding individual flags to "flag variables". +# +# Usage: traccc_add_flag( CMAKE_CXX_FLAGS "-Wall" ) +# +function( traccc_add_flag name value ) + + # Escape special characters in the value: + set( matchedValue "${value}" ) + foreach( c "*" "." "^" "$" "+" "?" ) + string( REPLACE "${c}" "\\${c}" matchedValue "${matchedValue}" ) + endforeach() + + # Check if the variable already has this value in it: + if( "${${name}}" MATCHES "${matchedValue}" ) + return() + endif() + + # If not, then let's add it now: + set( ${name} "${${name}} ${value}" PARENT_SCOPE ) + +endfunction( traccc_add_flag ) diff --git a/Traccc/cmake/traccc-packaging.cmake b/Traccc/cmake/traccc-packaging.cmake new file mode 100644 index 00000000000..01c5a3a251b --- /dev/null +++ b/Traccc/cmake/traccc-packaging.cmake @@ -0,0 +1,34 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2022 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# CMake include(s). +include( CPack ) + +# Export the configuration of the project. +include( CMakePackageConfigHelpers ) +set( CMAKE_INSTALL_CMAKEDIR + "${CMAKE_INSTALL_LIBDIR}/cmake/traccc-${PROJECT_VERSION}" ) +install( EXPORT traccc-exports + NAMESPACE "traccc::" + FILE "traccc-config-targets.cmake" + DESTINATION "${CMAKE_INSTALL_CMAKEDIR}" ) +configure_package_config_file( + "${CMAKE_CURRENT_SOURCE_DIR}/cmake/traccc-config.cmake.in" + "${CMAKE_CURRENT_BINARY_DIR}${CMAKE_FILES_DIRECTORY}/traccc-config.cmake" + INSTALL_DESTINATION "${CMAKE_INSTALL_CMAKEDIR}" + PATH_VARS CMAKE_INSTALL_INCLUDEDIR CMAKE_INSTALL_LIBDIR + CMAKE_INSTALL_CMAKEDIR + NO_CHECK_REQUIRED_COMPONENTS_MACRO ) +write_basic_package_version_file( + "${CMAKE_CURRENT_BINARY_DIR}${CMAKE_FILES_DIRECTORY}/traccc-config-version.cmake" + COMPATIBILITY "AnyNewerVersion" ) +install( FILES + "${CMAKE_CURRENT_BINARY_DIR}${CMAKE_FILES_DIRECTORY}/traccc-config.cmake" + "${CMAKE_CURRENT_BINARY_DIR}${CMAKE_FILES_DIRECTORY}/traccc-config-version.cmake" + DESTINATION "${CMAKE_INSTALL_CMAKEDIR}" ) + +# Clean up. +unset( CMAKE_INSTALL_CMAKEDIR ) diff --git a/Traccc/codegen/kernel_specialization/gen_kernel_specialization.py b/Traccc/codegen/kernel_specialization/gen_kernel_specialization.py new file mode 100644 index 00000000000..fe382e06c4b --- /dev/null +++ b/Traccc/codegen/kernel_specialization/gen_kernel_specialization.py @@ -0,0 +1,73 @@ +import string +import argparse +import pathlib + +if __name__ == "__main__": + parser = argparse.ArgumentParser() + + parser.add_argument( + "template", + type=pathlib.Path, + help="the template path", + ) + + parser.add_argument( + "-o", + "--output", + type=pathlib.Path, + help="the output path to write to", + required=True, + ) + + parser.add_argument( + "--detector", + type=str, + help="the name of the detector", + ) + + parser.add_argument( + "--bfield", + type=str, + help="the name of the bfield", + ) + + parser.add_argument( + "--model", + type=str, + help="the name of the programming mode", + default="cpu", + ) + + args = parser.parse_args() + + with open(args.template, "r") as f: + src = string.Template(f.read()) + + subs = {} + + subs["SOURCE_DIR"] = args.template.parent + + det = getattr(args, "detector", None) + if det is not None: + subs["DETECTOR_NAME"] = det + + bfield = getattr(args, "bfield", None) + if bfield is not None: + # HACK: Perform some transformations to make the types line up... + # This could be resolved in the C++ file itself in the future. + bfield_name = bfield + + if args.model != "cpu" and bfield != "const": + bfield_name = args.model + "::" + bfield_name + + bfield_name += "_bfield_backend_t" + + if bfield != "inhom_texture": + bfield_name += "" + + subs["BFIELD_NAME"] = bfield_name + + result = src.substitute(subs) + + with open(args.output, "w") as f: + f.write(result) diff --git a/Traccc/core/CMakeLists.txt b/Traccc/core/CMakeLists.txt new file mode 100644 index 00000000000..f57af9e5a80 --- /dev/null +++ b/Traccc/core/CMakeLists.txt @@ -0,0 +1,163 @@ +# TRACCC library, part of the ACTS project (R&D line) +# +# (c) 2021-2026 CERN for the benefit of the ACTS project +# +# Mozilla Public License Version 2.0 + +# Project include(s). +include( traccc-compiler-options-cpp ) + +# Set up the "build" of the traccc::core library. +traccc_add_library( traccc_core core TYPE SHARED + # Common definitions. + "include/traccc/definitions/track_parametrization.hpp" + "include/traccc/definitions/math.hpp" + "include/traccc/definitions/primitives.hpp" + "include/traccc/definitions/common.hpp" + "include/traccc/definitions/qualifiers.hpp" + # Event data model. + "include/traccc/edm/details/container_base.hpp" + "include/traccc/edm/details/container_element.hpp" + "include/traccc/edm/details/device_container.hpp" + "include/traccc/edm/details/host_container.hpp" + "include/traccc/edm/measurement_collection.hpp" + "include/traccc/edm/impl/measurement_collection.ipp" + "include/traccc/edm/measurement_helpers.hpp" + "include/traccc/edm/impl/measurement_helpers.ipp" + "include/traccc/edm/particle.hpp" + "include/traccc/edm/track_parameters.hpp" + "include/traccc/edm/container.hpp" + "include/traccc/edm/silicon_cell_collection.hpp" + "include/traccc/edm/impl/silicon_cell_collection.ipp" + "include/traccc/edm/silicon_cluster_collection.hpp" + "include/traccc/edm/spacepoint_collection.hpp" + "include/traccc/edm/impl/spacepoint_collection.ipp" + "include/traccc/edm/spacepoint_helpers.hpp" + "include/traccc/edm/impl/spacepoint_helpers.ipp" + "include/traccc/edm/seed_collection.hpp" + "include/traccc/edm/impl/seed_collection.ipp" + "include/traccc/edm/track_state_collection.hpp" + "include/traccc/edm/impl/track_state_collection.ipp" + "include/traccc/edm/track_state_helpers.hpp" + "include/traccc/edm/impl/track_state_helpers.ipp" + "include/traccc/edm/track_fit_outcome.hpp" + "include/traccc/edm/track_constituent_link.hpp" + "include/traccc/edm/track_collection.hpp" + "include/traccc/edm/impl/track_collection.ipp" + "include/traccc/edm/track_container.hpp" + # Magnetic field description. + "include/traccc/bfield/magnetic_field_types.hpp" + "include/traccc/bfield/magnetic_field.hpp" + "include/traccc/bfield/impl/magnetic_field.ipp" + "include/traccc/bfield/construct_const_bfield.hpp" + "include/traccc/bfield/impl/construct_const_bfield.ipp" + "src/bfield/construct_const_bfield.cpp" + # Geometry description. + "include/traccc/geometry/detector.hpp" + "include/traccc/geometry/module_map.hpp" + "include/traccc/geometry/geometry.hpp" + "include/traccc/geometry/pixel_data.hpp" + "include/traccc/geometry/detector_design_description.hpp" + "include/traccc/geometry/detector_conditions_description.hpp" + # Utilities. + "include/traccc/utils/algorithm.hpp" + "include/traccc/utils/detray_conversion.hpp" + "include/traccc/utils/impl/detray_conversion.ipp" + "include/traccc/utils/type_traits.hpp" + "include/traccc/utils/memory_resource.hpp" + "include/traccc/utils/seed_generator.hpp" + "include/traccc/utils/subspace.hpp" + "include/traccc/utils/logging.hpp" + "include/traccc/utils/prob.hpp" + "src/utils/logging.cpp" + # Clusterization algorithmic code. + "include/traccc/clusterization/details/sparse_ccl.hpp" + "include/traccc/clusterization/impl/sparse_ccl.ipp" + "include/traccc/clusterization/sparse_ccl_algorithm.hpp" + "src/clusterization/sparse_ccl_algorithm.cpp" + "include/traccc/clusterization/details/measurement_creation.hpp" + "include/traccc/clusterization/impl/measurement_creation.ipp" + "include/traccc/clusterization/measurement_creation_algorithm.hpp" + "src/clusterization/measurement_creation_algorithm.cpp" + "include/traccc/clusterization/measurement_sorting_algorithm.hpp" + "src/clusterization/measurement_sorting_algorithm.cpp" + "include/traccc/clusterization/clusterization_algorithm.hpp" + "src/clusterization/clusterization_algorithm.cpp" + # Finding algorithmic code + "include/traccc/finding/candidate_link.hpp" + "include/traccc/finding/finding_config.hpp" + "include/traccc/finding/actors/ckf_aborter.hpp" + "include/traccc/finding/details/combinatorial_kalman_filter_types.hpp" + "include/traccc/finding/details/combinatorial_kalman_filter.hpp" + "include/traccc/finding/combinatorial_kalman_filter_algorithm.hpp" + "src/finding/combinatorial_kalman_filter_algorithm.cpp" + # Fitting algorithmic code + "include/traccc/fitting/kalman_filter/gain_matrix_updater.hpp" + "include/traccc/fitting/kalman_filter/kalman_actor.hpp" + "include/traccc/fitting/kalman_filter/kalman_fitter.hpp" + "include/traccc/fitting/kalman_filter/kalman_step_aborter.hpp" + "include/traccc/fitting/kalman_filter/measurement_selector.hpp" + "include/traccc/fitting/kalman_filter/statistics_updater.hpp" + "include/traccc/fitting/kalman_filter/two_filters_smoother.hpp" + "include/traccc/fitting/details/kalman_fitting_types.hpp" + "include/traccc/fitting/details/kalman_fitting.hpp" + "include/traccc/fitting/kalman_fitting_algorithm.hpp" + "include/traccc/fitting/triplet_fit/triplet_fitter.hpp" + "include/traccc/fitting/triplet_fitting_algorithm.hpp" + "src/fitting/kalman_fitting_algorithm.cpp" + "src/fitting/triplet_fitting_algorithm.cpp" + # Seed finding algorithmic code. + "include/traccc/seeding/detail/lin_circle.hpp" + "include/traccc/seeding/detail/doublet.hpp" + "include/traccc/seeding/detail/triplet.hpp" + "include/traccc/seeding/detail/singlet.hpp" + "include/traccc/seeding/detail/seeding_config.hpp" + "include/traccc/seeding/detail/spacepoint_grid.hpp" + "include/traccc/seeding/seed_selecting_helper.hpp" + "src/seeding/seed_filtering.hpp" + "src/seeding/seed_filtering.cpp" + "include/traccc/seeding/seeding_algorithm.hpp" + "src/seeding/seeding_algorithm.cpp" + "include/traccc/seeding/track_params_estimation_helper.hpp" + "include/traccc/seeding/doublet_finding_helper.hpp" + "include/traccc/seeding/spacepoint_binning_helper.hpp" + "include/traccc/seeding/track_params_estimation.hpp" + "src/seeding/track_params_estimation.cpp" + "include/traccc/seeding/triplet_finding_helper.hpp" + "src/seeding/doublet_finding.hpp" + "src/seeding/triplet_finding.hpp" + "include/traccc/seeding/detail/seed_finding.hpp" + "src/seeding/seed_finding.cpp" + "include/traccc/seeding/detail/spacepoint_binning.hpp" + "src/seeding/spacepoint_binning.cpp" + "include/traccc/seeding/detail/spacepoint_formation.hpp" + "include/traccc/seeding/impl/spacepoint_formation.ipp" + "src/seeding/silicon_pixel_spacepoint_formation.hpp" + "include/traccc/seeding/silicon_pixel_spacepoint_formation_algorithm.hpp" + "src/seeding/silicon_pixel_spacepoint_formation_algorithm.cpp" + #gbts seed finding config + "include/traccc/gbts_seeding/gbts_seeding_config.hpp" + "src/gbts_seeding/gbts_seeding_config.cpp" + # Ambiguity resolution + "include/traccc/ambiguity_resolution/ambiguity_resolution_config.hpp" + "include/traccc/ambiguity_resolution/greedy_ambiguity_resolution_algorithm.hpp" + "src/ambiguity_resolution/greedy_ambiguity_resolution_algorithm.cpp" + "include/traccc/ambiguity_resolution/legacy/greedy_ambiguity_resolution_algorithm.hpp" + "src/ambiguity_resolution/legacy/greedy_ambiguity_resolution_algorithm.cpp") +target_link_libraries( traccc_core + PUBLIC vecmem::core covfie::core detray::core_array detray::detectors + Acts::Core ) + +string(REPLACE ";" ", " TRACCC_DETECTOR_TYPES "${TRACCC_SUPPORTED_DETECTORS}") +message(STATUS "Building with detector types: ${TRACCC_DETECTOR_TYPES}") +configure_file( + "${CMAKE_CURRENT_SOURCE_DIR}/include/traccc/geometry/detector_type_list.hpp.in" + "${CMAKE_CURRENT_BINARY_DIR}${CMAKE_FILES_DIRECTORY}/include/traccc/geometry/detector_type_list.hpp" + @ONLY) + +target_include_directories(traccc_core + PUBLIC $ ) + +install( + DIRECTORY "${CMAKE_CURRENT_BINARY_DIR}${CMAKE_FILES_DIRECTORY}/include/" + DESTINATION "${CMAKE_INSTALL_INCLUDEDIR}" ) diff --git a/Traccc/core/include/traccc/ambiguity_resolution/ambiguity_resolution_config.hpp b/Traccc/core/include/traccc/ambiguity_resolution/ambiguity_resolution_config.hpp new file mode 100644 index 00000000000..5f4fc514f43 --- /dev/null +++ b/Traccc/core/include/traccc/ambiguity_resolution/ambiguity_resolution_config.hpp @@ -0,0 +1,29 @@ +/** 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 + */ + +#pragma once + +// System include(s). +#include +#include + +namespace traccc { + +/// Configuration struct for ambiguity resolution +struct ambiguity_resolution_config { + + /// Minimum number of measurement to form a track. + unsigned int min_meas_per_track = 3; + + /// Max iteration to remove the bad tracks + unsigned int max_iterations = std::numeric_limits::max(); + + /// Max shared measurements to break the iteration + unsigned int max_shared_meas = 1; +}; + +} // namespace traccc diff --git a/Traccc/core/include/traccc/ambiguity_resolution/greedy_ambiguity_resolution_algorithm.hpp b/Traccc/core/include/traccc/ambiguity_resolution/greedy_ambiguity_resolution_algorithm.hpp new file mode 100644 index 00000000000..a665d84dd46 --- /dev/null +++ b/Traccc/core/include/traccc/ambiguity_resolution/greedy_ambiguity_resolution_algorithm.hpp @@ -0,0 +1,62 @@ +/** 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 + */ + +#pragma once + +// Project include(s). +#include "traccc/ambiguity_resolution/ambiguity_resolution_config.hpp" +#include "traccc/definitions/primitives.hpp" +#include "traccc/edm/track_container.hpp" +#include "traccc/utils/algorithm.hpp" +#include "traccc/utils/messaging.hpp" + +// VecMem include(s). +#include + +// System include(s). +#include + +namespace traccc::host { + +/// Evicts tracks that seem to be duplicates or fakes. This algorithm takes a +/// greedy approach in the sense that it will remove the track which looks "most +/// duplicate/fake" +class greedy_ambiguity_resolution_algorithm + : public algorithm::host( + const edm::track_container::const_view&)>, + public messaging { + + public: + using config_type = ambiguity_resolution_config; + + /// Constructor for the greedy ambiguity resolution algorithm + /// + /// @param cfg Configuration object + greedy_ambiguity_resolution_algorithm( + const config_type& cfg, vecmem::memory_resource& mr, + std::unique_ptr logger = getDummyLogger().clone()) + : messaging(std::move(logger)), m_config{cfg}, m_mr{mr} {} + + /// Run the algorithm + /// + /// @param tracks the container of found patterns + /// @return the container without ambiguous tracks + output_type operator()( + const edm::track_container::const_view& tracks) + const override; + + /// Get configuration + config_type& get_config() { return m_config; } + + private: + /// Algorithm configuration + config_type m_config; + /// The memory resource to use + std::reference_wrapper m_mr; +}; + +} // namespace traccc::host diff --git a/Traccc/core/include/traccc/ambiguity_resolution/legacy/greedy_ambiguity_resolution_algorithm.hpp b/Traccc/core/include/traccc/ambiguity_resolution/legacy/greedy_ambiguity_resolution_algorithm.hpp new file mode 100644 index 00000000000..84be4c60ef0 --- /dev/null +++ b/Traccc/core/include/traccc/ambiguity_resolution/legacy/greedy_ambiguity_resolution_algorithm.hpp @@ -0,0 +1,139 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2024-2025 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// System include +#include +#include +#include +#include +#include +#include +#include +#include +#include + +// VecMem include(s). +#include + +// Project include(s). +#include "traccc/definitions/qualifiers.hpp" +#include "traccc/edm/track_container.hpp" +#include "traccc/utils/algorithm.hpp" +#include "traccc/utils/messaging.hpp" + +// Greedy ambiguity resolution adapted from ACTS code + +namespace traccc::legacy { + +/// Evicts tracks that seem to be duplicates or fakes. This algorithm takes a +/// greedy approach in the sense that it will remove the track which looks "most +/// duplicate/fake" first and continues the same process with the rest. That +/// process continues until the final state conditions are met. +/// +/// The implementation works as follows: +/// 1) Calculate shared hits per track. +/// 2) If the maximum shared hits criteria is met, we are done. +/// This is the configurable amount of shared hits we are ok with +/// in our experiment. +/// 3) Else, remove the track with the highest relative shared hits (i.e. +/// shared hits / hits). +/// 4) Back to square 1. +class greedy_ambiguity_resolution_algorithm + : public algorithm::host( + const edm::track_container::host&)>, + public messaging { + + public: + struct config_t { + + config_t() {} + + /// Maximum amount of shared hits per track. One (1) means "no shared + /// hit allowed". + std::uint32_t maximum_shared_hits = 1; + + /// Maximum number of iterations. + std::uint32_t maximum_iterations = 1000000; + + /// Minimum number of measurement to form a track. + std::size_t n_measurements_min = 3; + }; + + struct state_t { + std::size_t number_of_tracks{}; + + /// For this whole comment section, track_index refers to the index of a + /// track in the initial input container. + /// + /// There is no (track_id) in this algorithm, only (track_index). + + /// Associates each track_index with the track's p-value + std::vector track_pval; + + /// Associates each track_index to the track's (measurement_id)s list + std::vector> measurements_per_track; + + /// Associates each measurement_id to a set of (track_index)es sharing + /// it + std::unordered_map> + tracks_per_measurement; + + /// Associates each track_index to its number of shared measurements + /// (among other tracks) + std::vector shared_measurements_per_track; + + /// Keeps the selected tracks indexes that have not (yet) been removed + /// by the algorithm + std::set> selected_tracks; + }; + + /// Constructor for the greedy ambiguity resolution algorithm + /// + /// @param cfg Configuration object + // greedy_ambiguity_resolution_algorithm(const config_type& cfg) : + // _config(cfg) {} + greedy_ambiguity_resolution_algorithm( + const config_t cfg, vecmem::memory_resource& mr, + std::unique_ptr logger = getDummyLogger().clone()) + : messaging(std::move(logger)), _config{cfg}, m_mr{mr} {} + + /// Run the algorithm + /// + /// @param track_states the container of the fitted track parameters + /// @return the container without ambiguous tracks + output_type operator()(const edm::track_container::host& + tracks) const override; + + /// Get configuration + config_t& get_config() { return _config; } + + private: + /// Computes the initial state for the input data. This function accumulates + /// information that will later be used to accelerate the ambiguity + /// resolution. + /// + /// @param tracks The input track container + /// @param state An empty state object which is expected to be default + /// constructed. + void compute_initial_state( + const edm::track_container::host& tracks, + state_t& state) const; + + /// Updates the state iteratively by evicting one track after the other + /// until the final state conditions are met. + /// + /// @param state A state object that was previously filled by the + /// initialization. + void resolve(state_t& state) const; + + config_t _config; + std::reference_wrapper m_mr; +}; + +} // namespace traccc::legacy diff --git a/Traccc/core/include/traccc/bfield/construct_const_bfield.hpp b/Traccc/core/include/traccc/bfield/construct_const_bfield.hpp new file mode 100644 index 00000000000..c718a98a910 --- /dev/null +++ b/Traccc/core/include/traccc/bfield/construct_const_bfield.hpp @@ -0,0 +1,36 @@ +/** 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 + */ + +#pragma once + +// Local include(s). +#include "traccc/bfield/magnetic_field.hpp" +#include "traccc/definitions/primitives.hpp" + +namespace traccc { + +/// Construct a constant magnetic field object +/// +/// @tparam scalar_t The scalar type to construct the field with +/// +/// @param x The X component of the constant field +/// @param y The Y component of the constant field +/// @param z The Z component of the constant field +/// +template +magnetic_field construct_const_bfield(scalar_t x, scalar_t y, scalar_t z); + +/// Construct a constant magnetic field object +/// +/// @param v The 3-vector describing the constant field +/// +magnetic_field construct_const_bfield(const vector3& v); + +} // namespace traccc + +// Include the implementation. +#include "traccc/bfield/impl/construct_const_bfield.ipp" diff --git a/Traccc/core/include/traccc/bfield/impl/construct_const_bfield.ipp b/Traccc/core/include/traccc/bfield/impl/construct_const_bfield.ipp new file mode 100644 index 00000000000..ad51143d4ab --- /dev/null +++ b/Traccc/core/include/traccc/bfield/impl/construct_const_bfield.ipp @@ -0,0 +1,27 @@ +/** 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 + */ + +#pragma once + +// Local include(s). +#include "traccc/bfield/magnetic_field_types.hpp" + +// Covfie include(s). +#include + +namespace traccc { + +template +magnetic_field construct_const_bfield(scalar_t x, scalar_t y, scalar_t z) { + + return magnetic_field{::covfie::field>{ + ::covfie::make_parameter_pack( + typename const_bfield_backend_t::configuration_t{x, y, + z})}}; +} + +} // namespace traccc diff --git a/Traccc/core/include/traccc/bfield/impl/magnetic_field.ipp b/Traccc/core/include/traccc/bfield/impl/magnetic_field.ipp new file mode 100644 index 00000000000..10e37af54ce --- /dev/null +++ b/Traccc/core/include/traccc/bfield/impl/magnetic_field.ipp @@ -0,0 +1,38 @@ +/** 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 + */ + +#pragma once + +namespace traccc { + +template +magnetic_field::magnetic_field(covfie::field&& obj) + : m_field(std::move(obj)) {} + +template +void magnetic_field::set(covfie::field&& obj) { + m_field = std::move(obj); +} + +template +bool magnetic_field::is() const { + return (m_field.type() == typeid(covfie::field)); +} + +template +typename covfie::field::view_t magnetic_field::as_view() + const { + return typename covfie::field::view_t{ + std::any_cast&>(m_field)}; +} + +template +const covfie::field& magnetic_field::as_field() const { + return std::any_cast&>(m_field); +} + +} // namespace traccc diff --git a/Traccc/core/include/traccc/bfield/magnetic_field.hpp b/Traccc/core/include/traccc/bfield/magnetic_field.hpp new file mode 100644 index 00000000000..015ff6929a3 --- /dev/null +++ b/Traccc/core/include/traccc/bfield/magnetic_field.hpp @@ -0,0 +1,120 @@ +/** 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 + */ + +#pragma once + +// Covfie include(s). +#include +#include + +// System include(s). +#include +#include + +namespace traccc { + +/// Typeless, owning, host-only magnetic field object +class magnetic_field { + + public: + /// Default constructor + magnetic_field() = default; + + /// Constructor from a specific b-field object + /// + /// @tparam bfield_backend_t The backend type of the b-field object + /// @param obj The b-field object to construct from + /// + template + explicit magnetic_field(covfie::field&& obj); + + /// Set a specific b-field object + /// + /// @tparam bfield_backend_t The backend type of the b-field object + /// @param obj The b-field object to set + /// + template + void set(covfie::field&& obj); + + /// Check if the b-field is of a certain type + /// + /// @tparam bfield_backend_t The covfie backend type to check + /// @return @c true if the b-field is of the specified type, + /// @c false otherwise + /// + template + bool is() const; + + /// Get a b-field view object as a specific type + /// + /// @tparam bfield_backend_t The covfie backend type to use + /// @return The b-field view object of the specified type + /// + template + typename covfie::field::view_t as_view() const; + + /// Get the b-field object as a specific type + /// + /// @tparam bfield_backend_t The covfie backend type to use + /// @return The b-field object cast to the specified type + /// + template + const covfie::field& as_field() const; + + /// @brief Return type information about the contained magnetic field. + const std::type_info& type() const { return m_field.type(); } + + private: + /// The actualy covfie b-field object + std::any m_field; + +}; // class magnetic_field + +/// @brief Helper function for `bfield_visitor` +template +auto magnetic_field_visitor_helper(const magnetic_field& bfield, + callable_t&& callable, + std::tuple*) + requires(covfie::concepts::field_backend && + (covfie::concepts::field_backend && ...)) +{ + if (bfield.is()) { + return callable(bfield.as_view()); + } else { + if constexpr (sizeof...(bfield_ts) > 0) { + return magnetic_field_visitor_helper( + bfield, std::forward(callable), + static_cast*>(nullptr)); + } else { + std::stringstream exception_message; + + exception_message + << "Invalid B-field type (" << bfield.type().name() + << ") received, but this type is not supported" << std::endl; + + throw std::invalid_argument(exception_message.str()); + } + } +} + +/// @brief Visitor for polymorphic magnetic field types +/// +/// This function takes a list of supported magnetic field types and checks +/// if the provided field is one of them. If it is, it will call the provided +/// callable on a view of it and otherwise it will throw an exception. +template +auto magnetic_field_visitor(const magnetic_field& bfield, + callable_t&& callable) { + return magnetic_field_visitor_helper(bfield, + std::forward(callable), + static_cast(nullptr)); +} + +} // namespace traccc + +// Include the implementation. +#include "traccc/bfield/impl/magnetic_field.ipp" diff --git a/Traccc/core/include/traccc/bfield/magnetic_field_types.hpp b/Traccc/core/include/traccc/bfield/magnetic_field_types.hpp new file mode 100644 index 00000000000..efdc7570226 --- /dev/null +++ b/Traccc/core/include/traccc/bfield/magnetic_field_types.hpp @@ -0,0 +1,58 @@ +/** 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 + */ + +#pragma once + +// Covfie include(s). +#include +#include +#include +#include +#include +#include +#include + +namespace traccc { + +/// Constant magnetic field backend type +template +using const_bfield_backend_t = + ::covfie::backend::constant<::covfie::vector::vector_d, + ::covfie::vector::vector_d>; +// Test that the type is a valid backend for a field +static_assert(covfie::concepts::field_backend>, + "const_bfield_backend_t is not a valid field backend type"); + +namespace host { + +/// Inhomogeneous magnetic field used for IO +template +using inhom_io_bfield_backend_t = + covfie::backend::affine, + covfie::backend::array>>>>; +// Test that the type is a valid backend for a field +static_assert(covfie::concepts::field_backend>, + "inhom_io_bfield_backend_t is not a valid field backend type"); + +/// Inhomogeneous magnetic field backend type +template +using inhom_bfield_backend_t = covfie::backend::affine< + covfie::backend::linear, + covfie::backend::array>>>>>; +// Test that the type is a valid backend for a field +static_assert(covfie::concepts::field_backend>, + "inhom_bfield_backend_t is not a valid field backend type"); + +/// @brief The standard list of host bfield types to support +template +using bfield_type_list = std::tuple, + inhom_bfield_backend_t>; + +} // namespace host +} // namespace traccc diff --git a/Traccc/core/include/traccc/clusterization/clustering_config.hpp b/Traccc/core/include/traccc/clusterization/clustering_config.hpp new file mode 100644 index 00000000000..e05e10254dd --- /dev/null +++ b/Traccc/core/include/traccc/clusterization/clustering_config.hpp @@ -0,0 +1,114 @@ +/** + * 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 + +#include "traccc/definitions/qualifiers.hpp" + +namespace traccc { +/** + * @brief The different strategies that can be used to compute the diameter + * of a cluster. + */ +enum class clustering_diameter_strategy { + /** + * @brief Compute the diameter as the size of the cluster in the channel0 + * dimension. + */ + CHANNEL0, + /** + * @brief Compute the diameter as the size of the cluster in the channel1 + * dimension. + */ + CHANNEL1, + /** + * @brief Compute the diameter as the maximum of the sizes of the cluster + * in the channel0 and channel1 dimensions. + */ + MAXIMUM, + /** + * @brief Compute the diameter as the diagonal (i.e. the hypotenuse) of + * the triangle with sides bounding both the channel0 and channel1 + * dimensions. + */ + DIAGONAL +}; + +/** + * @brief Configuration type for massively parallel clustering algorithms. + */ +struct clustering_config { + /** + * @brief The desired number of threads per partition. + * + * This directly correlates to the block size on most algorithms, so don't + * set this too low (which will reduce occupancy due to available thread + * slots) or too high (which may not be supported on a device). + */ + unsigned int threads_per_partition{256}; + + /** + * @brief The maximum number of cells per thread. + * + * This sets the maximum thread coarsening factor for the CCA algorithm. + * Increasing this value increases shared memory usage and may decrease + * occupancy. If this is too low, scratch space will need to be used which + * may slow the algorithm down. + */ + unsigned int max_cells_per_thread{16}; + + /** + * @brief The desired number of cells per thread. + * + * This sets the desired thread coarsening factor for the CCA algorithm. + * Decreasing this may decrease occupancy. Increasing this increases the + * probability that scratch space will need to be used. + */ + unsigned int target_cells_per_thread{8}; + + /** + * @brief The upscaling factor for the scratch space. + * + * The scratch space will be large enough to support partitions this number + * of times larger than the maximum partition size determined by + * `threads_per_partition` and `max_cells_per_thread` + */ + unsigned int backup_size_multiplier{256}; + + /** + * @brief The strategy for computing the cluster width. + * + * See the enum definition for more details. + */ + clustering_diameter_strategy diameter_strategy = + clustering_diameter_strategy::CHANNEL1; + + /** + * @brief The maximum number of cells per partition. + */ + TRACCC_HOST_DEVICE constexpr unsigned int max_partition_size() const { + return threads_per_partition * max_cells_per_thread; + } + + /** + * @brief The target number of cells per partition. + */ + TRACCC_HOST_DEVICE constexpr unsigned int target_partition_size() const { + return threads_per_partition * target_cells_per_thread; + } + + /** + * @brief The total size of the scratch space, in number of cells. + */ + TRACCC_HOST_DEVICE constexpr unsigned int backup_size() const { + return max_partition_size() * backup_size_multiplier; + } +}; +} // namespace traccc diff --git a/Traccc/core/include/traccc/clusterization/clusterization_algorithm.hpp b/Traccc/core/include/traccc/clusterization/clusterization_algorithm.hpp new file mode 100644 index 00000000000..cf7c78c0f16 --- /dev/null +++ b/Traccc/core/include/traccc/clusterization/clusterization_algorithm.hpp @@ -0,0 +1,81 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2026 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Library include(s). +#include "traccc/clusterization/measurement_creation_algorithm.hpp" +#include "traccc/clusterization/sparse_ccl_algorithm.hpp" +#include "traccc/edm/measurement_collection.hpp" +#include "traccc/edm/silicon_cell_collection.hpp" +#include "traccc/geometry/detector_conditions_description.hpp" +#include "traccc/geometry/detector_design_description.hpp" +#include "traccc/utils/algorithm.hpp" +#include "traccc/utils/messaging.hpp" + +// VecMem include(s). +#include + +// System include(s). +#include +#include + +namespace traccc::host { + +/// Clusterization algorithm, creating measurements from cells +/// +/// This algorithm creates local/2D measurements separately for each detector +/// module from the cells of the modules. +/// +class clusterization_algorithm + : public algorithm, + public messaging { + + public: + using config_type = std::monostate; + + /// Clusterization algorithm constructor + /// + /// @param mr The memory resource to use for the result objects + /// + clusterization_algorithm( + vecmem::memory_resource& mr, + std::unique_ptr logger = getDummyLogger().clone()); + + /// Construct measurements for each detector module + /// + /// @param cells_view The cells for every detector module in the event + /// @param dmd_view The detector segmentation description + /// @param dcd_view The detector conditins description + /// @return The measurements reconstructed for every detector module + /// + output_type operator()( + const edm::silicon_cell_collection::const_view& cells_view, + const detector_design_description::const_view& dmd_view, + const detector_conditions_description::const_view& dcd_view) + const override; + + private: + /// @name Sub-algorithms used by this algorithm + /// @{ + + /// Per-module cluster creation algorithm + sparse_ccl_algorithm m_cc; + + /// Per-module measurement creation algorithm + measurement_creation_algorithm m_mc; + + /// @} + + /// Reference to the host-accessible memory resource + std::reference_wrapper m_mr; +}; // class clusterization_algorithm + +} // namespace traccc::host diff --git a/Traccc/core/include/traccc/clusterization/details/measurement_creation.hpp b/Traccc/core/include/traccc/clusterization/details/measurement_creation.hpp new file mode 100644 index 00000000000..a3376d444d2 --- /dev/null +++ b/Traccc/core/include/traccc/clusterization/details/measurement_creation.hpp @@ -0,0 +1,72 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2025 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Project include(s). +#include "traccc/definitions/primitives.hpp" +#include "traccc/definitions/qualifiers.hpp" +#include "traccc/edm/measurement_collection.hpp" +#include "traccc/edm/silicon_cell_collection.hpp" +#include "traccc/edm/silicon_cluster_collection.hpp" +#include "traccc/geometry/detector_conditions_description.hpp" +#include "traccc/geometry/detector_design_description.hpp" + +namespace traccc::details { + +/// Get the local position of a cell on a module +/// +/// @param cell The cell to get the position of +/// @param det_descr The (silicon) detector description +/// @return The local position of the cell (upper bound) and optionality the +/// lower bound +/// +template +TRACCC_HOST_DEVICE inline vector2 position_from_cell( + const edm::silicon_cell& cell, + const traccc::detector_design_description_interface& module_dd); + +/// Function used for calculating the properties of the cluster during +/// measurement creation +/// +/// @param[in] cluster The silicon cluster to calculate the properties of +/// @param[in] cells All silicon cells in the event +/// @param[in] det_descr The detector segmentation description +/// @param[in] det_cond The detector conditionas data +/// @param[out] mean The mean position of the cluster/measurement +/// @param[out] var The variation on the mean position of the +/// cluster/measurement +/// @param[out] totalWeight The total weight of the cluster/measurement +/// +template +TRACCC_HOST_DEVICE inline void calc_cluster_properties( + const edm::silicon_cluster& cluster, + const edm::silicon_cell_collection::const_device& cells, + const traccc::detector_design_description_interface& module_dd, + point2& mean, point2& var, scalar& totalWeight); + +/// Function used for calculating the properties of the cluster during +/// measurement creation +/// +/// @param[out] measurement Measurement object to be filled +/// @param[in] cluster The silicon cluster to turn into a measurement +/// @param[in] index The index of the cluster/measurement in the collection +/// @param[in] cells All silicon cells in the event +/// @param[in] det_descr Detector segmentation description +/// @param[in] det_cond Detector conditions description +/// +template +TRACCC_HOST_DEVICE inline void fill_measurement( + edm::measurement& measurement, const edm::silicon_cluster& cluster, + unsigned int index, const edm::silicon_cell_collection::const_device& cells, + const detector_design_description::const_device& det_descr, + const detector_conditions_description::const_device& det_cond); + +} // namespace traccc::details + +// Include the implementation. +#include "traccc/clusterization/impl/measurement_creation.ipp" diff --git a/Traccc/core/include/traccc/clusterization/details/sparse_ccl.hpp b/Traccc/core/include/traccc/clusterization/details/sparse_ccl.hpp new file mode 100644 index 00000000000..0053906dd78 --- /dev/null +++ b/Traccc/core/include/traccc/clusterization/details/sparse_ccl.hpp @@ -0,0 +1,82 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2024 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Library include(s). +#include "traccc/definitions/qualifiers.hpp" +#include "traccc/edm/silicon_cell_collection.hpp" +#include "traccc/geometry/detector_conditions_description.hpp" + +// VecMem include(s). +#include + +namespace traccc::details { + +/// Implemementation of SparseCCL, following +/// [DOI: 10.1109/DASIP48288.2019.9049184] +/// +/// Requires cells to be sorted in column major + +/// Find root of the tree for entry @param e +/// +/// @param labels an equivalance table +/// +/// @return the root of @param e +/// +TRACCC_HOST_DEVICE inline unsigned int find_root( + const vecmem::device_vector& labels, unsigned int e); + +/// Create a union of two entries @param e1 and @param e2 +/// +/// @param labels an equivalance table +/// +/// @return the rleast common ancestor of the entries +/// +TRACCC_HOST_DEVICE inline unsigned int make_union( + vecmem::device_vector& labels, unsigned int e1, + unsigned int e2); + +/// Helper method to find adjacent cells +/// +/// @param a the first cell +/// @param b the second cell +/// +/// @return boolan to indicate 8-cell connectivity +/// +template +TRACCC_HOST_DEVICE inline bool is_adjacent(const edm::silicon_cell& a, + const edm::silicon_cell& b); + +/// Helper method to find define distance, +/// does not need abs, as channels are sorted in +/// column major +/// +/// @param a the first cell +/// @param b the second cell +/// +/// @return boolan to indicate !8-cell connectivity +/// +template +TRACCC_HOST_DEVICE inline bool is_far_enough(const edm::silicon_cell& a, + const edm::silicon_cell& b); + +/// Sparce CCL algorithm +/// +/// @param cells is the cell collection +/// @param labels is the vector of the output indices (to which cluster a cell +/// belongs to) +/// @return number of clusters +/// +TRACCC_HOST_DEVICE inline unsigned int sparse_ccl( + const edm::silicon_cell_collection::const_device& cells, + vecmem::device_vector& labels); + +} // namespace traccc::details + +// Include the implementation. +#include "traccc/clusterization/impl/sparse_ccl.ipp" diff --git a/Traccc/core/include/traccc/clusterization/impl/measurement_creation.ipp b/Traccc/core/include/traccc/clusterization/impl/measurement_creation.ipp new file mode 100644 index 00000000000..3fc30b11b09 --- /dev/null +++ b/Traccc/core/include/traccc/clusterization/impl/measurement_creation.ipp @@ -0,0 +1,182 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2026 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Library include(s). +#include "traccc/utils/detray_conversion.hpp" + +// System include(s). +#include + +namespace traccc::details { + +template +TRACCC_HOST_DEVICE inline vector2 position_from_cell( + const edm::silicon_cell& cell, + const traccc::detector_design_description_interface& module_dd) { + // Calculate / construct the local cell position. + vector2 cell_lower_position = { + (module_dd.bin_edges_x()).at(cell.channel0()), + (module_dd.bin_edges_y()).at(cell.channel1())}; + + vector2 cell_upper_position = { + (module_dd.bin_edges_x()).at(cell.channel0() + 1), + (module_dd.bin_edges_y()).at(cell.channel1() + 1)}; + + vector2 cell_middle_position = { + scalar{0.5f} * (cell_upper_position[0] + cell_lower_position[0]), + scalar{0.5f} * (cell_upper_position[1] + cell_lower_position[1])}; + + return cell_middle_position; +} + +template +TRACCC_HOST_DEVICE inline void calc_cluster_properties( + const edm::silicon_cluster& cluster, + const edm::silicon_cell_collection::const_device& cells, + const traccc::detector_design_description_interface& module_dd, + point2& mean, point2& var, scalar& totalWeight) { + + point2 offset{0.f, 0.f}, width{0.f, 0.f}; + bool first_processed = false; + + unsigned int min_channel0 = std::numeric_limits::max(); + unsigned int max_channel0 = std::numeric_limits::lowest(); + unsigned int min_channel1 = std::numeric_limits::max(); + unsigned int max_channel1 = std::numeric_limits::lowest(); + + // Loop over the cell indices of the cluster. + for (const unsigned int cell_idx : cluster.cell_indices()) { + + // The cell object. + const edm::silicon_cell cell = cells.at(cell_idx); + + // Translate the cell readout value into a weight. + const scalar weight = cell.activation(); + + // Update all output properties with this cell. + totalWeight += weight; + scalar weight_factor = weight / totalWeight; + + point2 cell_position = position_from_cell(cell, module_dd); + + min_channel0 = std::min(min_channel0, cell.channel0()); + min_channel1 = std::min(min_channel1, cell.channel1()); + max_channel0 = std::max(max_channel0, cell.channel0()); + max_channel1 = std::max(max_channel1, cell.channel1()); + + if (!first_processed) { + offset = cell_position; + first_processed = true; + } + + cell_position = cell_position - offset; + + const point2 diff_old = cell_position - mean; + mean = mean + diff_old * weight_factor; + const point2 diff_new = cell_position - mean; + + var[0] = (1.f - weight_factor) * var[0] + + weight_factor * (diff_old[0] * diff_new[0]); + var[1] = (1.f - weight_factor) * var[1] + + weight_factor * (diff_old[1] * diff_new[1]); + } + + // cluster width in the number of cells + unsigned int delta0 = (max_channel0 - min_channel0) + 1; + unsigned int delta1 = (max_channel1 - min_channel1) + 1; + + vector2 cluster_lower_position = { + (module_dd.bin_edges_x()).at(min_channel0), + (module_dd.bin_edges_y()).at(min_channel1)}; + + vector2 cluster_upper_position = { + (module_dd.bin_edges_x()).at(max_channel0 + 1), + (module_dd.bin_edges_y()).at(max_channel1 + 1)}; + + width[0] = cluster_upper_position[0] - cluster_lower_position[0]; + width[1] = cluster_upper_position[1] - cluster_lower_position[1]; + + point2 pitch = {width[0] / static_cast(delta0), + width[1] / static_cast(delta1)}; + + var = var + point2{pitch[0] * pitch[0] / static_cast(12.), + pitch[1] * pitch[1] / static_cast(12.)}; + + mean = mean + offset; +} + +template +TRACCC_HOST_DEVICE inline void fill_measurement( + edm::measurement& measurement, const edm::silicon_cluster& cluster, + const unsigned int index, + const edm::silicon_cell_collection::const_device& cells, + const detector_design_description::const_device& det_descr, + const detector_conditions_description::const_device& det_cond) { + + // To calculate the mean and variance with high numerical stability + // we use a weighted variant of Welford's algorithm. This is a + // single-pass online algorithm that works well for large numbers + // of samples, as well as samples with very high values. + // + // To learn more about this algorithm please refer to: + // [1] https://doi.org/10.1080/00401706.1962.10490022 + // [2] The Art of Computer Programming, Donald E. Knuth, second + // edition, chapter 4.2.2. + + // Security checks. + assert(cluster.cell_indices().empty() == false); + assert([&]() { + const unsigned int module_idx = + cells.module_index().at(cluster.cell_indices().front()); + for (const unsigned int cell_idx : cluster.cell_indices()) { + if (cells.module_index().at(cell_idx) != module_idx) { + return false; + } + } + return true; + }() == true); + + // The index of the module the cluster is on. + const unsigned int module_idx = + cells.module_index().at(cluster.cell_indices().front()); + const auto module_cd = det_cond.at(module_idx); + const unsigned int design_idx = module_cd.module_to_design_id(); + const auto module_dd = det_descr.at(design_idx); + + // Calculate the cluster properties + scalar totalWeight = 0.f; + point2 mean{0.f, 0.f}, var{0.f, 0.f}; + calc_cluster_properties(cluster, cells, module_dd, mean, var, totalWeight); + assert(totalWeight > 0.f); + + // Fill the measurement object. + measurement.surface_link() = module_cd.geometry_id(); + + // apply lorentz shift to the cell position + measurement.local_position() = utils::to_float_array( + mean + module_cd.measurement_translation()); + + // plus pitch^2 / 12 + measurement.local_variance() = utils::to_float_array(var); + + // For the ambiguity resolution algorithm, give a unique measurement ID + measurement.identifier() = index; + measurement.cluster_index() = index; + + // Set the measurement dimensionality. + measurement.dimensions() = module_dd.dimensions(); + + // Set the measurement's subspace. + measurement.set_subspace(module_dd.subspace()); + + // Save the index of the cluster that produced this measurement + measurement.cluster_index() = static_cast(index); +} + +} // namespace traccc::details diff --git a/Traccc/core/include/traccc/clusterization/impl/sparse_ccl.ipp b/Traccc/core/include/traccc/clusterization/impl/sparse_ccl.ipp new file mode 100644 index 00000000000..de5cff0ff7a --- /dev/null +++ b/Traccc/core/include/traccc/clusterization/impl/sparse_ccl.ipp @@ -0,0 +1,99 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2024 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// System include(s). +#include + +namespace traccc::details { + +TRACCC_HOST_DEVICE inline unsigned int find_root( + const vecmem::device_vector& labels, unsigned int e) { + + unsigned int r = e; + assert(r < labels.size()); + while (labels[r] != r) { + r = labels[r]; + assert(r < labels.size()); + } + return r; +} + +TRACCC_HOST_DEVICE inline unsigned int make_union( + vecmem::device_vector& labels, unsigned int e1, + unsigned int e2) { + + unsigned int e; + if (e1 < e2) { + e = e1; + assert(e2 < labels.size()); + labels[e2] = e; + } else { + e = e2; + assert(e1 < labels.size()); + labels[e1] = e; + } + return e; +} + +template +TRACCC_HOST_DEVICE inline bool is_adjacent(const edm::silicon_cell& a, + const edm::silicon_cell& b) { + + return (a.channel0() - b.channel0()) * (a.channel0() - b.channel0()) <= 1 && + (a.channel1() - b.channel1()) * (a.channel1() - b.channel1()) <= 1 && + a.module_index() == b.module_index(); +} + +template +TRACCC_HOST_DEVICE inline bool is_far_enough(const edm::silicon_cell& a, + const edm::silicon_cell& b) { + + assert((a.channel1() >= b.channel1()) || + (a.module_index() != b.module_index())); + return (a.channel1() > (b.channel1() + 1)) || + (a.module_index() != b.module_index()); +} + +TRACCC_HOST_DEVICE inline unsigned int sparse_ccl( + const edm::silicon_cell_collection::const_device& cells, + vecmem::device_vector& labels) { + + unsigned int nlabels = 0; + + // The number of cells. + const unsigned int n_cells = cells.size(); + + // first scan: pixel association + unsigned int start_j = 0; + for (unsigned int i = 0; i < n_cells; ++i) { + + labels[i] = i; + unsigned int ai = i; + for (unsigned int j = start_j; j < i; ++j) { + + if (is_adjacent(cells.at(i), cells.at(j))) { + ai = make_union(labels, ai, find_root(labels, j)); + } else if (is_far_enough(cells.at(i), cells.at(j))) { + ++start_j; + } + } + } + + // second scan: transitive closure + for (unsigned int i = 0; i < n_cells; ++i) { + if (labels[i] == i) { + labels[i] = nlabels++; + } else { + labels[i] = labels[labels[i]]; + } + } + return nlabels; +} + +} // namespace traccc::details diff --git a/Traccc/core/include/traccc/clusterization/measurement_creation_algorithm.hpp b/Traccc/core/include/traccc/clusterization/measurement_creation_algorithm.hpp new file mode 100644 index 00000000000..296e127d0e7 --- /dev/null +++ b/Traccc/core/include/traccc/clusterization/measurement_creation_algorithm.hpp @@ -0,0 +1,70 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2026 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Library include(s). +#include "traccc/edm/measurement_collection.hpp" +#include "traccc/edm/silicon_cell_collection.hpp" +#include "traccc/edm/silicon_cluster_collection.hpp" +#include "traccc/geometry/detector_conditions_description.hpp" +#include "traccc/geometry/detector_design_description.hpp" +#include "traccc/utils/algorithm.hpp" +#include "traccc/utils/messaging.hpp" + +// VecMem include(s). +#include + +// System include(s). +#include + +namespace traccc::host { + +/// Measurement creation out of clusters +/// +/// This algorithm can create measurements for a single detector module +/// for all of the clusters that were identified in that one detector +/// module. +/// +class measurement_creation_algorithm + : public algorithm, + public messaging { + + public: + /// Measurement_creation algorithm constructor + /// + /// @param mr The memory resource to use in the algorithm + /// + measurement_creation_algorithm( + vecmem::memory_resource &mr, + std::unique_ptr logger = getDummyLogger().clone()); + + /// Callable operator for the connected component, based on one single + /// module + /// + /// @param cells_view Cells that were clusterized + /// @param clusters_view Clusters to turn into measurements + /// @param dd_view The detector description + /// @return The reconstructed measurement collection + /// + output_type operator()( + const edm::silicon_cell_collection::const_view &cells_view, + const edm::silicon_cluster_collection::const_view &clusters_view, + const detector_design_description::const_view &dmd_view, + const detector_conditions_description::const_view &dcd_view) + const override; + + private: + /// The memory resource used by the algorithm + std::reference_wrapper m_mr; +}; // class measurement_creation_algorithm + +} // namespace traccc::host diff --git a/Traccc/core/include/traccc/clusterization/measurement_sorting_algorithm.hpp b/Traccc/core/include/traccc/clusterization/measurement_sorting_algorithm.hpp new file mode 100644 index 00000000000..4c667b3bbaf --- /dev/null +++ b/Traccc/core/include/traccc/clusterization/measurement_sorting_algorithm.hpp @@ -0,0 +1,61 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2024-2026 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Library include(s). +#include "traccc/edm/measurement_collection.hpp" +#include "traccc/utils/algorithm.hpp" +#include "traccc/utils/messaging.hpp" + +// VecMem include(s). +#include + +// System include(s). +#include +#include + +namespace traccc::host { + +/// Algorithm sorting the reconstructed measurements in their container +/// +/// The track finding algorithm expects measurements belonging to a single +/// detector module to be consecutive in memory. But certain clusterization / +/// measurement creation algorithms may not produce the measurements in such +/// an ordered state. In such cases this algorithm can be used to sort the +/// measurements "correctly" in place. +/// +class measurement_sorting_algorithm + : public algorithm, + public messaging { + + public: + /// Constructor + /// + /// @param mr The memory resource to use for the algorithm + /// @param logger The logger to use for the algorithm + /// + measurement_sorting_algorithm( + vecmem::memory_resource& mr, + std::unique_ptr logger = getDummyLogger().clone()); + + /// Callable operator performing the sorting on a container + /// + /// @param measurements The measurements to sort + /// + [[nodiscard]] output_type operator()( + const edm::measurement_collection::const_view& measurements) + const override; + + private: + /// The memory resource to use + std::reference_wrapper m_mr; + +}; // class measurement_sorting_algorithm + +} // namespace traccc::host diff --git a/Traccc/core/include/traccc/clusterization/sparse_ccl_algorithm.hpp b/Traccc/core/include/traccc/clusterization/sparse_ccl_algorithm.hpp new file mode 100644 index 00000000000..820e6ea6111 --- /dev/null +++ b/Traccc/core/include/traccc/clusterization/sparse_ccl_algorithm.hpp @@ -0,0 +1,67 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2024 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Library include(s). +#include "traccc/edm/silicon_cell_collection.hpp" +#include "traccc/edm/silicon_cluster_collection.hpp" +#include "traccc/geometry/detector_conditions_description.hpp" +#include "traccc/utils/algorithm.hpp" +#include "traccc/utils/messaging.hpp" + +// VecMem include(s). +#include + +// System include(s). +#include + +namespace traccc::host { + +/// Pixel cell clusterization based on a SparseCCL algorithm +/// +/// The implementation is based on the paper: +/// https://doi.org/10.1109/DASIP48288.2019.9049184 +/// +class sparse_ccl_algorithm + : public algorithm, + public messaging { + + public: + /// Constructor for component_connection + /// + /// @param mr is the memory resource + /// + sparse_ccl_algorithm( + vecmem::memory_resource& mr, + std::unique_ptr logger = getDummyLogger().clone()); + + /// @name Operator(s) to use in host code + /// @{ + + /// Callable operator for the connected component labelling + /// + /// @param cells_view Collection of input cells sorted by module + /// @param det_cond_view Collection of detector conditions + /// + /// @return a cluster container + /// + output_type operator()( + const edm::silicon_cell_collection::const_view& cells_view, + const detector_conditions_description::const_view& det_cond_view) + const override; + + /// @} + + private: + /// The memory resource used by the algorithm + std::reference_wrapper m_mr; +}; // class sparse_ccl_algorithm + +} // namespace traccc::host diff --git a/Traccc/core/include/traccc/definitions/common.hpp b/Traccc/core/include/traccc/definitions/common.hpp new file mode 100644 index 00000000000..4e3b512eb34 --- /dev/null +++ b/Traccc/core/include/traccc/definitions/common.hpp @@ -0,0 +1,27 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2022 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Project include(s). +#include "traccc/definitions/primitives.hpp" + +// Detray include(s). +#include + +namespace traccc { + +template +using unit = detray::unit; + +template +using constant = detray::constant; + +// epsilon for float variables +constexpr scalar float_epsilon = 1e-5f; + +} // namespace traccc diff --git a/Traccc/core/include/traccc/definitions/hints.hpp b/Traccc/core/include/traccc/definitions/hints.hpp new file mode 100644 index 00000000000..e707a6c689b --- /dev/null +++ b/Traccc/core/include/traccc/definitions/hints.hpp @@ -0,0 +1,24 @@ +/** + * 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 + +#if not defined __has_builtin +#define TRACCC_ASSUME(...) +#elif __has_builtin(__builtin_assume) +#define TRACCC_ASSUME(...) __builtin_assume(__VA_ARGS__) +#else +#define TRACCC_ASSUME(...) +#endif + +#if defined(__CUDACC__) || defined(__HIP__) || defined(__OPENMP) || \ + defined(__SYCL__) || defined(__clang__) +#define TRACCC_PRAGMA_UNROLL _Pragma("unroll") +#else +#define TRACCC_PRAGMA_UNROLL +#endif diff --git a/Traccc/core/include/traccc/definitions/math.hpp b/Traccc/core/include/traccc/definitions/math.hpp new file mode 100644 index 00000000000..c4290bf5405 --- /dev/null +++ b/Traccc/core/include/traccc/definitions/math.hpp @@ -0,0 +1,79 @@ +/** 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 "traccc/definitions/qualifiers.hpp" + +// SYCL include(s). +#if defined(CL_SYCL_LANGUAGE_VERSION) || defined(SYCL_LANGUAGE_VERSION) +#include +#endif + +// System include(s). +#include + +namespace traccc::math { + +#if defined(CL_SYCL_LANGUAGE_VERSION) || defined(SYCL_LANGUAGE_VERSION) +using ::sycl::abs; +using ::sycl::acos; +using ::sycl::asin; +using ::sycl::atan; +using ::sycl::atan2; +using ::sycl::cos; +using ::sycl::exp; +using ::sycl::fabs; +using ::sycl::floor; +using ::sycl::fmod; +using ::sycl::log; +using ::sycl::max; +using ::sycl::min; +using ::sycl::pow; +using ::sycl::sin; +using ::sycl::sqrt; +using ::sycl::tan; +#else +using std::abs; +using std::acos; +using std::asin; +using std::atan; +using std::atan2; +using std::cos; +using std::exp; +using std::fabs; +using std::floor; +using std::fmod; +using std::log; +using std::max; +using std::min; +using std::pow; +using std::sin; +using std::sqrt; +using std::tan; +#endif + +/** + * @brief Perform IEEE-754 division, even if fast math is enabled. + * + * @returns x divided by y + */ +template +TRACCC_HOST_DEVICE inline __attribute__((always_inline)) T div_ieee754(T x, + T y) { + static_assert(std::is_same_v || std::is_same_v); +#ifdef __CUDA_ARCH__ + if constexpr (std::is_same_v) { + return __ddiv_rn(x, y); + } else { + return __fdiv_rn(x, y); + } +#else + return x / y; +#endif +} +} // namespace traccc::math diff --git a/Traccc/core/include/traccc/definitions/primitives.hpp b/Traccc/core/include/traccc/definitions/primitives.hpp new file mode 100644 index 00000000000..6c1e97e8508 --- /dev/null +++ b/Traccc/core/include/traccc/definitions/primitives.hpp @@ -0,0 +1,46 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2026 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Detray include(s) +#include + +// System include(s) +#include + +namespace traccc { + +using measurement_id_type = unsigned int; +using particle_id = std::uint64_t; +using geometry_id = std::uint64_t; +using channel_id = unsigned int; + +// Default algebra type +using default_algebra = detray::array; + +using detray::algebra::array::operator*; +using detray::algebra::array::operator-; +using detray::algebra::array::operator+; + +using scalar = detray::dscalar; +using point2 = detray::dpoint2D; +using vector2 = point2; +using variance2 = point2; +using point3 = detray::dpoint3D; +using vector3 = detray::dvector3D; +using variance3 = point3; +using transform3 = detray::dtransform3D; + +namespace getter = detray::getter; +namespace vector = detray::vector; +namespace matrix = detray::matrix; + +// Pull in the print operator definitions for the algebra types +using detray::algebra::operator<<; + +} // namespace traccc diff --git a/Traccc/core/include/traccc/definitions/qualifiers.hpp b/Traccc/core/include/traccc/definitions/qualifiers.hpp new file mode 100644 index 00000000000..0fc8493511f --- /dev/null +++ b/Traccc/core/include/traccc/definitions/qualifiers.hpp @@ -0,0 +1,33 @@ +/** + * TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +#if defined(__CUDACC__) || defined(__HIP__) +#define TRACCC_DEVICE __device__ +#else +#define TRACCC_DEVICE +#endif + +#if defined(__CUDACC__) || defined(__HIP__) +#define TRACCC_HOST __host__ +#else +#define TRACCC_HOST +#endif + +#if defined(__CUDACC__) || defined(__HIP__) +#define TRACCC_HOST_DEVICE __host__ __device__ +#else +#define TRACCC_HOST_DEVICE +#endif + +#if defined(__CUDACC__) || defined(__HIP__) +#define TRACCC_ALIGN(x) __align__(x) +#else +#define TRACCC_ALIGN(x) alignas(x) +#endif diff --git a/Traccc/core/include/traccc/definitions/track_parametrization.hpp b/Traccc/core/include/traccc/definitions/track_parametrization.hpp new file mode 100644 index 00000000000..31f648c6922 --- /dev/null +++ b/Traccc/core/include/traccc/definitions/track_parametrization.hpp @@ -0,0 +1,18 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2024 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Detray include(s) +#include + +namespace traccc { + +using enum detray::bound_indices; +using enum detray::free_indices; + +} // namespace traccc diff --git a/Traccc/core/include/traccc/edm/container.hpp b/Traccc/core/include/traccc/edm/container.hpp new file mode 100644 index 00000000000..fb230cabc89 --- /dev/null +++ b/Traccc/core/include/traccc/edm/container.hpp @@ -0,0 +1,185 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2022 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Project include(s). +#include "traccc/definitions/qualifiers.hpp" +#include "traccc/edm/details/device_container.hpp" +#include "traccc/edm/details/host_container.hpp" +#include "traccc/utils/type_traits.hpp" + +// VecMem include(s). +#include +#include +#include +#include +#include +#include +#include + +// System include(s). +#include + +namespace traccc { + +/// @name Types used to send data back and forth between host and device code +/// @{ + +/// Structure holding (some of the) data about the container in host code +template +struct container_data { + using header_vector = vecmem::data::vector_view; + using item_vector = vecmem::data::jagged_vector_data; + header_vector headers; + item_vector items; +}; + +/// Structure holding (all of the) data about the container in host code +template +struct container_buffer { + using header_vector = vecmem::data::vector_buffer; + using item_vector = vecmem::data::jagged_vector_buffer; + header_vector headers; + item_vector items; +}; + +/// Structure used to send the data about the container to device code +/// +/// This is the type that can be passed to device code as-is. But since in +/// host code one needs to manage the data describing a +/// @c traccc::container either using @c traccc::container_data or +/// @c traccc::container_buffer, it needs to have constructors from +/// both of those types. +/// +/// In fact it needs to be created from one of those types, as such an +/// object can only function if an instance of one of those types exists +/// alongside it as well. +/// +template +struct container_view { + + /// Type for the header vector (view) + using header_vector = vecmem::data::vector_view; + /// Type for the item vector (view) + using item_vector = vecmem::data::jagged_vector_view; + + /// Constructor from a @c container_data object + template < + typename other_header_t, typename other_item_t, + std::enable_if_t::value, + bool> = true, + std::enable_if_t::value, + bool> = true> + container_view(const container_data& data) + : headers(data.headers), items(data.items) {} + + /// Constructor from a @c container_buffer object + template < + typename other_header_t, typename other_item_t, + std::enable_if_t::value, + bool> = true, + std::enable_if_t::value, + bool> = true> + container_view(const container_buffer& buffer) + : headers(buffer.headers), items(buffer.items) {} + + /// Constructor from a non-const view + template < + typename other_header_t, typename other_item_t, + std::enable_if_t::value, + bool> = true, + std::enable_if_t::value, + bool> = true> + container_view(const container_view& parent) + : headers(parent.headers), items(parent.items) {} + + /// View of the data describing the headers + header_vector headers; + + /// View of the data describing the items + item_vector items; +}; + +/// Helper function for making a "simple" object out of the container +/// (non-const) +template +inline container_data get_data( + host_container& cc, + vecmem::memory_resource* resource = nullptr) { + return {{vecmem::get_data(cc.get_headers())}, + {vecmem::get_data(cc.get_items(), resource)}}; +} + +/// Helper function for making a "simple" object out of the container (const) +template +inline container_data get_data( + const host_container& cc, + vecmem::memory_resource* resource = nullptr) { + return {{vecmem::get_data(cc.get_headers())}, + {vecmem::get_data(cc.get_items(), resource)}}; +} + +/// Type trait defining all "collection types" for an EDM class +template +struct collection_types { + + /// @c item_t must not be a constant type + static_assert(std::is_const::value == false, + "The template parameter must not be a constant type"); + + /// Host collection for @c item_t + using host = vecmem::vector; + /// Non-const device collection for @c item_t + using device = vecmem::device_vector; + /// Constant device collection for @c item_t + using const_device = vecmem::device_vector; + + /// Non-constant view of an @c item_t collection + using view = vecmem::data::vector_view; + /// Constant view of an @c item_t collection + using const_view = vecmem::data::vector_view; + + /// Buffer for an @c item_t collection + using buffer = vecmem::data::vector_buffer; + +}; // struct collection_types + +/// Type trait defining all "container types" for an EDM class pair +template +struct container_types { + + /// @c header_t must not be a constant type + static_assert(std::is_const::value == false, + "The header type must not be constant"); + /// @c item_t must not be a constant type + static_assert(std::is_const::value == false, + "The item type must not be constant"); + + /// Host container for @c header_t and @c item_t + using host = host_container; + /// Non-const device container for @c header_t and @c item_t + using device = device_container; + /// Constant device container for @c header_t and @c item_t + using const_device = device_container; + + /// Non-constant view of an @c header_t / @c item_t container + using view = container_view; + /// Constant view of an @c header_t / @c item_t container + using const_view = container_view; + + /// Non-constant data for an @c header_t / @c item_t container + using data = container_data; + /// Constant data for an @c header_t / @c item_t container + using const_data = container_data; + + /// Buffer for an @c header_t / @c item_t container + using buffer = container_buffer; + +}; // struct container_types + +} // namespace traccc diff --git a/Traccc/core/include/traccc/edm/details/container_base.hpp b/Traccc/core/include/traccc/edm/details/container_base.hpp new file mode 100644 index 00000000000..50181e29892 --- /dev/null +++ b/Traccc/core/include/traccc/edm/details/container_base.hpp @@ -0,0 +1,215 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2022 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Project include(s). +#include "traccc/definitions/qualifiers.hpp" +#include "traccc/edm/details/container_element.hpp" +#include "traccc/utils/pair.hpp" + +// VecMem include(s). +#include + +// System include(s). +#include +#include +#include + +namespace traccc { + +/// Container base class for describing objects in a given event +/// +/// This is the generic container of the code, holding all relevant +/// information about objcts in a given event. +/// +/// It can be instantiated with different vector types, to be able to use +/// the same container type in both host and device code. +/// +/// It also can be instantiated with different edm types represented by +/// header and item type. +/// +template class vector_t, + template class jagged_vector_t, + template class pair_t = traccc::pair> +class container_base { + public: + /// @name Type definitions + /// @{ + + /// Header type + using header_type = header_t; + + /// Item type + using item_type = item_t; + + /// Vector type used by the container + template + using vector_type = vector_t; + + /// Jagged vector type used by the container + template + using jagged_vector_type = jagged_vector_t; + + /// The header vector type + using header_vector = vector_type; + + /// The item vector type + using item_vector = jagged_vector_type; + + /** + * @brief The size type of this container, which is the type by which + * its elements are indexed. + */ + using size_type = typename header_vector::size_type; + + /// The element link type + using link_type = pair_t; + + /// @} + + /** + * @brief The type name of the element view which is returned by various + * methods in this class. + */ + using element_view = container_element; + + /** + * @brief The type name of the constant element view which is returned + * by various methods in this class. + */ + using const_element_view = + container_element; + + /** + * We need to assert that the header vector and the outer layer of the + * jagged vector have the same size type, so they can be indexed using + * the same type. + */ + static_assert( + std::is_convertible::value, + "Size type for container header and item vectors must be the same."); + + /// Default copy-constructor + container_base(const container_base&) = default; + /// Default move-constructor + container_base(container_base&&) = default; + + /** + * @brief (Copy) Constructor from a header and item vector + * + * To enforce the invariant that both vectors must be the same size, we + * check this in the constructor. This is also checked in release + * builds. + */ + TRACCC_HOST + container_base(const header_vector& hv, const item_vector& iv) + : m_headers(hv), m_items(iv) { + if (m_headers.size() != m_items.size()) { + throw std::logic_error("Header and item length not equal."); + } + } + + /** + * @brief (Move) Constructor from a header and item vector + * + * To enforce the invariant that both vectors must be the same size, we + * check this in the constructor. This is also checked in release + * builds. + */ + TRACCC_HOST + container_base(header_vector&& hv, item_vector&& iv) + : m_headers(hv), m_items(iv) { + if (m_headers.size() != m_items.size()) { + throw std::logic_error("Header and item length not equal."); + } + } + + /** + * @brief Constructor from a pair of "view type" objects + */ + template + requires(std::constructible_from && + std::constructible_from) + TRACCC_HOST_DEVICE container_base(const header_vector_tp& hv, + const item_vector_tp& iv) + : m_headers(hv), m_items(iv) { + + assert(m_headers.size() == m_items.size()); + } + + /** + * @brief Constructor with vector size and memory resource . + */ + template + TRACCC_HOST explicit container_base(size_type size, + vecmem::memory_resource* mr) + : m_headers(size, mr), m_items(size, mr) {} + + /** + * @brief Constructor with memory resource . + */ + + TRACCC_HOST explicit container_base(vecmem::memory_resource* mr) + : m_headers(mr), m_items(mr) {} + + /** + * @brief Default Constructor + */ + container_base() = default; + + /// Default (copy) assignment operator + container_base& operator=(const container_base&) = default; + /// Default (move) assignment operator + container_base& operator=(container_base&&) = default; + + /** + * @brief Accessor method for the internal header vector. + */ + TRACCC_HOST_DEVICE + const header_vector& get_headers() const { return m_headers; } + + /** + * @brief Non-const accessor method for the internal header vector. + * + * @warning Do not use this function! It is dangerous, and risks breaking + * invariants! + */ + TRACCC_HOST_DEVICE + header_vector& get_headers() { return m_headers; } + + /** + * @brief Accessor method for the internal item vector-of-vectors. + */ + TRACCC_HOST_DEVICE + const item_vector& get_items() const { return m_items; } + + /** + * @brief Non-const accessor method for the internal item vector-of-vectors. + * + * @warning Do not use this function! It is dangerous, and risks breaking + * invariants! + */ + TRACCC_HOST_DEVICE + item_vector& get_items() { return m_items; } + + protected: + /// Headers information related to the objects in the event + header_vector m_headers; + + /// All objects in the event + item_vector m_items; + +}; // class container_base + +} // namespace traccc diff --git a/Traccc/core/include/traccc/edm/details/container_element.hpp b/Traccc/core/include/traccc/edm/details/container_element.hpp new file mode 100644 index 00000000000..5bc8ce8cfb3 --- /dev/null +++ b/Traccc/core/include/traccc/edm/details/container_element.hpp @@ -0,0 +1,53 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2022 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Project include(s). +#include "traccc/definitions/qualifiers.hpp" + +namespace traccc { + +/** + * @brief View class for an element in a header-vector container. + * + * In order to enforce certain invariants on the @c container header-vector + * type, we access elements (which, of course, are product types of a header + * and a vector) through this wrapper class. This class provides + * low-overhead access to the struct-of-arrays container class to emulate an + * array-of-structs architecture. + * + * @tparam header_reference_t The type of reference of the header object. + * @tparam vector_reference_t The fully qualified vector reference type. + */ +template +struct container_element { + + /// Header reference type + using header_reference = header_reference_t; + /// Vector reference type + using vector_reference = vector_reference_t; + + /** + * @brief Construct a new container element view. + * + * This constructor is extremely trivial, as it simply takes a reference + * to a header, a reference to a vector, and saves them in this object's + * internal state. + * + * @param[in] h The header object reference. + * @param[in] v The vector object reference. + */ + TRACCC_HOST_DEVICE + container_element(header_reference h, vector_reference v) + : header(h), items(v) {} + + header_reference header; + vector_reference items; +}; // struct container_element + +} // namespace traccc diff --git a/Traccc/core/include/traccc/edm/details/device_container.hpp b/Traccc/core/include/traccc/edm/details/device_container.hpp new file mode 100644 index 00000000000..67cb7e344c0 --- /dev/null +++ b/Traccc/core/include/traccc/edm/details/device_container.hpp @@ -0,0 +1,143 @@ +/** TRACCC library, part of the ACTS project (R&D line) + * + * (c) 2021-2022 CERN for the benefit of the ACTS project + * + * Mozilla Public License Version 2.0 + */ + +#pragma once + +// Project include(s). +#include "traccc/definitions/qualifiers.hpp" +#include "traccc/edm/details/container_base.hpp" + +// VecMem include(s). +#include +#include + +// System include(s). +#include + +namespace traccc { + +/// Host container describing objects in a given event +template +class device_container + : public container_base { + public: + /// Base class type + using base_type = container_base; + + /// Inherit all of the base class's constructors + using base_type::base_type; + + /// Constructor from a "view object" + /// + /// Note that it may also be a "data" or "buffer" object. It just needs to + /// look like a "view object". + /// + template