GPU TPC: Tracking Developments - Part 2 - DO NOT MERGE #14651

davidrohr · 2025-09-04T19:37:38Z

Developed on top of #14542

github-actions · 2025-09-04T19:37:52Z

REQUEST FOR PRODUCTION RELEASES:
To request your PR to be included in production software, please add the corresponding labels called "async-" to your PR. Add the labels directly (if you have the permissions) or add a comment of the form (note that labels are separated by a ",")

+async-label <label1>, <label2>, !<label3> ...

This will add <label1> and <label2> and removes <label3>.

The following labels are available
async-2023-pbpb-apass4
async-2023-pp-apass4
async-2024-pp-apass1
async-2022-pp-apass7
async-2024-pp-cpass0
async-2024-PbPb-apass1
async-2024-ppRef-apass1
async-2024-PbPb-apass2
async-2023-PbPb-apass5

alibuild · 2025-09-04T23:07:04Z

Error while checking build/O2/fullCI_slc9 for 15f2f6e at 2025-09-12 18:57:

## sw/BUILD/O2-full-system-test-latest/log
Detected critical problem in logfile reco_NOGPU.log


## sw/BUILD/o2checkcode-latest/log
--
========== List of errors found ==========
++ GRERR=0
++ grep -v clang-diagnostic-error error-log.txt
++ grep ' error:'
grep: error-log.txt: binary file matches
++ GRERR=1
++ [[ 1 == 0 ]]
++ mkdir -p /sw/INSTALLROOT/6174c862f76bd37149836b7f0ed19bd774a6a133/slc9_x86-64/o2checkcode/1.0-local506/etc/modulefiles
++ cat
--

Full log here.

alibuild · 2025-09-16T20:11:21Z

Error while checking build/O2/fullCI_slc9 for 4180a10 at 2025-09-16 22:11:

## sw/BUILD/O2-full-system-test-latest/log
Detected critical problem in logfile reco_NOGPU.log


## sw/BUILD/o2checkcode-latest/log
--
========== List of errors found ==========
++ GRERR=0
++ grep -v clang-diagnostic-error error-log.txt
++ grep ' error:'
grep: error-log.txt: binary file matches
++ GRERR=1
++ [[ 1 == 0 ]]
++ mkdir -p /sw/INSTALLROOT/4ba08558a8db6cd46a2ce0af7067555717824a08/slc9_x86-64/o2checkcode/1.0-local558/etc/modulefiles
++ cat
--

Full log here.

alibuild · 2025-09-29T20:23:23Z

Error while checking build/O2/fullCI_slc9 for ce1131b at 2025-09-29 22:23:

## sw/BUILD/O2-latest/log
/sw/BUILD/29e1610f7d072a30ff0c37f2b7718d1a79011a42/O2/GPU/include_gpu_onthefly/krnl_GPUTPCGMMergerHitWeights_compute.cu(35): error: identifier "GPUCA_LB_GPUTPCGMMergerHitWeights_compute" is undefined
/sw/BUILD/29e1610f7d072a30ff0c37f2b7718d1a79011a42/O2/GPU/include_gpu_onthefly/krnl_GPUTPCGMMergerHitWeights_resolve.cu(35): error: identifier "GPUCA_LB_GPUTPCGMMergerHitWeights_resolve" is undefined
/sw/BUILD/29e1610f7d072a30ff0c37f2b7718d1a79011a42/O2/GPU/include_gpu_onthefly/krnl_GPUTPCGMMergerHitWeights_resolveShared.cu(35): error: identifier "GPUCA_LB_GPUTPCGMMergerHitWeights_resolveShared" is undefined
ninja: build stopped: subcommand failed.

Full log here.

alibuild · 2025-10-05T08:45:12Z

Error while checking build/O2/fullCI_slc9 for 0982c20 at 2025-10-05 10:45:

