Back to home page

Project CMSSW displayed by LXR

 
 

    


File indexing completed on 2025-02-09 23:41:52

0001 #ifndef Geometry_CommonTopologies_SimplePixelTopology_h
0002 #define Geometry_CommonTopologies_SimplePixelTopology_h
0003 
0004 #include <array>
0005 #include <cstdint>
0006 #include <type_traits>
0007 #include "FWCore/Utilities/interface/HostDeviceConstant.h"
0008 
0009 namespace pixelTopology {
0010 
0011   constexpr auto maxNumberOfLadders = 160;
0012   constexpr uint32_t maxLayers = 28;
0013 
0014   template <typename TrackerTraits>
0015   struct AverageGeometryT {
0016     //
0017     float ladderZ[TrackerTraits::numberOfLaddersInBarrel];
0018     float ladderX[TrackerTraits::numberOfLaddersInBarrel];
0019     float ladderY[TrackerTraits::numberOfLaddersInBarrel];
0020     float ladderR[TrackerTraits::numberOfLaddersInBarrel];
0021     float ladderMinZ[TrackerTraits::numberOfLaddersInBarrel];
0022     float ladderMaxZ[TrackerTraits::numberOfLaddersInBarrel];
0023     float endCapZ[2];  // just for pos and neg Layer1
0024   };
0025 
0026   constexpr int16_t phi0p05 = 522;  // round(521.52189...) = phi2short(0.05);
0027   constexpr int16_t phi0p06 = 626;  // round(625.82270...) = phi2short(0.06);
0028   constexpr int16_t phi0p07 = 730;  // round(730.12648...) = phi2short(0.07);
0029   constexpr int16_t phi0p09 = 900;
0030 
0031   constexpr uint16_t last_barrel_layer = 3;  // this is common between all the topologies
0032 
0033   template <class Function, std::size_t... Indices>
0034   constexpr auto map_to_array_helper(Function f, std::index_sequence<Indices...>)
0035       -> std::array<std::invoke_result_t<Function, std::size_t>, sizeof...(Indices)> {
0036     return {{f(Indices)...}};
0037   }
0038 
0039   template <int N, class Function>
0040   constexpr auto map_to_array(Function f) -> std::array<std::invoke_result_t<Function, std::size_t>, N> {
0041     return map_to_array_helper(f, std::make_index_sequence<N>{});
0042   }
0043 
0044   template <typename TrackerTraits>
0045   constexpr uint16_t findMaxModuleStride() {
0046     bool go = true;
0047     int n = 2;
0048     while (go) {
0049       for (uint8_t i = 1; i < TrackerTraits::numberOfLayers + 1; ++i) {
0050         if (TrackerTraits::layerStart[i] % n != 0) {
0051           go = false;
0052           break;
0053         }
0054       }
0055       if (!go)
0056         break;
0057       n *= 2;
0058     }
0059     return n / 2;
0060   }
0061 
0062   template <typename TrackerTraits>
0063   constexpr uint16_t maxModuleStride = findMaxModuleStride<TrackerTraits>();
0064 
0065   template <typename TrackerTraits>
0066   constexpr uint8_t findLayer(uint32_t detId, uint8_t sl = 0) {
0067     for (uint8_t i = sl; i < TrackerTraits::numberOfLayers + 1; ++i)
0068       if (detId < TrackerTraits::layerStart[i + 1])
0069         return i;
0070     return TrackerTraits::numberOfLayers + 1;
0071   }
0072 
0073   template <typename TrackerTraits>
0074   constexpr uint8_t findLayerFromCompact(uint32_t detId) {
0075     detId *= maxModuleStride<TrackerTraits>;
0076     for (uint8_t i = 0; i < TrackerTraits::numberOfLayers + 1; ++i)
0077       if (detId < TrackerTraits::layerStart[i + 1])
0078         return i;
0079     return TrackerTraits::numberOfLayers + 1;
0080   }
0081 
0082   template <typename TrackerTraits>
0083   constexpr uint32_t layerIndexSize = TrackerTraits::numberOfModules / maxModuleStride<TrackerTraits>;
0084 
0085   template <typename TrackerTraits>
0086 #ifdef __CUDA_ARCH__
0087   __device__
0088 #endif
0089       constexpr std::array<uint8_t, layerIndexSize<TrackerTraits>>
0090           layer = map_to_array<layerIndexSize<TrackerTraits>>(findLayerFromCompact<TrackerTraits>);
0091 
0092   template <typename TrackerTraits>
0093   constexpr uint8_t getLayer(uint32_t detId) {
0094     return layer<TrackerTraits>[detId / maxModuleStride<TrackerTraits>];
0095   }
0096 
0097   template <typename TrackerTraits>
0098   constexpr bool validateLayerIndex() {
0099     bool res = true;
0100     for (auto i = 0U; i < TrackerTraits::numberOfModules; ++i) {
0101       auto j = i / maxModuleStride<TrackerTraits>;
0102       res &= (layer<TrackerTraits>[j] < TrackerTraits::numberOfLayers);
0103       res &= (i >= TrackerTraits::layerStart[layer<TrackerTraits>[j]]);
0104       res &= (i < TrackerTraits::layerStart[layer<TrackerTraits>[j] + 1]);
0105     }
0106     return res;
0107   }
0108 
0109   template <typename TrackerTraits>
0110 #ifdef __CUDA_ARCH__
0111   __device__
0112 #endif
0113       constexpr inline uint32_t
0114       layerStart(uint32_t i) {
0115     return TrackerTraits::layerStart[i];
0116   }
0117 
0118   constexpr inline uint16_t divu52(uint16_t n) {
0119     n = n >> 2;
0120     uint16_t q = (n >> 1) + (n >> 4);
0121     q = q + (q >> 4) + (q >> 5);
0122     q = q >> 3;
0123     uint16_t r = n - q * 13;
0124     return q + ((r + 3) >> 4);
0125   }
0126 }  // namespace pixelTopology
0127 
0128 namespace phase1PixelTopology {
0129 
0130   using pixelTopology::phi0p05;
0131   using pixelTopology::phi0p06;
0132   using pixelTopology::phi0p07;
0133 
0134   constexpr uint32_t numberOfLayers = 28;
0135   constexpr int nPairs = 13 + 2 + 4;
0136   constexpr uint16_t numberOfModules = 1856;
0137 
0138   constexpr uint32_t maxNumClustersPerModules = 1024;
0139 
0140   constexpr uint32_t max_ladder_bpx0 = 12;
0141   constexpr uint32_t first_ladder_bpx0 = 0;
0142   constexpr float module_length_bpx0 = 6.7f;
0143   constexpr float module_tolerance_bpx0 = 0.4f;  // projection to cylinder is inaccurate on BPIX1
0144   constexpr uint32_t max_ladder_bpx4 = 64;
0145   constexpr uint32_t first_ladder_bpx4 = 84;
0146   constexpr float radius_even_ladder = 15.815f;
0147   constexpr float radius_odd_ladder = 16.146f;
0148   constexpr float module_length_bpx4 = 6.7f;
0149   constexpr float module_tolerance_bpx4 = 0.2f;
0150   constexpr float barrel_z_length = 26.f;
0151   constexpr float forward_z_begin = 32.f;
0152 
0153   HOST_DEVICE_CONSTANT uint8_t layerPairs[2 * nPairs] = {
0154       0, 1, 0, 4, 0, 7,              // BPIX1 (3)
0155       1, 2, 1, 4, 1, 7,              // BPIX2 (6)
0156       4, 5, 7, 8,                    // FPIX1 (8)
0157       2, 3, 2, 4, 2, 7, 5, 6, 8, 9,  // BPIX3 & FPIX2 (13)
0158       0, 2, 1, 3,                    // Jumping Barrel (15)
0159       0, 5, 0, 8,                    // Jumping Forward (BPIX1,FPIX2)
0160       4, 6, 7, 9                     // Jumping Forward (19)
0161   };
0162 
0163   HOST_DEVICE_CONSTANT int16_t phicuts[nPairs]{phi0p05,
0164                                                phi0p07,
0165                                                phi0p07,
0166                                                phi0p05,
0167                                                phi0p06,
0168                                                phi0p06,
0169                                                phi0p05,
0170                                                phi0p05,
0171                                                phi0p06,
0172                                                phi0p06,
0173                                                phi0p06,
0174                                                phi0p05,
0175                                                phi0p05,
0176                                                phi0p05,
0177                                                phi0p05,
0178                                                phi0p05,
0179                                                phi0p05,
0180                                                phi0p05,
0181                                                phi0p05};
0182   HOST_DEVICE_CONSTANT float minz[nPairs] = {
0183       -20., 0., -30., -22., 10., -30., -70., -70., -22., 15., -30, -70., -70., -20., -22., 0, -30., -70., -70.};
0184   HOST_DEVICE_CONSTANT float maxz[nPairs] = {
0185       20., 30., 0., 22., 30., -10., 70., 70., 22., 30., -15., 70., 70., 20., 22., 30., 0., 70., 70.};
0186   HOST_DEVICE_CONSTANT float maxr[nPairs] = {
0187       20., 9., 9., 20., 7., 7., 5., 5., 20., 6., 6., 5., 5., 20., 20., 9., 9., 9., 9.};
0188 
0189   static constexpr uint32_t layerStart[numberOfLayers + 1] = {0,
0190                                                               96,
0191                                                               320,
0192                                                               672,  // barrel
0193                                                               1184,
0194                                                               1296,
0195                                                               1408,  // positive endcap
0196                                                               1520,
0197                                                               1632,
0198                                                               1744,  // negative endcap
0199                                                               numberOfModules};
0200 }  // namespace phase1PixelTopology
0201 
0202 namespace phase2PixelTopology {
0203 
0204   using pixelTopology::phi0p05;
0205   using pixelTopology::phi0p06;
0206   using pixelTopology::phi0p07;
0207   using pixelTopology::phi0p09;
0208 
0209   constexpr uint32_t numberOfLayers = 28;
0210   constexpr int nPairs = 23 + 6 + 14 + 8 + 4;  // include far forward layer pairs
0211   constexpr uint16_t numberOfModules = 4000;
0212 
0213   constexpr uint32_t maxNumClustersPerModules = 1024;
0214 
0215   HOST_DEVICE_CONSTANT uint8_t layerPairs[2 * nPairs] = {
0216 
0217       0,  1,  0,  4,  0,  16,  // BPIX1 (3)
0218       1,  2,  1,  4,  1,  16,  // BPIX2 (6)
0219       2,  3,  2,  4,  2,  16,  // BPIX3 & Forward (9)
0220 
0221       4,  5,  5,  6,  6,  7,  7,  8,  8,  9,  9,  10, 10, 11,  // POS (16)
0222       16, 17, 17, 18, 18, 19, 19, 20, 20, 21, 21, 22, 22, 23,  // NEG (23)
0223 
0224       0,  2,  0,  5,  0,  17, 0,  6,  0,  18,  // BPIX1 Jump (28)
0225       1,  3,  1,  5,  1,  17, 1,  6,  1,  18,  // BPIX2 Jump (33)
0226 
0227       11, 12, 12, 13, 13, 14, 14, 15,  // Late POS (37)
0228       23, 24, 24, 25, 25, 26, 26, 27,  // Late NEG (41)
0229 
0230       4,  6,  5,  7,  6,  8,  7,  9,  8,  10, 9,  11, 10, 12,  // POS Jump (48)
0231       16, 18, 17, 19, 18, 20, 19, 21, 20, 22, 21, 23, 22, 24,  // NEG Jump (55)
0232   };
0233   HOST_DEVICE_CONSTANT uint32_t layerStart[numberOfLayers + 1] = {0,
0234                                                                   108,
0235                                                                   324,
0236                                                                   504,  // Barrel
0237                                                                   756,
0238                                                                   864,
0239                                                                   972,
0240                                                                   1080,
0241                                                                   1188,
0242                                                                   1296,
0243                                                                   1404,
0244                                                                   1512,
0245                                                                   1620,
0246                                                                   1796,
0247                                                                   1972,
0248                                                                   2148,  // Fp
0249                                                                   2324,
0250                                                                   2432,
0251                                                                   2540,
0252                                                                   2648,
0253                                                                   2756,
0254                                                                   2864,
0255                                                                   2972,
0256                                                                   3080,
0257                                                                   3188,
0258                                                                   3364,
0259                                                                   3540,
0260                                                                   3716,  // Np
0261                                                                   numberOfModules};
0262 
0263   HOST_DEVICE_CONSTANT int16_t phicuts[nPairs]{
0264       phi0p05, phi0p05, phi0p05, phi0p06, phi0p07, phi0p07, phi0p06, phi0p07, phi0p07, phi0p05, phi0p05,
0265       phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05,
0266       phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p07, phi0p07, phi0p07, phi0p07,
0267       phi0p07, phi0p07, phi0p07, phi0p07, phi0p07, phi0p07, phi0p07, phi0p07, phi0p07, phi0p07, phi0p07,
0268       phi0p07, phi0p07, phi0p07, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05, phi0p05};
0269 
0270   HOST_DEVICE_CONSTANT float minz[nPairs] = {
0271       -16.0, 4.0,   -22.0, -17.0, 6.0,   -22.0, -18.0, 11.0,  -22.0,  23.0,   30.0,   39.0,   50.0,   65.0,
0272       82.0,  109.0, -28.0, -35.0, -44.0, -55.0, -70.0, -87.0, -113.0, -16.,   7.0,    -22.0,  11.0,   -22.0,
0273       -17.0, 9.0,   -22.0, 13.0,  -22.0, 137.0, 173.0, 199.0, 229.0,  -142.0, -177.0, -203.0, -233.0, 23.0,
0274       30.0,  39.0,  50.0,  65.0,  82.0,  109.0, -28.0, -35.0, -44.0,  -55.0,  -70.0,  -87.0,  -113.0};
0275 
0276   HOST_DEVICE_CONSTANT float maxz[nPairs] = {
0277 
0278       17.0, 22.0,  -4.0,  17.0,  22.0,  -6.0,  18.0,  22.0,  -11.0,  28.0,   35.0,   44.0,   55.0,   70.0,
0279       87.0, 113.0, -23.0, -30.0, -39.0, -50.0, -65.0, -82.0, -109.0, 17.0,   22.0,   -7.0,   22.0,   -10.0,
0280       17.0, 22.0,  -9.0,  22.0,  -13.0, 142.0, 177.0, 203.0, 233.0,  -137.0, -173.0, -199.0, -229.0, 28.0,
0281       35.0, 44.0,  55.0,  70.0,  87.0,  113.0, -23.0, -30.0, -39.0,  -50.0,  -65.0,  -82.0,  -109.0};
0282 
0283   HOST_DEVICE_CONSTANT float maxr[nPairs] = {5.0, 5.0, 5.0, 7.0, 8.0, 8.0,  7.0, 7.0, 7.0, 6.0, 6.0, 6.0, 6.0, 5.0,
0284                                              6.0, 5.0, 6.0, 6.0, 6.0, 6.0,  5.0, 6.0, 5.0, 5.0, 5.0, 5.0, 5.0, 5.0,
0285                                              5.0, 8.0, 8.0, 8.0, 8.0, 6.0,  5.0, 5.0, 5.0, 6.0, 5.0, 5.0, 5.0, 9.0,
0286                                              9.0, 9.0, 8.0, 8.0, 8.0, 11.0, 9.0, 9.0, 9.0, 8.0, 8.0, 8.0, 11.0};
0287 }  // namespace phase2PixelTopology
0288 
0289 namespace phase1HIonPixelTopology {
0290   // Storing here the needed constants different w.r.t. pp Phase1 topology.
0291   // All the other defined by inheritance in the HIon topology struct.
0292   using pixelTopology::phi0p09;
0293 
0294   constexpr uint32_t maxNumClustersPerModules = 2048;
0295 
0296   HOST_DEVICE_CONSTANT int16_t phicuts[phase1PixelTopology::nPairs]{phi0p09,
0297                                                                     phi0p09,
0298                                                                     phi0p09,
0299                                                                     phi0p09,
0300                                                                     phi0p09,
0301                                                                     phi0p09,
0302                                                                     phi0p09,
0303                                                                     phi0p09,
0304                                                                     phi0p09,
0305                                                                     phi0p09,
0306                                                                     phi0p09,
0307                                                                     phi0p09,
0308                                                                     phi0p09,
0309                                                                     phi0p09,
0310                                                                     phi0p09,
0311                                                                     phi0p09,
0312                                                                     phi0p09,
0313                                                                     phi0p09,
0314                                                                     phi0p09};
0315 
0316 }  // namespace phase1HIonPixelTopology
0317 
0318 namespace pixelTopology {
0319 
0320   struct Phase2 {
0321     // types
0322     using hindex_type = uint32_t;  // FIXME from siPixelRecHitsHeterogeneousProduct
0323     using tindex_type = uint32_t;  // for tuples
0324     using cindex_type = uint32_t;  // for cells
0325 
0326     static constexpr uint32_t maxCellNeighbors = 64;
0327     static constexpr uint32_t maxCellTracks = 302;
0328     static constexpr uint32_t maxHitsOnTrack = 15;
0329     static constexpr uint32_t maxHitsOnTrackForFullFit = 6;
0330     static constexpr uint32_t avgHitsPerTrack = 7;
0331     static constexpr uint32_t maxCellsPerHit = 256;
0332     static constexpr uint32_t avgTracksPerHit = 10;
0333     static constexpr uint32_t maxNumberOfTuples = 256 * 1024;
0334     // this is well above thanks to maxNumberOfTuples
0335     static constexpr uint32_t maxHitsForContainers = avgHitsPerTrack * maxNumberOfTuples;
0336     static constexpr uint32_t maxNumberOfDoublets = 5 * 512 * 1024;
0337     static constexpr uint32_t maxNumOfActiveDoublets = maxNumberOfDoublets / 8;
0338     static constexpr uint32_t maxNumberOfQuadruplets = maxNumberOfTuples;
0339     static constexpr uint32_t maxDepth = 12;
0340     static constexpr uint32_t numberOfLayers = 28;
0341 
0342     static constexpr uint32_t maxSizeCluster = 2047;
0343 
0344     static constexpr uint32_t getDoubletsFromHistoMaxBlockSize = 64;  // for both x and y
0345     static constexpr uint32_t getDoubletsFromHistoMinBlocksPerMP = 16;
0346 
0347     static constexpr uint16_t last_bpix1_detIndex = 108;
0348     static constexpr uint16_t last_bpix2_detIndex = 324;
0349     static constexpr uint16_t last_barrel_detIndex = 504;
0350 
0351     static constexpr uint32_t maxPixInModule = 6000;
0352     static constexpr uint32_t maxNumClustersPerModules = phase2PixelTopology::maxNumClustersPerModules;
0353     static constexpr uint32_t maxHitsInModule = phase2PixelTopology::maxNumClustersPerModules;
0354 
0355     static constexpr float moduleLength = 4.345f;
0356     static constexpr float endcapCorrection = 0.0f;
0357 
0358     static constexpr float xerr_barrel_l1_def = 0.00035f;
0359     static constexpr float yerr_barrel_l1_def = 0.00125f;
0360     static constexpr float xerr_barrel_ln_def = 0.00035f;
0361     static constexpr float yerr_barrel_ln_def = 0.00125f;
0362     static constexpr float xerr_endcap_def = 0.00060f;
0363     static constexpr float yerr_endcap_def = 0.00180f;
0364 
0365     static constexpr float bigPixXCorrection = 0.0f;
0366     static constexpr float bigPixYCorrection = 0.0f;
0367 
0368     static constexpr float dzdrFact = 8 * 0.0285 / 0.015;  // from dz/dr to "DY"
0369 
0370     static constexpr int nPairsMinimal = 33;
0371     static constexpr int nPairsFarForwards = nPairsMinimal + 8;  // include barrel "jumping" layer pairs
0372     static constexpr int nPairs = phase2PixelTopology::nPairs;   // include far forward layer pairs
0373 
0374     static constexpr int maxDYsize12 = 12;
0375     static constexpr int maxDYsize = 10;
0376     static constexpr int maxDYPred = 20;
0377 
0378     static constexpr uint16_t numberOfModules = phase2PixelTopology::numberOfModules;
0379 
0380     // 1000 bins < 1024 bins (10 bits) must be:
0381     // - < 32*32 (warpSize*warpSize for block prefix scan for CUDA)
0382     // - > number of columns (y) in any module. This is due to the fact
0383     //     that in pixel clustering we give for granted that in each
0384     //     bin we only have the pixel belonging to the same column.
0385     //     See RecoLocalTracker/SiPixelClusterizer/plugins/alpaka/PixelClustering.h#L325-L347
0386     static constexpr uint16_t clusterBinning = 1000;
0387     static constexpr uint16_t clusterBits = 10;
0388 
0389     static constexpr uint16_t numberOfModulesInBarrel = 756;
0390     static constexpr uint16_t numberOfModulesInLadder = 9;
0391     static constexpr uint16_t numberOfLaddersInBarrel = numberOfModulesInBarrel / numberOfModulesInLadder;
0392 
0393     static constexpr uint16_t firstEndcapPos = 4;
0394     static constexpr uint16_t firstEndcapNeg = 16;
0395 
0396     static constexpr int16_t xOffset = -1e4;  // not used actually, to suppress static analyzer warnings
0397 
0398     static constexpr char const *nameModifier = "Phase2";
0399 
0400     static constexpr uint32_t const *layerStart = phase2PixelTopology::layerStart;
0401     static constexpr float const *minz = phase2PixelTopology::minz;
0402     static constexpr float const *maxz = phase2PixelTopology::maxz;
0403     static constexpr float const *maxr = phase2PixelTopology::maxr;
0404 
0405     static constexpr uint8_t const *layerPairs = phase2PixelTopology::layerPairs;
0406     static constexpr int16_t const *phicuts = phase2PixelTopology::phicuts;
0407 
0408     static constexpr inline bool isBigPixX(uint16_t px) { return false; }
0409     static constexpr inline bool isBigPixY(uint16_t py) { return false; }
0410 
0411     static constexpr inline uint16_t localX(uint16_t px) { return px; }
0412     static constexpr inline uint16_t localY(uint16_t py) { return py; }
0413   };
0414 
0415   struct Phase1 {
0416     // types
0417     using hindex_type = uint32_t;  // FIXME from siPixelRecHitsHeterogeneousProduct
0418     using tindex_type = uint16_t;  // for tuples
0419     using cindex_type = uint32_t;  // for cells
0420 
0421     static constexpr uint32_t maxCellNeighbors = 36;
0422     static constexpr uint32_t maxCellTracks = 48;
0423     static constexpr uint32_t maxHitsOnTrack = 10;
0424     static constexpr uint32_t maxHitsOnTrackForFullFit = 6;
0425     static constexpr uint32_t avgHitsPerTrack = 5;
0426     static constexpr uint32_t maxCellsPerHit = 256;
0427     static constexpr uint32_t avgTracksPerHit = 6;
0428     static constexpr uint32_t maxNumberOfTuples = 32 * 1024;
0429     static constexpr uint32_t maxHitsForContainers = avgHitsPerTrack * maxNumberOfTuples;
0430     static constexpr uint32_t maxNumberOfDoublets = 512 * 1024;
0431     static constexpr uint32_t maxNumOfActiveDoublets = maxNumberOfDoublets / 8;
0432     static constexpr uint32_t maxNumberOfQuadruplets = maxNumberOfTuples;
0433     static constexpr uint32_t maxDepth = 6;
0434     static constexpr uint32_t numberOfLayers = 10;
0435 
0436     static constexpr uint32_t maxSizeCluster = 1023;
0437 
0438     static constexpr uint32_t getDoubletsFromHistoMaxBlockSize = 64;  // for both x and y
0439     static constexpr uint32_t getDoubletsFromHistoMinBlocksPerMP = 16;
0440 
0441     static constexpr uint16_t last_bpix1_detIndex = 96;
0442     static constexpr uint16_t last_bpix2_detIndex = 320;
0443     static constexpr uint16_t last_barrel_detIndex = 1184;
0444 
0445     static constexpr uint32_t maxPixInModule = 6000;
0446     static constexpr uint32_t maxNumClustersPerModules = phase1PixelTopology::maxNumClustersPerModules;
0447     static constexpr uint32_t maxHitsInModule = phase1PixelTopology::maxNumClustersPerModules;
0448 
0449     static constexpr float moduleLength = 6.7f;
0450     static constexpr float endcapCorrection = 1.5f;
0451 
0452     static constexpr float xerr_barrel_l1_def = 0.00200f;
0453     static constexpr float yerr_barrel_l1_def = 0.00210f;
0454     static constexpr float xerr_barrel_ln_def = 0.00200f;
0455     static constexpr float yerr_barrel_ln_def = 0.00210f;
0456     static constexpr float xerr_endcap_def = 0.0020f;
0457     static constexpr float yerr_endcap_def = 0.00210f;
0458 
0459     static constexpr float bigPixXCorrection = 1.0f;
0460     static constexpr float bigPixYCorrection = 8.0f;
0461 
0462     static constexpr float dzdrFact = 8 * 0.0285 / 0.015;  // from dz/dr to "DY"
0463 
0464     static constexpr int nPairsForQuadruplets = 13;                     // quadruplets require hits in all layers
0465     static constexpr int nPairsForTriplets = nPairsForQuadruplets + 2;  // include barrel "jumping" layer pairs
0466     static constexpr int nPairs = nPairsForTriplets + 4;                // include forward "jumping" layer pairs
0467 
0468     static constexpr int maxDYsize12 = 28;
0469     static constexpr int maxDYsize = 20;
0470     static constexpr int maxDYPred = 20;
0471 
0472     static constexpr uint16_t numberOfModules = phase1PixelTopology::numberOfModules;
0473 
0474     static constexpr uint16_t numRowsInRoc = 80;
0475     static constexpr uint16_t numColsInRoc = 52;
0476     static constexpr uint16_t lastRowInRoc = numRowsInRoc - 1;
0477     static constexpr uint16_t lastColInRoc = numColsInRoc - 1;
0478 
0479     static constexpr uint16_t numRowsInModule = 2 * numRowsInRoc;
0480     static constexpr uint16_t numColsInModule = 8 * numColsInRoc;
0481     static constexpr uint16_t lastRowInModule = numRowsInModule - 1;
0482     static constexpr uint16_t lastColInModule = numColsInModule - 1;
0483 
0484     // 418 bins < 512, 9 bits are enough
0485     static constexpr uint16_t clusterBinning = numColsInModule + 2;
0486     static constexpr uint16_t clusterBits = 9;
0487 
0488     static constexpr uint16_t numberOfModulesInBarrel = 1184;
0489     static constexpr uint16_t numberOfModulesInLadder = 8;
0490     static constexpr uint16_t numberOfLaddersInBarrel = numberOfModulesInBarrel / numberOfModulesInLadder;
0491 
0492     static constexpr uint16_t firstEndcapPos = 4;
0493     static constexpr uint16_t firstEndcapNeg = 7;
0494 
0495     static constexpr int16_t xOffset = -81;
0496 
0497     static constexpr char const *nameModifier = "";
0498 
0499     static constexpr uint32_t const *layerStart = phase1PixelTopology::layerStart;
0500     static constexpr float const *minz = phase1PixelTopology::minz;
0501     static constexpr float const *maxz = phase1PixelTopology::maxz;
0502     static constexpr float const *maxr = phase1PixelTopology::maxr;
0503 
0504     static constexpr uint8_t const *layerPairs = phase1PixelTopology::layerPairs;
0505     static constexpr int16_t const *phicuts = phase1PixelTopology::phicuts;
0506 
0507     static constexpr inline bool isEdgeX(uint16_t px) { return (px == 0) | (px == lastRowInModule); }
0508 
0509     static constexpr inline bool isEdgeY(uint16_t py) { return (py == 0) | (py == lastColInModule); }
0510 
0511     static constexpr inline uint16_t toRocX(uint16_t px) { return (px < numRowsInRoc) ? px : px - numRowsInRoc; }
0512 
0513     static constexpr inline uint16_t toRocY(uint16_t py) {
0514       auto roc = divu52(py);
0515       return py - 52 * roc;
0516     }
0517 
0518     static constexpr inline bool isBigPixX(uint16_t px) { return (px == 79) | (px == 80); }
0519     static constexpr inline bool isBigPixY(uint16_t py) {
0520       auto ly = toRocY(py);
0521       return (ly == 0) | (ly == lastColInRoc);
0522     }
0523 
0524     static constexpr inline uint16_t localX(uint16_t px) {
0525       auto shift = 0;
0526       if (px > lastRowInRoc)
0527         shift += 1;
0528       if (px > numRowsInRoc)
0529         shift += 1;
0530       return px + shift;
0531     }
0532 
0533     static constexpr inline uint16_t localY(uint16_t py) {
0534       auto roc = divu52(py);
0535       auto shift = 2 * roc;
0536       auto yInRoc = py - 52 * roc;
0537       if (yInRoc > 0)
0538         shift += 1;
0539       return py + shift;
0540     }
0541   };
0542 
0543   struct HIonPhase1 : public Phase1 {
0544     // Storing here the needed constants different w.r.t. pp Phase1 topology.
0545     // All the other defined by inheritance in the HIon topology struct.
0546 
0547     using tindex_type = uint32_t;  // for tuples
0548 
0549     static constexpr uint32_t maxCellNeighbors = 90;
0550     static constexpr uint32_t maxCellTracks = 90;
0551     static constexpr uint32_t maxNumberOfTuples = 256 * 1024;
0552     static constexpr uint32_t maxNumberOfDoublets = 6 * 512 * 1024;
0553     static constexpr uint32_t maxHitsForContainers = avgHitsPerTrack * maxNumberOfTuples;
0554     static constexpr uint32_t maxNumberOfQuadruplets = maxNumberOfTuples;
0555 
0556     static constexpr uint32_t maxPixInModule = 10000;
0557 
0558     static constexpr uint32_t maxNumOfActiveDoublets =
0559         maxNumberOfDoublets / 4;  // TODO need to think a better way to avoid this duplication
0560     static constexpr uint32_t maxCellsPerHit = 256;
0561 
0562     static constexpr uint32_t maxNumClustersPerModules = phase1HIonPixelTopology::maxNumClustersPerModules;
0563     static constexpr uint32_t maxHitsInModule = phase1HIonPixelTopology::maxNumClustersPerModules;
0564 
0565     static constexpr char const *nameModifier = "HIonPhase1";
0566   };
0567 
0568   template <typename T>
0569   using isPhase1Topology = typename std::enable_if<std::is_base_of<Phase1, T>::value>::type;
0570 
0571   template <typename T>
0572   using isPhase2Topology = typename std::enable_if<std::is_base_of<Phase2, T>::value>::type;
0573 
0574 }  // namespace pixelTopology
0575 
0576 #endif  // Geometry_CommonTopologies_SimplePixelTopology_h