Skip to content

Commit fce6045

Browse files
ChSonnabendChSonnabend
authored
Merge branch 'AliceO2Group:dev' into devel_fst_numactl
2 parents 18943ef + 4ea5007 commit fce6045

26 files changed

Lines changed: 345 additions & 124 deletions

File tree

‎Common/Constants/include/CommonConstants/PhysicsConstants.h‎

Lines changed: 3 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -97,7 +97,8 @@ enum Pdg {
9797
kHyperHelium4Sigma = 1110020040,
9898
kLambda1520_Py = 102134,
9999
kK1_1270_0 = 10313,
100-
kK1_1270Plus = 10323
100+
kK1_1270Plus = 10323,
101+
kCDeuteron = 2010010020
101102
};
102103

103104
/// \brief Declarations of masses for additional particles
@@ -168,6 +169,7 @@ constexpr double MassHyperHelium4Sigma = 3.995;
168169
constexpr double MassLambda1520_Py = 1.5195;
169170
constexpr double MassK1_1270_0 = 1.253;
170171
constexpr double MassK1_1270Plus = 1.272;
172+
constexpr double MassCDeuteron = 3.226;
171173

172174
/// \brief Declarations of masses for particles in ROOT PDG_t
173175
constexpr double MassDown = 0.00467;

‎Common/Constants/include/CommonConstants/make_pdg_header.py‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -156,6 +156,7 @@ class Pdg(Enum):
156156
kLambda1520_Py = 102134 # PYTHIA code different from PDG
157157
kK1_1270_0 = 10313
158158
kK1_1270Plus = 10323
159+
kCDeuteron = 2010010020
159160

160161

161162
dbPdg = o2.O2DatabasePDG

‎DataFormats/simulation/include/SimulationDataFormat/O2DatabasePDG.h‎

Lines changed: 15 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -732,6 +732,21 @@ inline void O2DatabasePDG::addALICEParticles(TDatabasePDG* db)
732732
db->AddParticle("Xi_c_0_3080", "Xi_c_0_3080", 3.0799, false, 0.0056, 0, "Resonance", ionCode);
733733
}
734734
db->AddAntiParticle("Anti-Xi_c_0_3080", -ionCode);
735+
ionCode = 24124;
736+
if (!db->GetParticle(ionCode)) {
737+
db->AddParticle("Lambda_c_Plus_2860", "Lambda_c_Plus_2860", 2.8561, false, 0.0680, 0, "Resonance", ionCode);
738+
}
739+
db->AddAntiParticle("Anti-Lambda_c_Minus_2860", -ionCode);
740+
ionCode = 24126;
741+
if (!db->GetParticle(ionCode)) {
742+
db->AddParticle("Lambda_c_Plus_2880", "Lambda_c_Plus_2880", 2.8816, false, 0.0056, 0, "Resonance", ionCode);
743+
}
744+
db->AddAntiParticle("Anti-Lambda_c_Minus_2880", -ionCode);
745+
ionCode = 4125;
746+
if (!db->GetParticle(ionCode)) {
747+
db->AddParticle("Lambda_c_Plus_2940", "Lambda_c_Plus_2940", 2.9396, false, 0.0200, 0, "Resonance", ionCode);
748+
}
749+
db->AddAntiParticle("Anti-Lambda_c_Minus_2940", -ionCode);
735750

736751
// d*(2380) - dibaryon resonance
737752