## sw/BUILD/O2-latest/log
/sw/BUILD/29e1610f7d072a30ff0c37f2b7718d1a79011a42/O2/GPU/include_gpu_onthefly/krnl_GPUTPCGMMergerHitWeights_resolve.cu(35): error: identifier "GPUCA_LB_GPUTPCGMMergerHitWeights_resolve" is undefined
/sw/BUILD/29e1610f7d072a30ff0c37f2b7718d1a79011a42/O2/GPU/include_gpu_onthefly/krnl_GPUTPCGMMergerHitWeights_compute.cu(35): error: identifier "GPUCA_LB_GPUTPCGMMergerHitWeights_compute" is undefined
/sw/BUILD/29e1610f7d072a30ff0c37f2b7718d1a79011a42/O2/GPU/include_gpu_onthefly/krnl_GPUTPCGMMergerHitWeights_resolveShared.cu(35): error: identifier "GPUCA_LB_GPUTPCGMMergerHitWeights_resolveShared" is undefined
ninja: build stopped: subcommand failed.

Full log here.

alibuild · 2025-10-07T04:55:13Z

Error while checking build/O2/fullCI_slc9 for 0b43a47 at 2025-10-08 04:14:

No log files found

Full log here.

alibuild · 2025-10-21T19:20:51Z

Error while checking build/O2/fullCI_slc9 for 4b6c00b at 2025-10-21 21:20:

## sw/BUILD/O2-latest/log
/sw/BUILD/87b097c39f2b3f8462eee6662e094d10fe639277/O2/GPU/include_gpu_onthefly/krnl_GPUTPCGMMergerHitWeights_resolve1.hip:35:1: error: use of undeclared identifier 'GPUCA_LB_GPUTPCGMMergerHitWeights_resolve1'
ninja: build stopped: subcommand failed.

Full log here.

alibuild · 2025-10-27T17:59:05Z

Error while checking build/O2/fullCI_slc9 for fb0384b at 2025-10-27 18:59:

## sw/BUILD/O2-latest/log
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/DataTypes/GPUTPCGeometry.h:148:59: error: 'M_PI' was not declared in this scope
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/DataTypes/GPUTPCGeometry.h:148:59: error: 'M_PI' was not declared in this scope
ninja: build stopped: subcommand failed.

Full log here.

alibuild · 2025-10-27T21:07:46Z

Error while checking build/O2/fullCI_slc9 for 6346222 at 2025-10-27 22:07:

## sw/BUILD/O2-RTC-test-latest/log
/tmp/o2cagpu_rtc_132_203408360_26.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_5.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_19.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_0.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_4.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_80.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_14.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_9.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_6.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_73.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_13.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_1.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_18.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_22.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_24.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_17.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_69.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_93.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_15.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_71.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_54.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_88.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_53.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_86.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_66.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_10.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_64.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_83.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_70.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_11.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_46.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_59.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_103.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_78.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_68.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_30.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_62.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_60.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_32.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_91.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_3.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_2.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_92.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_81.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_56.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_76.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_74.cu(17569): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_203408360_7.cu(17569): error: expression must have a constant value
[0 more errors; see full log]

Full log here.

alibuild · 2025-10-28T11:13:33Z

Error while checking build/O2/fullCI_slc9 for 89a46ed at 2025-10-29 06:59:

## sw/BUILD/O2-RTC-test-latest/log
/tmp/o2cagpu_rtc_132_723026579_53.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_80.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_0.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_39.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_56.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_13.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_73.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_74.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_59.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_26.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_100.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_32.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_66.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_19.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_17.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_62.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_86.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_46.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_76.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_78.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_16.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_27.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_40.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_1.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_69.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_64.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_29.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_35.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_93.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_14.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_57.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_22.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_42.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_77.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_25.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_103.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_61.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_23.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_49.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_71.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_60.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_18.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_30.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_55.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_84.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_83.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_4.cu(17571): error: expression must have a constant value
/tmp/o2cagpu_rtc_132_723026579_72.cu(17571): error: expression must have a constant value
[0 more errors; see full log]

Full log here.

alibuild · 2025-10-31T11:57:21Z

Error while checking build/O2/fullCI_slc9 for 85f2345 at 2025-10-31 12:57:

## sw/BUILD/O2-latest/log
clang: error: llvm-spirv command failed with exit code 23 (use -v to see invocation)
ninja: build stopped: subcommand failed.

Full log here.

alibuild · 2025-11-02T19:23:06Z

