From 8fd9596d426dd8dc15449cee0aa60ce555fa1e61 Mon Sep 17 00:00:00 2001 From: Felix Schlepper Date: Thu, 5 Nov 2026 21:28:58 +0100 Subject: [PATCH 01/46] GPUCommonDefAPI: add GPUhdnii(), host and device, inlined on the host and never on the device Co-Authored-By: Claude Opus 5.5 --- GPU/Common/GPUCommonDefAPI.h | 5 +++++ 1 file changed, 5 insertions(+) diff --git a/GPU/Common/GPUCommonDefAPI.h b/GPU/Common/GPUCommonDefAPI.h index faddcb027e0c5..21f1e7c1bc2f0 100644 --- a/GPU/Common/GPUCommonDefAPI.h +++ b/GPU/Common/GPUCommonDefAPI.h @@ -43,6 +43,7 @@ #define GPUhd() // Host and device function, inlined during GPU compilation to avoid symbol clashes in host code #define GPUhdi() inline // Host and device function, to-be-inlined on host and device #define GPUhdni() // Host and device function, not to-be-inlined automatically + #define GPUhdnii() inline // Host and device function, not-to-be-inlined on device, inlined on host #define GPUg() INVALID_TRIGGER_ERROR_NO_GPU_CODE // GPU kernel #define GPUshared() // shared memory variable declaration #define GPUglobal() // global memory variable declaration (only used for kernel input pointers) @@ -89,6 +90,7 @@ #define GPUhd() inline #define GPUhdi() inline #define GPUhdni() + #define GPUhdnii() inline #define GPUg() __kernel #define GPUshared() __local #define GPUglobal() __global @@ -157,6 +159,7 @@ #define GPUhd() inline #define GPUhdi() inline #define GPUhdni() + #define GPUhdnii() inline #define GPUg() kernel #define GPUshared() threadgroup #define GPUglobal() device @@ -187,6 +190,7 @@ #define GPUhd() __host__ __device__ inline #define GPUhdi() __host__ __device__ inline #define GPUhdni() __host__ __device__ + #define GPUhdnii() __host__ __device__ inline __attribute__((noinline)) #define GPUg() __global__ #define GPUshared() __shared__ #if defined(GPUCA_GPUCODE_DEVICE) && 0 // TODO: Fix for HIP @@ -222,6 +226,7 @@ #define GPUhd() __host__ __device__ inline #define GPUhdi() __host__ __device__ inline #define GPUhdni() __host__ __device__ + #define GPUhdnii() __host__ __device__ inline __attribute__((noinline)) #define GPUg() __global__ #define GPUshared() __shared__ #define GPUglobal() From 4369e0e526abaa6ec24f05ca8d72537e8ec86d63 Mon Sep 17 00:00:00 2001 From: Felix Schlepper Date: Thu, 5 Nov 2026 23:15:38 +0100 Subject: [PATCH 02/46] MFT: o2-mft-track-writer-workflow, the track writer on its own as ITS has it Co-Authored-By: Claude Opus 5.5 --- Detectors/ITSMFT/MFT/workflow/CMakeLists.txt | 5 +++ .../src/mft-track-writer-workflow.cxx | 35 +++++++++++++++++++ 2 files changed, 40 insertions(+) create mode 100644 Detectors/ITSMFT/MFT/workflow/src/mft-track-writer-workflow.cxx diff --git a/Detectors/ITSMFT/MFT/workflow/CMakeLists.txt b/Detectors/ITSMFT/MFT/workflow/CMakeLists.txt index 5f904c24011fb..30e7f5a42ca08 100644 --- a/Detectors/ITSMFT/MFT/workflow/CMakeLists.txt +++ b/Detectors/ITSMFT/MFT/workflow/CMakeLists.txt @@ -47,6 +47,11 @@ o2_add_executable(assessment-workflow COMPONENT_NAME mft PUBLIC_LINK_LIBRARIES O2::MFTWorkflow) +o2_add_executable(track-writer-workflow + SOURCES src/mft-track-writer-workflow.cxx + COMPONENT_NAME mft + PUBLIC_LINK_LIBRARIES O2::ITSMFTCAWriter) + o2_add_executable(cluster-writer-workflow SOURCES src/mft-cluster-writer-workflow.cxx COMPONENT_NAME mft diff --git a/Detectors/ITSMFT/MFT/workflow/src/mft-track-writer-workflow.cxx b/Detectors/ITSMFT/MFT/workflow/src/mft-track-writer-workflow.cxx new file mode 100644 index 0000000000000..973026ebf8f19 --- /dev/null +++ b/Detectors/ITSMFT/MFT/workflow/src/mft-track-writer-workflow.cxx @@ -0,0 +1,35 @@ +// Copyright 2019-2026 CERN and copyright holders of ALICE O2. +// See https://alice-o2.web.cern.ch/copyright for details of the copyright holders. +// All rights not expressly granted are reserved. +// +// This software is distributed under the terms of the GNU General Public +// License v3 (GPL Version 3), copied verbatim in the file "COPYING". +// +// In applying this license CERN does not waive the privileges and immunities +// granted to it by virtue of its status as an Intergovernmental Organization +// or submit itself to any jurisdiction. + +#include "ITSMFTCAWriter/MFTCATrackWriterSpec.h" +#include "Framework/ConfigParamSpec.h" +#include "Framework/CompletionPolicyHelpers.h" + +using namespace o2::framework; + +void customize(std::vector& policies) +{ + // ordered policies for the writers + policies.push_back(CompletionPolicyHelpers::consumeWhenAllOrdered(".*(?:MFT|mft).*[W,w]riter.*")); +} + +void customize(std::vector& workflowOptions) +{ + workflowOptions.push_back(ConfigParamSpec{"disable-mc", o2::framework::VariantType::Bool, false, {"disable MC propagation even if available"}}); +} + +#include "Framework/runDataProcessing.h" + +WorkflowSpec defineDataProcessing(ConfigContext const& configcontext) +{ + auto useMC = !configcontext.options().get("disable-mc"); + return WorkflowSpec{o2::mft::getTrackWriterSpec(useMC)}; +} From 07f57671acb3731e021a4acf5a7218a6ad014bcb Mon Sep 17 00:00:00 2001 From: Felix Schlepper Date: Thu, 5 Nov 2026 22:22:54 +0100 Subject: [PATCH 03/46] GenericTracking: the generic tracking, its workflow, study and tools Detectors/GenericTracking grows from the bounded allocator to a detector-generic tracking library with its ALICE adapters and devices: - base/ (O2::GTrkBase): the bounded allocator, the capacity estimator, the bounded columns (memlayout), the geometry and time model and the layouts of ITS, MFT and ALICE 3's TRK+FT3. - tracking/ (O2::GTrkTracking, O2::GTrkTrackingCUDA/HIP): the header tracker (tracklets, triplets, roads, fit, selection, follow, refit), the vertexer, and the device tracker on O2's GPU framework: a GPU chain of its own holds the arena's block of the reconstruction's pool, and each backend library registers its runtime when it is loaded. - workflow/ (O2::GTrkWorkflow): the cluster conversion, the measurement producer (with the mean vertex per timeframe) and the tracking device gtrk--tracker (host, or --gtrk-gpu-backend cuda|hip), publishing GenericTracking's own track records, vertices and contributor counts; the ITS and MFT adaptors to O2's TrackITS and TrackMFT with their labels; one workflow per detector, o2-gtrk-its-, -mft- and -trk-reco-workflow, the ITS's and the MFT's with the adaptor and O2's track writers. - study/ and tools/ (with the upgrades): the bench, the study sink and its QA, o2-gtrk--study-workflow, the analyses, the material LUT builder and the checks. The only lines outside Detectors/GenericTracking are in ITSMFT: the ITS and ITSMFT tracking include GenericTracking/Base/BoundedAllocator.h and link O2::GTrkBase, the allocator's new location and library. Needs GPUhdnii() (GPU/Common) and o2-mft-track-writer-workflow (MFT). Co-Authored-By: Claude Opus 5.5 --- Detectors/GenericTracking/CMakeLists.txt | 15 +- Detectors/GenericTracking/base/CMakeLists.txt | 30 + .../base/External/memlayout/wrapper.h | 570 ++ .../GenericTracking/Base}/BoundedAllocator.h | 77 +- .../GenericTracking/Base/CapacityEstimator.h | 61 + .../include/GenericTracking/Base/Columns.h | 127 + .../GenericTracking/Base/Definitions.h | 95 + .../GenericTracking/Base/FrameTiming.h | 54 + .../include/GenericTracking/Base/Geometry.h | 112 + .../GenericTracking/Base/HostAlgorithms.h | 120 + .../base/include/GenericTracking/Base/Lanes.h | 60 + .../GenericTracking/Base/LayerEnvelopes.h | 122 + .../include/GenericTracking/Base/LayerLinks.h | 223 + .../include/GenericTracking/Base/LayerSpec.h | 271 + .../GenericTracking/Base/PortableMath.h | 503 + .../include/GenericTracking/Base/Sensor.h | 54 + .../include/GenericTracking/Base/TimeEstPS.h | 99 + .../GenericTracking/Detectors/Alpide.h | 50 + .../include/GenericTracking/Detectors/ITS.h | 77 + .../include/GenericTracking/Detectors/MFT.h | 101 + .../GenericTracking/Detectors/TRKFT3.h | 303 + .../{ => base}/src/BoundedAllocator.cxx | 33 +- .../base/src/CapacityEstimator.cxx | 67 + .../base/src/LayerEnvelopes.cxx | 223 + .../{ => base}/test/CMakeLists.txt | 14 +- .../{ => base}/test/testBoundedAllocator.cxx | 6 +- .../base/test/testPortableMath.cxx | 220 + .../base/test/testTimeEstPS.cxx | 142 + .../GenericTracking/study/CMakeLists.txt | 108 + .../GenericTracking/Study/Alice/Hits.h | 98 + .../GenericTracking/Study/Alice/Tracks.h | 71 + .../include/GenericTracking/Study/Digest.h | 308 + .../include/GenericTracking/Study/MCTruth.h | 139 + .../GenericTracking/Study/MCVertexParam.h | 40 + .../Study/MCVertexReaderSpec.h | 32 + .../Study/MeasurementSinkSpec.h | 44 + .../GenericTracking/Study/ProfileReport.h | 63 + .../GenericTracking/Study/SpeciesOracle.h | 131 + .../GenericTracking/Study/StageComparison.h | 244 + .../GenericTracking/Study/StageFunnel.h | 1460 +++ .../GenericTracking/Study/Statistics.h | 58 + .../GenericTracking/Study/TrackComparison.h | 363 + .../GenericTracking/Study/TrackPulls.h | 69 + .../GenericTracking/Study/TrackingInput.h | 111 + .../GenericTracking/Study/TrackingQA.h | 358 + .../GenericTracking/Study/Validation.h | 150 + .../include/GenericTracking/Study/VertexQA.h | 217 + .../study/src/GTrkStudyLinkDef.h | 21 + Detectors/GenericTracking/study/src/Hits.cxx | 109 + .../study/src/MCVertexParam.cxx | 14 + .../study/src/MCVertexReaderSpec.cxx | 173 + .../study/src/MeasurementSinkSpec.cxx | 4519 +++++++++ .../study/src/ProfileReport.cxx | 188 + .../study/src/TrackComparison.cxx | 839 ++ .../GenericTracking/study/src/TrackPulls.cxx | 98 + .../study/src/TrackingInput.cxx | 441 + .../GenericTracking/study/src/TrackingQA.cxx | 330 + .../GenericTracking/study/src/Tracks.cxx | 134 + .../GenericTracking/study/test/CMakeLists.txt | 29 + .../study/test/testMCTruth.cxx | 60 + .../study/test/testProfile.cxx | 114 + .../study/test/testStageComparison.cxx | 264 + .../study/test/testStageFunnel.cxx | 191 + .../study/test/testTrackComparison.cxx | 380 + .../study/test/testTrackingQA.cxx | 380 + .../study/test/testTrackingTruth.cxx | 169 + .../study/test/testVertexQA.cxx | 84 + .../study/tools/CompleteAnatomy.h | 8590 +++++++++++++++++ .../GenericTracking/study/tools/FitInputs.h | 126 + .../study/tools/FollowAnatomy.h | 521 + .../study/tools/SeedingStrategy.h | 829 ++ .../study/tools/StudyWorkflow.h | 106 + .../GenericTracking/study/tools/TrackBench.h | 2242 +++++ .../study/tools/TrackBreakdown.h | 325 + .../GenericTracking/study/tools/VertexStudy.h | 2186 +++++ .../GenericTracking/study/tools/checkHits.cxx | 173 + .../study/tools/checkMatLUT.cxx | 649 ++ .../study/tools/clusterClosure.cxx | 735 ++ .../GenericTracking/study/tools/fitBench.cxx | 248 + .../study/tools/gtrk-its-study-workflow.cxx | 48 + .../study/tools/gtrk-mft-study-workflow.cxx | 48 + .../study/tools/gtrk-trk-study-workflow.cxx | 48 + .../study/tools/trackBench.cxx | 235 + .../study/tools/trackBenchITS.cxx | 20 + .../study/tools/trackBenchMFT.cxx | 20 + .../study/tools/trackBenchTRKFT3.cxx | 20 + .../GenericTracking/tools/CMakeLists.txt | 40 + .../include/GenericTracking/Tools/Material.h | 105 + .../GenericTracking/Tools/SensorFrame.h | 63 + .../GenericTracking/tools/src/Material.cxx | 259 + .../GenericTracking/tools/src/buildMatLUT.cxx | 392 + .../tools/src/importGeometry.cxx | 687 ++ .../GenericTracking/tools/src/layerLinks.cxx | 523 + .../GenericTracking/tracking/CMakeLists.txt | 42 + .../tracking/GPU/CMakeLists.txt | 120 + .../tracking/GPU/cuda/DeviceAlgorithms.cuh | 328 + .../tracking/GPU/cuda/DeviceRuntime.cu | 124 + .../tracking/GPU/cuda/DeviceTracker.cuh | 1426 +++ .../tracking/GPU/cuda/DeviceTrackerITS.cu | 35 + .../tracking/GPU/cuda/DeviceTrackerMFT.cu | 34 + .../tracking/GPU/cuda/DeviceTrackerTRKFT3.cu | 35 + .../tracking/GPU/cuda/FlatStage.cuh | 418 + .../tracking/GPU/cuda/GpuStage.cuh | 441 + .../tracking/GPU/cuda/LaneGroup.cuh | 96 + .../tracking/GPU/cuda/PathGlue.cuh | 313 + .../tracking/GPU/cuda/Platform.cuh | 130 + .../tracking/GPU/cuda/RoadGlue.cuh | 219 + .../tracking/GPU/cuda/SelectionGlue.cuh | 381 + .../tracking/GPU/cuda/SliceGlue.cuh | 252 + .../tracking/GPU/cuda/VertexGlue.cuh | 1047 ++ .../GenericTracking/Tracking/BeamConstraint.h | 180 + .../Tracking/BoundTrackState.h | 102 + .../Tracking/CovarianceTransport.h | 341 + .../GenericTracking/Tracking/CpuBackend.h | 502 + .../GenericTracking/Tracking/Definitions.h | 263 + .../GenericTracking/Tracking/DisplacedPair.h | 424 + .../include/GenericTracking/Tracking/Follow.h | 2421 +++++ .../Tracking/GPU/DeviceArena.h | 144 + .../Tracking/GPU/DeviceContext.h | 70 + .../Tracking/GPU/DeviceRuntime.h | 114 + .../Tracking/GPU/DeviceStageOutput.h | 137 + .../Tracking/GPU/DeviceStorage.h | 454 + .../Tracking/GPU/DeviceStream.h | 46 + .../Tracking/GPU/DeviceTracker.h | 333 + .../Tracking/GPU/MaterialPort.h | 41 + .../Tracking/GPU/RuntimeFactory.h | 40 + .../GenericTracking/Tracking/GridSizing.h | 143 + .../GenericTracking/Tracking/Iteration.h | 596 ++ .../GenericTracking/Tracking/LayerMaterial.h | 294 + .../GenericTracking/Tracking/LayerRowMap.h | 89 + .../GenericTracking/Tracking/LinkHoles.h | 228 + .../GenericTracking/Tracking/Material.h | 160 + .../GenericTracking/Tracking/MathUtils.h | 294 + .../GenericTracking/Tracking/Measurement.h | 193 + .../Tracking/MeasurementGrid.h | 978 ++ .../Tracking/MeasurementGridBuilder.h | 203 + .../Tracking/MeasurementRange.h | 35 + .../Tracking/MeasurementUpdate.h | 167 + .../GenericTracking/Tracking/Neighbour.h | 45 + .../GenericTracking/Tracking/NeighbourCut.h | 125 + .../GenericTracking/Tracking/NeighbourStage.h | 196 + .../Tracking/OrderedRoadCounts.h | 80 + .../GenericTracking/Tracking/PairCut.h | 873 ++ .../GenericTracking/Tracking/Parallel.h | 87 + .../GenericTracking/Tracking/Profile.h | 327 + .../GenericTracking/Tracking/Propagator.h | 514 + .../GenericTracking/Tracking/RoadCount.h | 251 + .../GenericTracking/Tracking/RoadFit.h | 312 + .../GenericTracking/Tracking/SearchWindow.h | 859 ++ .../include/GenericTracking/Tracking/Seed.h | 350 + .../GenericTracking/Tracking/Selection.h | 495 + .../GenericTracking/Tracking/SensorCells.h | 282 + .../GenericTracking/Tracking/ShapeCut.h | 69 + .../GenericTracking/Tracking/SlicePlan.h | 290 + .../GenericTracking/Tracking/StageOutput.h | 164 + .../GenericTracking/Tracking/TimeFrame.h | 102 + .../Tracking/TimeFrameScratch.h | 141 + .../GenericTracking/Tracking/TrackFitter.h | 594 ++ .../GenericTracking/Tracking/TrackRefit.h | 229 + .../GenericTracking/Tracking/Tracker.h | 1253 +++ .../Tracking/TrackingSurface.h | 235 + .../GenericTracking/Tracking/Tracklet.h | 77 + .../GenericTracking/Tracking/TrackletPaths.h | 195 + .../GenericTracking/Tracking/TrackletSort.h | 153 + .../GenericTracking/Tracking/TrackletStage.h | 732 ++ .../include/GenericTracking/Tracking/Trig.h | 154 + .../GenericTracking/Tracking/Triplet.h | 49 + .../GenericTracking/Tracking/TripletCut.h | 421 + .../Tracking/TripletEndIndex.h | 171 + .../GenericTracking/Tracking/TripletFit.h | 836 ++ .../GenericTracking/Tracking/TripletPaths.h | 142 + .../Tracking/TripletPermutation.h | 114 + .../GenericTracking/Tracking/TripletStage.h | 258 + .../GenericTracking/Tracking/VertexStage.h | 879 ++ .../GenericTracking/Tracking/VertexingPass.h | 117 + .../GenericTracking/Tracking/WorkCounters.h | 272 + .../tracking/src/GPU/DeviceArena.cxx | 167 + .../tracking/src/GPU/DeviceContext.cxx | 71 + .../tracking/src/GPU/DeviceStream.cxx | 55 + .../src/GPU/GPUChainGenericTracking.cxx | 79 + .../src/GPU/GPUChainGenericTracking.h | 83 + .../tracking/src/GPU/MaterialPort.cxx | 28 + .../tracking/src/GPU/RuntimeFactory.cxx | 57 + .../GenericTracking/tracking/src/Parallel.cxx | 179 + .../GenericTracking/tracking/src/Profile.cxx | 438 + .../tracking/src/WorkCounters.cxx | 115 + .../tracking/test/AnalyticLayout.h | 198 + .../tracking/test/BetweenBoundCheck.h | 147 + .../tracking/test/CMakeLists.txt | 60 + .../tracking/test/FixedPointReferences.h | 275 + .../tracking/test/GeneratedHelixFixture.h | 130 + .../tracking/test/GridTestSupport.h | 186 + .../GenericTracking/tracking/test/Oracles.h | 498 + .../test/PropagationPrecisionFixture.h | 108 + .../tracking/test/PullStatistics.h | 150 + .../tracking/test/SerialReferences.h | 262 + .../tracking/test/StageOutputComparison.h | 52 + .../tracking/test/SyntheticStage.h | 113 + .../tracking/test/ToyBarrel3.h | 66 + .../tracking/test/ToyBarrel4Disc8.h | 121 + .../tracking/test/ToyBarrel8.h | 81 + .../tracking/test/ToyBarrel8Fixture.h | 326 + .../tracking/test/UniformMaterial.h | 66 + .../tracking/test/device/CMakeLists.txt | 63 + .../tracking/test/device/DeviceTestFixture.h | 157 + .../tracking/test/device/DeviceTestHost.cxx | 27 + .../tracking/test/device/DeviceTestKernels.cu | 848 ++ .../tracking/test/device/DeviceTestKernels.h | 327 + .../test/device/testDeviceMaterial.cxx | 97 + .../test/device/testDeviceRuntime.cxx | 264 + .../test/device/testDeviceStorage.cxx | 205 + .../test/device/testDeviceTracker.cxx | 903 ++ .../tracking/test/device/testFlatStage.cxx | 200 + .../tracking/test/device/testGpuStage.cxx | 214 + .../tracking/test/device/testRoadGlue.cxx | 215 + .../test/device/testSelectionGlue.cxx | 235 + .../tracking/test/device/testSliceGlue.cxx | 289 + .../tracking/test/testCovarianceTransport.cxx | 474 + .../tracking/test/testDeviceArena.cxx | 212 + .../tracking/test/testDisplacedPair.cxx | 461 + .../tracking/test/testFollow.cxx | 1389 +++ .../tracking/test/testHostAlgorithms.cxx | 162 + .../tracking/test/testIteration.cxx | 516 + .../tracking/test/testIterations.cxx | 804 ++ .../tracking/test/testLayerLinks.cxx | 307 + .../tracking/test/testMathUtils.cxx | 75 + .../tracking/test/testMeasurementGrid.cxx | 776 ++ .../tracking/test/testMeasurementUpdate.cxx | 349 + .../tracking/test/testNeighbourStage.cxx | 447 + .../tracking/test/testPassiveScattering.cxx | 481 + .../tracking/test/testPlacedLayouts.cxx | 383 + .../tracking/test/testProfile.cxx | 159 + .../test/testPropagationPrecision.cxx | 185 + .../tracking/test/testPropagator.cxx | 538 ++ .../tracking/test/testRoadFit.cxx | 661 ++ .../tracking/test/testRoads.cxx | 594 ++ .../tracking/test/testSeedCuts.cxx | 353 + .../tracking/test/testSelection.cxx | 260 + .../tracking/test/testStageOutput.cxx | 410 + .../tracking/test/testTimeFrame.cxx | 161 + .../tracking/test/testTimeFrameScratch.cxx | 90 + .../tracking/test/testTrackFitter.cxx | 766 ++ .../tracking/test/testTracker.cxx | 212 + .../tracking/test/testTrackingSurface.cxx | 186 + .../tracking/test/testTrackletPaths.cxx | 306 + .../tracking/test/testTrackletSort.cxx | 249 + .../tracking/test/testTrackletStage.cxx | 731 ++ .../tracking/test/testTripletEndIndex.cxx | 186 + .../tracking/test/testTripletFit.cxx | 540 ++ .../tracking/test/testTripletStage.cxx | 850 ++ .../tracking/test/testVertexStage.cxx | 392 + .../tracking/test/testVertexingPass.cxx | 248 + .../GenericTracking/workflow/CMakeLists.txt | 69 + .../include/GenericTracking/Alice/Clusters.h | 168 + .../include/GenericTracking/Alice/Placement.h | 123 + .../include/GenericTracking/Alice/TRK.h | 110 + .../GenericTracking/Clustering/Cluster.h | 62 + .../GenericTracking/Workflow/AdaptorSpec.h | 33 + .../GenericTracking/Workflow/BeamLine.h | 75 + .../Workflow/ClusterConverterSpec.h | 34 + .../Workflow/ClusterShapeBand.h | 95 + .../GenericTracking/Workflow/Conditions.h | 175 + .../GenericTracking/Workflow/ConditionsTRK.h | 183 + .../GenericTracking/Workflow/DataSpecs.h | 177 + .../Workflow/GeometryProvider.h | 55 + .../GenericTracking/Workflow/GridConfig.h | 391 + .../GenericTracking/Workflow/MaterialLUT.h | 116 + .../Workflow/MeasurementProducerSpec.h | 33 + .../GenericTracking/Workflow/O2Parameters.h | 194 + .../GenericTracking/Workflow/RecoParam.h | 313 + .../GenericTracking/Workflow/TrackLabel.h | 130 + .../GenericTracking/Workflow/TrackerSpec.h | 42 + .../Workflow/TrackingInterface.h | 534 + .../workflow/src/AdaptorDevice.h | 565 ++ .../workflow/src/AdaptorITS.cxx | 19 + .../workflow/src/AdaptorMFT.cxx | 19 + .../workflow/src/ClusterConverterITSMFT.h | 120 + .../GenericTracking/workflow/src/Clusters.cxx | 70 + .../workflow/src/GTrkWorkflowLinkDef.h | 29 + .../GenericTracking/workflow/src/ITS.cxx | 46 + .../GenericTracking/workflow/src/MFT.cxx | 46 + .../workflow/src/MeasurementProducer.h | 284 + .../workflow/src/Placement.cxx | 83 + .../workflow/src/RecoParam.cxx | 23 + .../workflow/src/RecoWorkflow.h | 70 + .../GenericTracking/workflow/src/TRK.cxx | 354 + .../workflow/src/TrackerDevice.h | 170 + .../workflow/src/TrackerITS.cxx | 25 + .../workflow/src/TrackerMFT.cxx | 25 + .../workflow/src/TrackerTRKFT3.cxx | 25 + .../workflow/src/gtrk-its-reco-workflow.cxx | 77 + .../workflow/src/gtrk-mft-reco-workflow.cxx | 74 + .../workflow/src/gtrk-trk-reco-workflow.cxx | 56 + .../workflow/test/CMakeLists.txt | 30 + .../workflow/test/testAliceClusters.cxx | 112 + .../workflow/test/testAliceLayouts.cxx | 154 + .../workflow/test/testClusterShapeBand.cxx | 69 + .../workflow/test/testDiscChainPulls.cxx | 347 + .../workflow/test/testGridConfig.cxx | 92 + .../workflow/test/testO2Parameters.cxx | 327 + .../workflow/test/testTrackerMemoryPool.cxx | 49 + .../workflow/test/testTrkChainPulls.cxx | 745 ++ Detectors/ITSMFT/ITS/tracking/CMakeLists.txt | 2 +- .../GPU/ITStrackingGPU/TimeFrameGPU.h | 2 +- .../GPU/ITStrackingGPU/TrackingKernels.h | 2 +- .../ITS/tracking/GPU/cuda/TimeFrameGPU.cu | 2 +- .../include/ITStracking/LineVertexerHelpers.h | 2 +- .../tracking/include/ITStracking/TimeFrame.h | 2 +- .../tracking/include/ITStracking/Tracker.h | 2 +- .../include/ITStracking/TrackerTraits.h | 2 +- .../include/ITStracking/TrackingInterface.h | 2 +- .../tracking/include/ITStracking/Vertexer.h | 2 +- .../include/ITStracking/VertexerTraits.h | 2 +- .../ITSMFT/ITS/tracking/src/TimeFrame.cxx | 2 +- Detectors/ITSMFT/ITS/tracking/src/Tracker.cxx | 2 +- .../ITSMFT/ITS/tracking/src/TrackerTraits.cxx | 2 +- .../ITS/tracking/src/TrackingInterface.cxx | 2 +- .../ITSMFT/ITS/tracking/src/Vertexer.cxx | 2 +- .../ITS/tracking/src/VertexerTraits.cxx | 2 +- .../ITSMFT/common/tracking/CMakeLists.txt | 2 +- .../include/ITSMFTTracking/BoundedAllocator.h | 4 +- 321 files changed, 92758 insertions(+), 111 deletions(-) create mode 100644 Detectors/GenericTracking/base/CMakeLists.txt create mode 100644 Detectors/GenericTracking/base/External/memlayout/wrapper.h rename Detectors/GenericTracking/{include/GenericTracking => base/include/GenericTracking/Base}/BoundedAllocator.h (69%) create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/CapacityEstimator.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/Columns.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/Definitions.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/FrameTiming.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/Geometry.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/HostAlgorithms.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/Lanes.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/LayerEnvelopes.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/LayerLinks.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/LayerSpec.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/PortableMath.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/Sensor.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Base/TimeEstPS.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Detectors/Alpide.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Detectors/ITS.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Detectors/MFT.h create mode 100644 Detectors/GenericTracking/base/include/GenericTracking/Detectors/TRKFT3.h rename Detectors/GenericTracking/{ => base}/src/BoundedAllocator.cxx (77%) create mode 100644 Detectors/GenericTracking/base/src/CapacityEstimator.cxx create mode 100644 Detectors/GenericTracking/base/src/LayerEnvelopes.cxx rename Detectors/GenericTracking/{ => base}/test/CMakeLists.txt (60%) rename Detectors/GenericTracking/{ => base}/test/testBoundedAllocator.cxx (97%) create mode 100644 Detectors/GenericTracking/base/test/testPortableMath.cxx create mode 100644 Detectors/GenericTracking/base/test/testTimeEstPS.cxx create mode 100644 Detectors/GenericTracking/study/CMakeLists.txt create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/Alice/Hits.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/Alice/Tracks.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/Digest.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/MCTruth.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/MCVertexParam.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/MCVertexReaderSpec.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/MeasurementSinkSpec.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/ProfileReport.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/SpeciesOracle.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/StageComparison.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/StageFunnel.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/Statistics.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/TrackComparison.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/TrackPulls.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/TrackingInput.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/TrackingQA.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/Validation.h create mode 100644 Detectors/GenericTracking/study/include/GenericTracking/Study/VertexQA.h create mode 100644 Detectors/GenericTracking/study/src/GTrkStudyLinkDef.h create mode 100644 Detectors/GenericTracking/study/src/Hits.cxx create mode 100644 Detectors/GenericTracking/study/src/MCVertexParam.cxx create mode 100644 Detectors/GenericTracking/study/src/MCVertexReaderSpec.cxx create mode 100644 Detectors/GenericTracking/study/src/MeasurementSinkSpec.cxx create mode 100644 Detectors/GenericTracking/study/src/ProfileReport.cxx create mode 100644 Detectors/GenericTracking/study/src/TrackComparison.cxx create mode 100644 Detectors/GenericTracking/study/src/TrackPulls.cxx create mode 100644 Detectors/GenericTracking/study/src/TrackingInput.cxx create mode 100644 Detectors/GenericTracking/study/src/TrackingQA.cxx create mode 100644 Detectors/GenericTracking/study/src/Tracks.cxx create mode 100644 Detectors/GenericTracking/study/test/CMakeLists.txt create mode 100644 Detectors/GenericTracking/study/test/testMCTruth.cxx create mode 100644 Detectors/GenericTracking/study/test/testProfile.cxx create mode 100644 Detectors/GenericTracking/study/test/testStageComparison.cxx create mode 100644 Detectors/GenericTracking/study/test/testStageFunnel.cxx create mode 100644 Detectors/GenericTracking/study/test/testTrackComparison.cxx create mode 100644 Detectors/GenericTracking/study/test/testTrackingQA.cxx create mode 100644 Detectors/GenericTracking/study/test/testTrackingTruth.cxx create mode 100644 Detectors/GenericTracking/study/test/testVertexQA.cxx create mode 100644 Detectors/GenericTracking/study/tools/CompleteAnatomy.h create mode 100644 Detectors/GenericTracking/study/tools/FitInputs.h create mode 100644 Detectors/GenericTracking/study/tools/FollowAnatomy.h create mode 100644 Detectors/GenericTracking/study/tools/SeedingStrategy.h create mode 100644 Detectors/GenericTracking/study/tools/StudyWorkflow.h create mode 100644 Detectors/GenericTracking/study/tools/TrackBench.h create mode 100644 Detectors/GenericTracking/study/tools/TrackBreakdown.h create mode 100644 Detectors/GenericTracking/study/tools/VertexStudy.h create mode 100644 Detectors/GenericTracking/study/tools/checkHits.cxx create mode 100644 Detectors/GenericTracking/study/tools/checkMatLUT.cxx create mode 100644 Detectors/GenericTracking/study/tools/clusterClosure.cxx create mode 100644 Detectors/GenericTracking/study/tools/fitBench.cxx create mode 100644 Detectors/GenericTracking/study/tools/gtrk-its-study-workflow.cxx create mode 100644 Detectors/GenericTracking/study/tools/gtrk-mft-study-workflow.cxx create mode 100644 Detectors/GenericTracking/study/tools/gtrk-trk-study-workflow.cxx create mode 100644 Detectors/GenericTracking/study/tools/trackBench.cxx create mode 100644 Detectors/GenericTracking/study/tools/trackBenchITS.cxx create mode 100644 Detectors/GenericTracking/study/tools/trackBenchMFT.cxx create mode 100644 Detectors/GenericTracking/study/tools/trackBenchTRKFT3.cxx create mode 100644 Detectors/GenericTracking/tools/CMakeLists.txt create mode 100644 Detectors/GenericTracking/tools/include/GenericTracking/Tools/Material.h create mode 100644 Detectors/GenericTracking/tools/include/GenericTracking/Tools/SensorFrame.h create mode 100644 Detectors/GenericTracking/tools/src/Material.cxx create mode 100644 Detectors/GenericTracking/tools/src/buildMatLUT.cxx create mode 100644 Detectors/GenericTracking/tools/src/importGeometry.cxx create mode 100644 Detectors/GenericTracking/tools/src/layerLinks.cxx create mode 100644 Detectors/GenericTracking/tracking/CMakeLists.txt create mode 100644 Detectors/GenericTracking/tracking/GPU/CMakeLists.txt create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/DeviceAlgorithms.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/DeviceRuntime.cu create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/DeviceTracker.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/DeviceTrackerITS.cu create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/DeviceTrackerMFT.cu create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/DeviceTrackerTRKFT3.cu create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/FlatStage.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/GpuStage.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/LaneGroup.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/PathGlue.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/Platform.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/RoadGlue.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/SelectionGlue.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/SliceGlue.cuh create mode 100644 Detectors/GenericTracking/tracking/GPU/cuda/VertexGlue.cuh create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/BeamConstraint.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/BoundTrackState.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/CovarianceTransport.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/CpuBackend.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Definitions.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/DisplacedPair.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Follow.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GPU/DeviceArena.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GPU/DeviceContext.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GPU/DeviceRuntime.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GPU/DeviceStageOutput.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GPU/DeviceStorage.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GPU/DeviceStream.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GPU/DeviceTracker.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GPU/MaterialPort.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GPU/RuntimeFactory.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/GridSizing.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Iteration.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/LayerMaterial.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/LayerRowMap.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/LinkHoles.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Material.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/MathUtils.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Measurement.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/MeasurementGrid.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/MeasurementGridBuilder.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/MeasurementRange.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/MeasurementUpdate.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Neighbour.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/NeighbourCut.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/NeighbourStage.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/OrderedRoadCounts.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/PairCut.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Parallel.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Profile.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Propagator.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/RoadCount.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/RoadFit.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/SearchWindow.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Seed.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Selection.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/SensorCells.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/ShapeCut.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/SlicePlan.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/StageOutput.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TimeFrame.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TimeFrameScratch.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TrackFitter.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TrackRefit.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Tracker.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TrackingSurface.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Tracklet.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TrackletPaths.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TrackletSort.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TrackletStage.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Trig.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/Triplet.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TripletCut.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TripletEndIndex.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TripletFit.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TripletPaths.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TripletPermutation.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/TripletStage.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/VertexStage.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/VertexingPass.h create mode 100644 Detectors/GenericTracking/tracking/include/GenericTracking/Tracking/WorkCounters.h create mode 100644 Detectors/GenericTracking/tracking/src/GPU/DeviceArena.cxx create mode 100644 Detectors/GenericTracking/tracking/src/GPU/DeviceContext.cxx create mode 100644 Detectors/GenericTracking/tracking/src/GPU/DeviceStream.cxx create mode 100644 Detectors/GenericTracking/tracking/src/GPU/GPUChainGenericTracking.cxx create mode 100644 Detectors/GenericTracking/tracking/src/GPU/GPUChainGenericTracking.h create mode 100644 Detectors/GenericTracking/tracking/src/GPU/MaterialPort.cxx create mode 100644 Detectors/GenericTracking/tracking/src/GPU/RuntimeFactory.cxx create mode 100644 Detectors/GenericTracking/tracking/src/Parallel.cxx create mode 100644 Detectors/GenericTracking/tracking/src/Profile.cxx create mode 100644 Detectors/GenericTracking/tracking/src/WorkCounters.cxx create mode 100644 Detectors/GenericTracking/tracking/test/AnalyticLayout.h create mode 100644 Detectors/GenericTracking/tracking/test/BetweenBoundCheck.h create mode 100644 Detectors/GenericTracking/tracking/test/CMakeLists.txt create mode 100644 Detectors/GenericTracking/tracking/test/FixedPointReferences.h create mode 100644 Detectors/GenericTracking/tracking/test/GeneratedHelixFixture.h create mode 100644 Detectors/GenericTracking/tracking/test/GridTestSupport.h create mode 100644 Detectors/GenericTracking/tracking/test/Oracles.h create mode 100644 Detectors/GenericTracking/tracking/test/PropagationPrecisionFixture.h create mode 100644 Detectors/GenericTracking/tracking/test/PullStatistics.h create mode 100644 Detectors/GenericTracking/tracking/test/SerialReferences.h create mode 100644 Detectors/GenericTracking/tracking/test/StageOutputComparison.h create mode 100644 Detectors/GenericTracking/tracking/test/SyntheticStage.h create mode 100644 Detectors/GenericTracking/tracking/test/ToyBarrel3.h create mode 100644 Detectors/GenericTracking/tracking/test/ToyBarrel4Disc8.h create mode 100644 Detectors/GenericTracking/tracking/test/ToyBarrel8.h create mode 100644 Detectors/GenericTracking/tracking/test/ToyBarrel8Fixture.h create mode 100644 Detectors/GenericTracking/tracking/test/UniformMaterial.h create mode 100644 Detectors/GenericTracking/tracking/test/device/CMakeLists.txt create mode 100644 Detectors/GenericTracking/tracking/test/device/DeviceTestFixture.h create mode 100644 Detectors/GenericTracking/tracking/test/device/DeviceTestHost.cxx create mode 100644 Detectors/GenericTracking/tracking/test/device/DeviceTestKernels.cu create mode 100644 Detectors/GenericTracking/tracking/test/device/DeviceTestKernels.h create mode 100644 Detectors/GenericTracking/tracking/test/device/testDeviceMaterial.cxx create mode 100644 Detectors/GenericTracking/tracking/test/device/testDeviceRuntime.cxx create mode 100644 Detectors/GenericTracking/tracking/test/device/testDeviceStorage.cxx create mode 100644 Detectors/GenericTracking/tracking/test/device/testDeviceTracker.cxx create mode 100644 Detectors/GenericTracking/tracking/test/device/testFlatStage.cxx create mode 100644 Detectors/GenericTracking/tracking/test/device/testGpuStage.cxx create mode 100644 Detectors/GenericTracking/tracking/test/device/testRoadGlue.cxx create mode 100644 Detectors/GenericTracking/tracking/test/device/testSelectionGlue.cxx create mode 100644 Detectors/GenericTracking/tracking/test/device/testSliceGlue.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testCovarianceTransport.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testDeviceArena.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testDisplacedPair.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testFollow.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testHostAlgorithms.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testIteration.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testIterations.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testLayerLinks.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testMathUtils.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testMeasurementGrid.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testMeasurementUpdate.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testNeighbourStage.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testPassiveScattering.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testPlacedLayouts.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testProfile.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testPropagationPrecision.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testPropagator.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testRoadFit.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testRoads.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testSeedCuts.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testSelection.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testStageOutput.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTimeFrame.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTimeFrameScratch.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTrackFitter.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTracker.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTrackingSurface.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTrackletPaths.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTrackletSort.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTrackletStage.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTripletEndIndex.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTripletFit.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testTripletStage.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testVertexStage.cxx create mode 100644 Detectors/GenericTracking/tracking/test/testVertexingPass.cxx create mode 100644 Detectors/GenericTracking/workflow/CMakeLists.txt create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Alice/Clusters.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Alice/Placement.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Alice/TRK.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Clustering/Cluster.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/AdaptorSpec.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/BeamLine.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/ClusterConverterSpec.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/ClusterShapeBand.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/Conditions.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/ConditionsTRK.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/DataSpecs.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/GeometryProvider.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/GridConfig.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/MaterialLUT.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/MeasurementProducerSpec.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/O2Parameters.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/RecoParam.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/TrackLabel.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/TrackerSpec.h create mode 100644 Detectors/GenericTracking/workflow/include/GenericTracking/Workflow/TrackingInterface.h create mode 100644 Detectors/GenericTracking/workflow/src/AdaptorDevice.h create mode 100644 Detectors/GenericTracking/workflow/src/AdaptorITS.cxx create mode 100644 Detectors/GenericTracking/workflow/src/AdaptorMFT.cxx create mode 100644 Detectors/GenericTracking/workflow/src/ClusterConverterITSMFT.h create mode 100644 Detectors/GenericTracking/workflow/src/Clusters.cxx create mode 100644 Detectors/GenericTracking/workflow/src/GTrkWorkflowLinkDef.h create mode 100644 Detectors/GenericTracking/workflow/src/ITS.cxx create mode 100644 Detectors/GenericTracking/workflow/src/MFT.cxx create mode 100644 Detectors/GenericTracking/workflow/src/MeasurementProducer.h create mode 100644 Detectors/GenericTracking/workflow/src/Placement.cxx create mode 100644 Detectors/GenericTracking/workflow/src/RecoParam.cxx create mode 100644 Detectors/GenericTracking/workflow/src/RecoWorkflow.h create mode 100644 Detectors/GenericTracking/workflow/src/TRK.cxx create mode 100644 Detectors/GenericTracking/workflow/src/TrackerDevice.h create mode 100644 Detectors/GenericTracking/workflow/src/TrackerITS.cxx create mode 100644 Detectors/GenericTracking/workflow/src/TrackerMFT.cxx create mode 100644 Detectors/GenericTracking/workflow/src/TrackerTRKFT3.cxx create mode 100644 Detectors/GenericTracking/workflow/src/gtrk-its-reco-workflow.cxx create mode 100644 Detectors/GenericTracking/workflow/src/gtrk-mft-reco-workflow.cxx create mode 100644 Detectors/GenericTracking/workflow/src/gtrk-trk-reco-workflow.cxx create mode 100644 Detectors/GenericTracking/workflow/test/CMakeLists.txt create mode 100644 Detectors/GenericTracking/workflow/test/testAliceClusters.cxx create mode 100644 Detectors/GenericTracking/workflow/test/testAliceLayouts.cxx create mode 100644 Detectors/GenericTracking/workflow/test/testClusterShapeBand.cxx create mode 100644 Detectors/GenericTracking/workflow/test/testDiscChainPulls.cxx create mode 100644 Detectors/GenericTracking/workflow/test/testGridConfig.cxx create mode 100644 Detectors/GenericTracking/workflow/test/testO2Parameters.cxx create mode 100644 Detectors/GenericTracking/workflow/test/testTrackerMemoryPool.cxx create mode 100644 Detectors/GenericTracking/workflow/test/testTrkChainPulls.cxx diff --git a/Detectors/GenericTracking/CMakeLists.txt b/Detectors/GenericTracking/CMakeLists.txt index 2c865e20d88a5..2d6b535792065 100644 --- a/Detectors/GenericTracking/CMakeLists.txt +++ b/Detectors/GenericTracking/CMakeLists.txt @@ -8,8 +8,15 @@ # In applying this license CERN does not waive the privileges and immunities # granted to it by virtue of its status as an Intergovernmental Organization # or submit itself to any jurisdiction. -o2_add_library(GenericTracking - SOURCES src/BoundedAllocator.cxx - PRIVATE_LINK_LIBRARIES O2::GPUCommon O2::FrameworkLogger) +# Every unit that instantiates the header tracker -- the tracking device's, the +# study library's, the sink's and the benches' -- compiles at -O3 in +# RelWithDebInfo, where the tracker runs faster than at O2's -O2. +set(GTRK_TRACKER_OPTIMIZATION "$<$,$>:-O3>") -add_subdirectory(test) +add_subdirectory(base) +add_subdirectory(tracking) +add_subdirectory(workflow) +if(ENABLE_UPGRADES) + add_subdirectory(tools) + add_subdirectory(study) +endif() diff --git a/Detectors/GenericTracking/base/CMakeLists.txt b/Detectors/GenericTracking/base/CMakeLists.txt new file mode 100644 index 0000000000000..4915ff0996a82 --- /dev/null +++ b/Detectors/GenericTracking/base/CMakeLists.txt @@ -0,0 +1,30 @@ +# Copyright 2019-2020 CERN and copyright holders of ALICE O2. +# See https://alice-o2.web.cern.ch/copyright for details of the copyright holders. +# All rights not expressly granted are reserved. +# +# This software is distributed under the terms of the GNU General Public +# License v3 (GPL Version 3), copied verbatim in the file "COPYING". +# +# In applying this license CERN does not waive the privileges and immunities +# granted to it by virtue of its status as an Intergovernmental Organization +# or submit itself to any jurisdiction. + +# The base: what includes no tracking code -- the geometry and time model +# (Base/, the detectors' layouts in Detectors/), the bounded allocator, the SoA +# columns, the host algorithms, the lanes, the portable math and the capacity +# estimator. Header code but for the three sources; host only. +o2_add_library(GTrkBase + SOURCES src/BoundedAllocator.cxx + src/CapacityEstimator.cxx + src/LayerEnvelopes.cxx + PUBLIC_LINK_LIBRARIES O2::GPUCommon + TARGETVARNAME targetName) + +# o2_add_library installs the headers of include/ only. +target_include_directories(${targetName} PUBLIC $) +target_include_directories(${targetName} SYSTEM PUBLIC $ + $) +install(DIRECTORY include/GenericTracking DESTINATION ${CMAKE_INSTALL_INCLUDEDIR}) +install(DIRECTORY External/memlayout DESTINATION ${CMAKE_INSTALL_INCLUDEDIR}/GenericTracking/External) + +add_subdirectory(test) diff --git a/Detectors/GenericTracking/base/External/memlayout/wrapper.h b/Detectors/GenericTracking/base/External/memlayout/wrapper.h new file mode 100644 index 0000000000000..904e0c8691c08 --- /dev/null +++ b/Detectors/GenericTracking/base/External/memlayout/wrapper.h @@ -0,0 +1,570 @@ +// Copyright 2019-2020 CERN and copyright holders of ALICE O2. +// See https://alice-o2.web.cern.ch/copyright for details of the copyright holders. +// All rights not expressly granted are reserved. +// +// This software is distributed under the terms of the GNU General Public +// License v3 (GPL Version 3), copied verbatim in the file "COPYING". +// +// In applying this license CERN does not waive the privileges and immunities +// granted to it by virtue of its status as an Intergovernmental Organization +// or submit itself to any jurisdiction. + +#ifndef MEMLAYOUT_H +#define MEMLAYOUT_H + +// Upstream forward-declares std::iterator_traits here. libc++ declares it +// differently, so the two are ambiguous in any translation unit that also +// includes a standard container. Including the real header instead. +#include +#include + +#ifndef MEMLAYOUT_HOST_DEVICE +#ifdef GPUhdi +#define MEMLAYOUT_HOST_DEVICE GPUhdi() +#else +#define MEMLAYOUT_HOST_DEVICE +#endif +#endif + +// Local addition: the element access -- operator[] and what it calls -- is +// always inlined. It builds a proxy of one reference per column; inlined, +// the references a caller never reads fold away and `view[i].x` is one load, +// but a compiler that keeps it out of line (GCC at -O2 does, for a record of +// many columns) computes and stores every one of them on each call. nvcc and +// HIP's clang take the attribute as they take __forceinline__. +// +// It only reaches as far as the proxy does inline: a proxy returned by, or +// passed to, a function the compiler keeps out of line is built whole again. +// Code that needs a few columns of a row is better served reading them from +// the view, `view.x[i]`, which builds no proxy at all. +#ifndef MEMLAYOUT_INLINE +#if defined(__GNUC__) || defined(__clang__) +#define MEMLAYOUT_INLINE __attribute__((always_inline)) +#else +#define MEMLAYOUT_INLINE +#endif +#endif + +namespace memlayout +{ + +using size_t = decltype(sizeof(0)); +using ptrdiff_t = decltype(static_cast(nullptr) - static_cast(nullptr)); + +template +using value = T; +template +using reference = T&; +template +using const_reference = const T&; +template +using pointer = T*; +template +using const_pointer = const T*; + +// Local addition: a default value for a field of a struct templated on its +// storage. In the record form, Container = value, the field takes `v`; in +// every other form it is value-initialised -- a null pointer, an empty +// container -- so one declaration serves as record and as columns: +// +// C x = memlayout::default_value(0.F); +// +// The reference forms have no default; that is safe because a default member +// initializer of a class template is instantiated only where it is used, and +// the wrapper always initialises every member. +template