‎Detectors/ITSMFT/ITS/tracking/GPU/ITStrackingGPU/TrackingKernels.h‎

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -193,8 +193,9 @@ void processNeighboursHandler(const int startLevel,
193193
const float MaxChi2ClusterAttachment,
194194
const float maxChi2NDF,
195195
const int maxHoles,
196-
const int minTrackLength,
196+
const int minSeedingClusters,
197197
const LayerMask holeLayerMask,
198+
const LayerMask nonSeedingLayerMask,
198199
const std::vector<float>& layerxX0Host,
199200
const o2::base::Propagator* propagator,
200201
const o2::base::PropagatorF::MatCorrType matCorrType,

‎Detectors/ITSMFT/ITS/tracking/GPU/cuda/TrackerTraitsGPU.cxx‎

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -331,8 +331,9 @@ void TrackerTraitsGPU<NLayers>::findRoads(const int iteration)
331331
this->mTrkParams[iteration].MaxChi2ClusterAttachment,
332332
this->mTrkParams[iteration].MaxChi2NDF,
333333
this->mTrkParams[iteration].MaxHoles,
334-
this->mTrkParams[iteration].MinTrackLength,
334+
this->mTrkParams[iteration].getMinSeedingClusters(),
335335
this->mTrkParams[iteration].HoleLayerMask,
336+
this->mTrkParams[iteration].getNonSeedingLayerMask(),
336337
this->mTrkParams[iteration].LayerxX0,
337338
mTimeFrameGPU->getDevicePropagator(),
338339
this->mTrkParams[iteration].CorrType,

‎Detectors/ITSMFT/ITS/tracking/GPU/cuda/TrackingKernels.cu‎

Lines changed: 8 additions & 22 deletions
Original file line numberDiff line numberDiff line change
@@ -32,8 +32,8 @@
3232
#include "ITStracking/Tracklet.h"
3333
#include "ITStracking/Cluster.h"
3434
#include "ITStracking/Cell.h"
35-
#include "ITStracking/TrackFollower.h"
3635
#include "ITStracking/TrackHelpers.h"
36+
#include "ITStracking/TrackFollower.h"
3737
#include "DataFormatsITS/TrackITS.h"
3838
#include "ITStrackingGPU/TrackingKernels.h"
3939
#include "ITStrackingGPU/Utils.h"
@@ -86,23 +86,6 @@ struct is_valid_pair {
8686
}
8787
};
8888