Error while checking build/O2/fullCI_slc9 for e2c6886 at 2025-11-02 20:23:

## sw/BUILD/O2-latest/log
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(66): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid()") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::Fit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(66): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(392): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(392): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(337): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitAddRow") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(337): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(338): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid()") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitAddRow") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(338): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(453): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FindBestInterpolatedHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(453): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(467): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FindBestInterpolatedHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(467): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(66): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid()") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::Fit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(66): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(392): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(392): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(337): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitAddRow") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(337): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(338): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid()") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitAddRow") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(338): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(453): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FindBestInterpolatedHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(453): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(467): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FindBestInterpolatedHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(467): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(66): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid()") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::Fit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(66): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(392): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(392): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(337): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitAddRow") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(337): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(338): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid()") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitAddRow") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(338): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(453): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FindBestInterpolatedHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(453): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(467): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FindBestInterpolatedHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(467): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(66): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid()") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::Fit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(66): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(392): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(392): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(337): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitAddRow") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(337): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(338): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid()") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FitAddRow") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(338): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::markInvalid" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(453): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FindBestInterpolatedHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(453): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(467): error: calling a __host__ function("o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid() const") from a __device__ function("o2::gpu::GPUTPCGMTrackParam::FindBestInterpolatedHit") is not allowed
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Merger/GPUTPCGMTrackParam.cxx(467): error: identifier "o2::gpu::gputpcgmmergertypes::InterpolationErrorHit::isValid const" is undefined in device code
[0 more errors; see full log]

Full log here.

alibuild · 2025-12-11T21:41:38Z

Error while checking build/O2/fullCI_slc9 for 810e9a0 at 2025-12-13 08:15:

## sw/BUILD/O2-latest/log
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/qa/GPUQA.h:151:14: error: 'o2::gpu::gpudatatypes::gpuQATaskIds' has not been declared
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/qa/GPUQA.h:151:14: error: 'o2::gpu::gpudatatypes::gpuQATaskIds' has not been declared
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Global/GPUChainTracking.cxx:482:105: error: 'tasksAutomatic' is not a member of 'o2::gpu::GPUQA'
/sw/SOURCES/O2/14651-slc9_x86-64/0/GPU/GPUTracking/Global/GPUChainTracking.cxx:697:102: error: 'tasksAutomatic' is not a member of 'o2::gpu::GPUQA'
ninja: build stopped: subcommand failed.

Full log here.

… track, we will anyway rebuild the track in the future

…l to rebuild track ion between

…iation

…g with new hit

…toring interpolated position

…ly current value

…ed in NoUpdateHighIncl region

…n not available

…eaks the track, we will anyway rebuild the track in the future" This reverts commit 5c3d113. We can actually properly retry also for rebuilt tracks, we just run the retry on the last iteration after rebuilding.

…is out of range, but instead reinitialize the model

…ged tracks

davidrohr requested review from shahor02 and wiechula as code owners September 4, 2025 19:37

alibuild mentioned this pull request Sep 4, 2025

Please consider the following formatting changes to #14651 davidrohr/AliceO2#127

Open

davidrohr force-pushed the dev_pull_request5 branch from 15f2f6e to 4180a10 Compare September 16, 2025 09:19

davidrohr force-pushed the dev_pull_request5 branch 2 times, most recently from 5537581 to ce1131b Compare September 29, 2025 20:04

davidrohr force-pushed the dev_pull_request5 branch from ce1131b to 0982c20 Compare October 5, 2025 06:45

davidrohr force-pushed the dev_pull_request5 branch from 0982c20 to 0b43a47 Compare October 6, 2025 21:15

davidrohr force-pushed the dev_pull_request5 branch 2 times, most recently from 28ceb99 to 4b6c00b Compare October 21, 2025 18:59

davidrohr force-pushed the dev_pull_request5 branch from 4b6c00b to ef72a95 Compare October 21, 2025 20:13

davidrohr mentioned this pull request Oct 23, 2025

GPU: Some fixes #14766

Merged

davidrohr force-pushed the dev_pull_request5 branch from ef72a95 to fb0384b Compare October 27, 2025 14:53

davidrohr force-pushed the dev_pull_request5 branch from 6346222 to 89a46ed Compare October 28, 2025 08:56

davidrohr force-pushed the dev_pull_request5 branch 3 times, most recently from 1012461 to 85f2345 Compare October 31, 2025 09:01

davidrohr force-pushed the dev_pull_request5 branch from 85f2345 to e2c6886 Compare November 2, 2025 16:06

davidrohr requested review from arossi81 and njacazio as code owners December 11, 2025 20:05

davidrohr force-pushed the dev_pull_request5 branch from 810e9a0 to 8fe4deb Compare December 24, 2025 15:05

davidrohr added 25 commits December 25, 2025 01:48

GPU TPC: Remove option to retry refit if cluster rejection breaks the…

eebb9e5

… track, we will anyway rebuild the track in the future

GPU TPC: Add possibility to run last way of TPC fit in separate kerne…

0d00fa6

…l to rebuild track ion between

GPU TPC: Search hit closest to interpolated point in track rebuilding

258b520

GPU TPC: Fix index for hit interpolation in rows with multiple clusters

f9ee898

GPU TPC: Implement ambiguity solving for rebuild tracks cluster assoc…

37e9889

…iation

GPU: Additional debug dumps

ff97b9d

GPU TPC: Split TPC Merger Fit() into multiple functions

108982e

GPU: Add looperFollowMode option

65d12bc

GPU: simplifaction, refactoring and some minor improvements

7f4f8dc

GPU TPC: Refactor cluster merging

6d295be

GPU: Change indexing of interpolated hit temporary data

8dff506

GPU: Move some functionality to CAMath

3264c46

GPU TPC: Store track position for interpolation also when not updatin…

b135e8a

…g with new hit

GPU TPC: Extract CE crossing handling in function, and apply before s…

04036ae

…toring interpolated position

GPU TPC: If value from other side of interpolation not stored, use on…

09ae94d

…ly current value

GPU TPC: Mark clusters of rebuilt track as high incli when interpolat…

611a4b2

…ed in NoUpdateHighIncl region

GPU TPC: Pick up best hit from interpolation only, if current positio…

900d462

…n not available

GPU TPC: Interpolate missing pad rows when rebuilding track

eec1649

GPU TPC: Relax chi2cut for single-pad clusters

e299926

GPU TPC: Clean up and fix resetT0 code

e2fac1c

GPU TPC: Don't break for high sinPhi if only the linearization model …

a2b6e9a

…is out of range, but instead reinitialize the model

GPU: Add option to steer which attachment steps to run during rebuild

c2dc568

GPU TPC: Extrapolate track inward and outward when rebuilding

e4416a0

GPU: Add debug option to create temporary MC labels for collected mer…

2b03a37

…ged tracks

davidrohr force-pushed the dev_pull_request5 branch from 8fe4deb to 2b03a37 Compare December 25, 2025 00:48

GPU TPC: Tracking Developments - Part 2 - DO NOT MERGE #14651

Are you sure you want to change the base?

GPU TPC: Tracking Developments - Part 2 - DO NOT MERGE #14651

Conversation

davidrohr commented Sep 4, 2025

Uh oh!

github-actions bot commented Sep 4, 2025

Uh oh!

alibuild commented Sep 4, 2025 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

alibuild commented Sep 16, 2025

Uh oh!

alibuild commented Sep 29, 2025

Uh oh!

alibuild commented Oct 5, 2025

Uh oh!

alibuild commented Oct 7, 2025 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

alibuild commented Oct 21, 2025

Uh oh!

alibuild commented Oct 27, 2025

Uh oh!

alibuild commented Oct 27, 2025

Uh oh!

alibuild commented Oct 28, 2025 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

alibuild commented Oct 31, 2025

Uh oh!

alibuild commented Nov 2, 2025

Uh oh!

alibuild commented Dec 11, 2025 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

Reviewers

Assignees

Labels

Milestone

Development

Uh oh!

2 participants

alibuild commented Sep 4, 2025 •

edited

Loading

alibuild commented Oct 7, 2025 •

edited

Loading

alibuild commented Oct 28, 2025 •

edited

Loading

alibuild commented Dec 11, 2025 •

edited

Loading