89-
template <int NLayers>
90-
struct seed_selector {
91-
float mMaxQ2Pt;
92-
float mMaxChi2;
93-
int mMaxHoles;
94-
int mMinTrackLength;
95-
LayerMask mHoleLayerMask;
96-
97-
GPUhd() seed_selector(float maxQ2Pt, float maxChi2, int maxHoles, int minTrackLength, LayerMask holeLayerMask) : mMaxQ2Pt(maxQ2Pt), mMaxChi2(maxChi2), mMaxHoles(maxHoles), mMinTrackLength(minTrackLength), mHoleLayerMask(holeLayerMask) {}
98-
GPUhd() bool operator()(const TrackSeed<NLayers>& seed) const
99-
{
100-
return !(seed.getQ2Pt() > mMaxQ2Pt || seed.getChi2() > mMaxChi2) &&
101-
seed.getHitLayerMask().length() >= mMinTrackLength &&
102-
seed.getHitLayerMask().isAllowed(mMaxHoles, mHoleLayerMask);
103-
}
104-
};
105-
10689
struct compare_track_chi2 {
10790
GPUhd() bool operator()(const TrackITSExt& a, const TrackITSExt& b) const
10891
{
@@ -975,8 +958,9 @@ void processNeighboursHandler(const int startLevel,
975958
const float maxChi2ClusterAttachment,
976959
const float maxChi2NDF,
977960
const int maxHoles,
978-
const int minTrackLength,
961+
const int minSeedingClusters,
979962
const LayerMask holeLayerMask,
963+
const LayerMask nonSeedingLayerMask,
980964
const std::vector<float>& layerxX0Host,
981965
const o2::base::Propagator* propagator,
982966
const o2::base::PropagatorF::MatCorrType matCorrType,
@@ -1110,7 +1094,7 @@ void processNeighboursHandler(const int startLevel,
11101094
}
11111095
GPUChkErrS(cudaStreamSynchronize(gpu::Stream::DefaultStream));
11121096
thrust::device_vector<TrackSeed<NLayers>, gpu::TypedAllocator<TrackSeed<NLayers>>> outSeeds(updatedCellSeed.size(), allocTrackSeed);
1113-
auto end = thrust::copy_if(nosync_policy, updatedCellSeed.begin(), updatedCellSeed.end(), outSeeds.begin(), gpu::seed_selector<NLayers>(1.e3, maxChi2NDF * ((startLevel + 2) * 2 - 5), maxHoles, minTrackLength, holeLayerMask));
1097+
auto end = thrust::copy_if(nosync_policy, updatedCellSeed.begin(), updatedCellSeed.end(), outSeeds.begin(), track::TrackSeedSelector<NLayers>{constants::MaxTrackSeedQ2Pt, maxChi2NDF, startLevel, maxHoles, minSeedingClusters, holeLayerMask, nonSeedingLayerMask});
11141098
auto s{end - outSeeds.begin()};
11151099
seedsHost.reserve(seedsHost.size() + s);
11161100
thrust::copy(outSeeds.begin(), outSeeds.begin() + s, std::back_inserter(seedsHost));
@@ -1380,8 +1364,9 @@ template void processNeighboursHandler<7>(const int startLevel,
13801364
const float maxChi2ClusterAttachment,
13811365
const float maxChi2NDF,
13821366
const int maxHoles,
1383-
const int minTrackLength,
1367+
const int minSeedingClusters,
13841368
const LayerMask holeLayerMask,
1369+
const LayerMask nonSeedingLayerMask,
13851370
const std::vector<float>& layerxX0Host,
13861371
const o2::base::Propagator* propagator,
13871372
const o2::base::PropagatorF::MatCorrType matCorrType,
@@ -1580,8 +1565,9 @@ template void processNeighboursHandler<11>(const int startLevel,
15801565
const float maxChi2ClusterAttachment,
15811566
const float maxChi2NDF,
15821567
const int maxHoles,
1583-
const int minTrackLength,
1568+
const int minSeedingClusters,
15841569
const LayerMask holeLayerMask,
1570+
const LayerMask nonSeedingLayerMask,
15851571
const std::vector<float>& layerxX0Host,
15861572
const o2::base::Propagator* propagator,
15871573
const o2::base::PropagatorF::MatCorrType matCorrType,

‎Detectors/ITSMFT/ITS/tracking/include/ITStracking/Configuration.h‎

Lines changed: 35 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -46,15 +46,42 @@ enum class IterationStep : uint16_t {
4646
using IterationSteps = o2::utils::EnumFlags<IterationStep>;
4747

4848
struct TrackingParameters {
49-
int CellMinimumLevel() const noexcept
49+
LayerMask getActiveLayerMask() const noexcept
50+
{
51+
return LayerMask::span(0, NLayers - 1) & ~InactiveLayerMask;
52+
}
53+
54+
LayerMask getSeedingLayerMask() const noexcept
55+
{
56+
const auto activeLayers = getActiveLayerMask();
57+
return SeedingLayers.empty() ? activeLayers : (SeedingLayers & activeLayers);
58+
}
59+
60+
LayerMask getNonSeedingLayerMask() const noexcept
61+
{
62+
return ~(getSeedingLayerMask());
63+
}
64+
65+
int getNSeedingLayers() const noexcept
66+
{
67+
return getSeedingLayerMask().count();
68+
}
69+
70+
int getMinSeedingClusters() const noexcept
5071
{
5172
const int minClusters = MinTrackLength - (MaxHoles > 0 ? MaxHoles : 0);
52-
const int effectiveMinClusters = minClusters > constants::ClustersPerCell ? minClusters : constants::ClustersPerCell;
53-
return effectiveMinClusters - constants::ClustersPerCell + 1;
73+
const int minClustersWithCells = minClusters > constants::ClustersPerCell ? minClusters : constants::ClustersPerCell;
74+
const int nSeedingLayers = getNSeedingLayers();
75+
return minClustersWithCells < nSeedingLayers ? minClustersWithCells : nSeedingLayers;
76+
}
77+
78+
int CellMinimumLevel() const noexcept
79+
{
80+
return getMinSeedingClusters() - constants::ClustersPerCell + 1;
5481
}
55-
int NeighboursPerRoad() const noexcept { return NLayers - 3; }
56-
int CellsPerRoad() const noexcept { return NLayers - 2; }
57-
int TrackletsPerRoad() const noexcept { return NLayers - 1; }
82+
int NeighboursPerRoad() const noexcept { return getNSeedingLayers() - 3; }
83+
int CellsPerRoad() const noexcept { return getNSeedingLayers() - 2; }
84+
int TrackletsPerRoad() const noexcept { return getNSeedingLayers() - 1; }
5885
std::string asString() const;
5986

6087
IterationSteps PassFlags{IterationStep::FirstPass, IterationStep::RebuildClusterLUT};
@@ -76,6 +103,8 @@ struct TrackingParameters {
76103
int MinTrackLength = 7;
77104
int MaxHoles = 0;
78105
LayerMask HoleLayerMask = 0;
106+
LayerMask InactiveLayerMask = 0;
107+
LayerMask SeedingLayers = 0;
79108
float NSigmaCut = 5;
80109
float PVres = 1.e-2f;
81110
/// Trackleting cuts

‎Detectors/ITSMFT/ITS/tracking/include/ITStracking/Constants.h‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -29,6 +29,7 @@ constexpr bool DoTimeBenchmarks = true;
2929
constexpr bool SaveTimeBenchmarks = false;
3030
constexpr float Tolerance = 1e-12; // numerical tolerance
3131
constexpr int ClustersPerCell = 3; // number of clusters for a cell
32+
constexpr float MaxTrackSeedQ2Pt = 1.e3f; // maximum q/pt for track seeds
3233
constexpr int UnusedIndex = -1; // global unused flag
3334
constexpr float UnsetValue = -999.f; // global unset value
3435
constexpr float Radl = 9.36f; // Radiation length of Si [cm]

‎Detectors/ITSMFT/ITS/tracking/include/ITStracking/MathUtils.h‎

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -122,6 +122,9 @@ GPUhdi() constexpr float SqDiff(float x, float y)
122122

123123
GPUhdi() float MSangle(float mass, float p, float xX0)
124124
{
125+
if (xX0 <= 0.f) {
126+
return 0.f;
127+
}
125128
float beta = p / o2::gpu::CAMath::Hypot(mass, p);
126129
return 0.0136f * o2::gpu::CAMath::Sqrt(xX0) * (1.f + 0.038f * o2::gpu::CAMath::Log(xX0)) / (beta * p);
127130
}

‎Detectors/ITSMFT/ITS/tracking/include/ITStracking/TrackHelpers.h‎

Lines changed: 36 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -21,6 +21,7 @@
2121
#include "ITStracking/Cell.h"
2222
#include "ITStracking/Cluster.h"
2323
#include "ITStracking/Constants.h"
24+
#include "ITStracking/LayerMask.h"
2425
#include "ITStracking/MathUtils.h"
2526
#include "ITStracking/TrackITSInternal.h"
2627
#include "DetectorsBase/Propagator.h"
@@ -34,16 +35,46 @@ GPUhdi() bool isBetter(const int nClustersA, const float chi2A, const int nClust
3435
return (nClustersA > nClustersB) || (nClustersA == nClustersB && chi2A < chi2B);
3536
}
3637

37-
GPUhdi() bool isBetter(const o2::its::TrackITS& a, const o2::its::TrackITS& b)
38+
GPUhdi() bool isBetter(const auto& a, const auto& b)
3839
{
3940
return isBetter(a.getNumberOfClusters(), a.getChi2(), b.getNumberOfClusters(), b.getChi2());
4041
}
4142

4243
template <int NLayers>
43-
GPUhdi() bool isBetter(const o2::its::TrackITSInternal<NLayers>& a, const o2::its::TrackITSInternal<NLayers>& b)
44-
{
45-
return isBetter(a.getNumberOfClusters(), a.getChi2(), b.getNumberOfClusters(), b.getChi2());
46-
}
44+
struct TrackSeedSelector {
45+
float maxQ2Pt;
46+
float maxChi2;
47+
int maxHoles;
48+
int minTrackLength;
49+
LayerMask holeLayerMask;
50+
LayerMask nonSeedingLayerMask;
51+
52+
GPUhd() TrackSeedSelector(float maxQ2Pt, float maxChi2NDF, int startLevel, int maxHoles, int minTrackLength, LayerMask holeLayerMask, LayerMask nonSeedingLayerMask)
53+
: maxQ2Pt{maxQ2Pt}, maxChi2{maxChi2NDF * ((startLevel + 2) * 2 - 5)}, maxHoles{maxHoles}, minTrackLength{minTrackLength}, holeLayerMask{holeLayerMask}, nonSeedingLayerMask{nonSeedingLayerMask}
54+
{
55+
}
56+
57+
static GPUhdi() int getEffectiveTrackLength(LayerMask hitLayerMask, LayerMask excludedLayerMask)
58+
{
59+
if (hitLayerMask.empty()) {
60+
return 0;
61+
}
62+
return hitLayerMask.length() - (LayerMask::span(hitLayerMask.first(), hitLayerMask.last()) & excludedLayerMask).count();
63+
}
64+
65+
static GPUhdi() LayerMask getEffectiveHoleMask(LayerMask hitLayerMask, LayerMask excludedLayerMask)
66+
{
67+
return hitLayerMask.holeMask() & ~excludedLayerMask;
68+
}
69+
70+
GPUhd() bool operator()(const TrackSeed<NLayers>& seed) const
71+
{
72+
const auto hitLayerMask = seed.getHitLayerMask();
73+
return !(seed.getQ2Pt() > maxQ2Pt || seed.getChi2() > maxChi2) &&
74+
getEffectiveTrackLength(hitLayerMask, nonSeedingLayerMask) >= minTrackLength &&
75+
getEffectiveHoleMask(hitLayerMask, nonSeedingLayerMask).isAllowedHoleMask(maxHoles, holeLayerMask);
76+
}
77+
};
4778

4879
// Find the populated interior layer closest to the radial midpoint.
4980
// If no layer can be found, return constants::UnusedIndex.

0 commit comments

Comments
 (0)