Back to home page

Project CMSSW displayed by LXR

 
 

    


File indexing completed on 2024-08-07 23:11:43

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 minYsizeB1 = 25;
0371     static constexpr int minYsizeB2 = 15;
0372 
0373     static constexpr int nPairsMinimal = 33;
0374     static constexpr int nPairsFarForwards = nPairsMinimal + 8;  // include barrel "jumping" layer pairs
0375     static constexpr int nPairs = phase2PixelTopology::nPairs;   // include far forward layer pairs
0376 
0377     static constexpr int maxDYsize12 = 12;
0378     static constexpr int maxDYsize = 10;
0379     static constexpr int maxDYPred = 20;
0380 
0381     static constexpr uint16_t numberOfModules = phase2PixelTopology::numberOfModules;
0382 
0383     // 1000 bins < 1024 bins (10 bits) must be:
0384     // - < 32*32 (warpSize*warpSize for block prefix scan for CUDA)
0385     // - > number of columns (y) in any module. This is due to the fact
0386     //     that in pixel clustering we give for granted that in each
0387     //     bin we only have the pixel belonging to the same column.
0388     //     See RecoLocalTracker/SiPixelClusterizer/plugins/alpaka/PixelClustering.h#L325-L347
0389     static constexpr uint16_t clusterBinning = 1000;
0390     static constexpr uint16_t clusterBits = 10;
0391 
0392     static constexpr uint16_t numberOfModulesInBarrel = 756;
0393     static constexpr uint16_t numberOfModulesInLadder = 9;
0394     static constexpr uint16_t numberOfLaddersInBarrel = numberOfModulesInBarrel / numberOfModulesInLadder;
0395 
0396     static constexpr uint16_t firstEndcapPos = 4;
0397     static constexpr uint16_t firstEndcapNeg = 16;
0398 
0399     static constexpr int16_t xOffset = -1e4;  // not used actually, to suppress static analyzer warnings
0400 
0401     static constexpr char const *nameModifier = "Phase2";
0402 
0403     static constexpr uint32_t const *layerStart = phase2PixelTopology::layerStart;
0404     static constexpr float const *minz = phase2PixelTopology::minz;
0405     static constexpr float const *maxz = phase2PixelTopology::maxz;
0406     static constexpr float const *maxr = phase2PixelTopology::maxr;
0407 
0408     static constexpr uint8_t const *layerPairs = phase2PixelTopology::layerPairs;
0409     static constexpr int16_t const *phicuts = phase2PixelTopology::phicuts;
0410 
0411     static constexpr inline bool isBigPixX(uint16_t px) { return false; }
0412     static constexpr inline bool isBigPixY(uint16_t py) { return false; }
0413 
0414     static constexpr inline uint16_t localX(uint16_t px) { return px; }
0415     static constexpr inline uint16_t localY(uint16_t py) { return py; }
0416   };
0417 
0418   struct Phase1 {
0419     // types
0420     using hindex_type = uint32_t;  // FIXME from siPixelRecHitsHeterogeneousProduct
0421     using tindex_type = uint16_t;  // for tuples
0422     using cindex_type = uint32_t;  // for cells
0423 
0424     static constexpr uint32_t maxCellNeighbors = 36;
0425     static constexpr uint32_t maxCellTracks = 48;
0426     static constexpr uint32_t maxHitsOnTrack = 10;
0427     static constexpr uint32_t maxHitsOnTrackForFullFit = 6;
0428     static constexpr uint32_t avgHitsPerTrack = 5;
0429     static constexpr uint32_t maxCellsPerHit = 256;
0430     static constexpr uint32_t avgTracksPerHit = 6;
0431     static constexpr uint32_t maxNumberOfTuples = 32 * 1024;
0432     static constexpr uint32_t maxHitsForContainers = avgHitsPerTrack * maxNumberOfTuples;
0433     static constexpr uint32_t maxNumberOfDoublets = 512 * 1024;
0434     static constexpr uint32_t maxNumOfActiveDoublets = maxNumberOfDoublets / 8;
0435     static constexpr uint32_t maxNumberOfQuadruplets = maxNumberOfTuples;
0436     static constexpr uint32_t maxDepth = 6;
0437     static constexpr uint32_t numberOfLayers = 10;
0438 
0439     static constexpr uint32_t maxSizeCluster = 1023;
0440 
0441     static constexpr uint32_t getDoubletsFromHistoMaxBlockSize = 64;  // for both x and y
0442     static constexpr uint32_t getDoubletsFromHistoMinBlocksPerMP = 16;
0443 
0444     static constexpr uint16_t last_bpix1_detIndex = 96;
0445     static constexpr uint16_t last_bpix2_detIndex = 320;
0446     static constexpr uint16_t last_barrel_detIndex = 1184;
0447 
0448     static constexpr uint32_t maxPixInModule = 6000;
0449     static constexpr uint32_t maxNumClustersPerModules = phase1PixelTopology::maxNumClustersPerModules;
0450     static constexpr uint32_t maxHitsInModule = phase1PixelTopology::maxNumClustersPerModules;
0451 
0452     static constexpr float moduleLength = 6.7f;
0453     static constexpr float endcapCorrection = 1.5f;
0454 
0455     static constexpr float xerr_barrel_l1_def = 0.00200f;
0456     static constexpr float yerr_barrel_l1_def = 0.00210f;
0457     static constexpr float xerr_barrel_ln_def = 0.00200f;
0458     static constexpr float yerr_barrel_ln_def = 0.00210f;
0459     static constexpr float xerr_endcap_def = 0.0020f;
0460     static constexpr float yerr_endcap_def = 0.00210f;
0461 
0462     static constexpr float bigPixXCorrection = 1.0f;
0463     static constexpr float bigPixYCorrection = 8.0f;
0464 
0465     static constexpr float dzdrFact = 8 * 0.0285 / 0.015;  // from dz/dr to "DY"
0466 
0467     static constexpr int minYsizeB1 = 36;
0468     static constexpr int minYsizeB2 = 28;
0469 
0470     static constexpr int nPairsForQuadruplets = 13;                     // quadruplets require hits in all layers
0471     static constexpr int nPairsForTriplets = nPairsForQuadruplets + 2;  // include barrel "jumping" layer pairs
0472     static constexpr int nPairs = nPairsForTriplets + 4;                // include forward "jumping" layer pairs
0473 
0474     static constexpr int maxDYsize12 = 28;
0475     static constexpr int maxDYsize = 20;
0476     static constexpr int maxDYPred = 20;
0477 
0478     static constexpr uint16_t numberOfModules = phase1PixelTopology::numberOfModules;
0479 
0480     static constexpr uint16_t numRowsInRoc = 80;
0481     static constexpr uint16_t numColsInRoc = 52;
0482     static constexpr uint16_t lastRowInRoc = numRowsInRoc - 1;
0483     static constexpr uint16_t lastColInRoc = numColsInRoc - 1;
0484 
0485     static constexpr uint16_t numRowsInModule = 2 * numRowsInRoc;
0486     static constexpr uint16_t numColsInModule = 8 * numColsInRoc;
0487     static constexpr uint16_t lastRowInModule = numRowsInModule - 1;
0488     static constexpr uint16_t lastColInModule = numColsInModule - 1;
0489 
0490     // 418 bins < 512, 9 bits are enough
0491     static constexpr uint16_t clusterBinning = numColsInModule + 2;
0492     static constexpr uint16_t clusterBits = 9;
0493 
0494     static constexpr uint16_t numberOfModulesInBarrel = 1184;
0495     static constexpr uint16_t numberOfModulesInLadder = 8;
0496     static constexpr uint16_t numberOfLaddersInBarrel = numberOfModulesInBarrel / numberOfModulesInLadder;
0497 
0498     static constexpr uint16_t firstEndcapPos = 4;
0499     static constexpr uint16_t firstEndcapNeg = 7;
0500 
0501     static constexpr int16_t xOffset = -81;
0502 
0503     static constexpr char const *nameModifier = "";
0504 
0505     static constexpr uint32_t const *layerStart = phase1PixelTopology::layerStart;
0506     static constexpr float const *minz = phase1PixelTopology::minz;
0507     static constexpr float const *maxz = phase1PixelTopology::maxz;
0508     static constexpr float const *maxr = phase1PixelTopology::maxr;
0509 
0510     static constexpr uint8_t const *layerPairs = phase1PixelTopology::layerPairs;
0511     static constexpr int16_t const *phicuts = phase1PixelTopology::phicuts;
0512 
0513     static constexpr inline bool isEdgeX(uint16_t px) { return (px == 0) | (px == lastRowInModule); }
0514 
0515     static constexpr inline bool isEdgeY(uint16_t py) { return (py == 0) | (py == lastColInModule); }
0516 
0517     static constexpr inline uint16_t toRocX(uint16_t px) { return (px < numRowsInRoc) ? px : px - numRowsInRoc; }
0518 
0519     static constexpr inline uint16_t toRocY(uint16_t py) {
0520       auto roc = divu52(py);
0521       return py - 52 * roc;
0522     }
0523 
0524     static constexpr inline bool isBigPixX(uint16_t px) { return (px == 79) | (px == 80); }
0525     static constexpr inline bool isBigPixY(uint16_t py) {
0526       auto ly = toRocY(py);
0527       return (ly == 0) | (ly == lastColInRoc);
0528     }
0529 
0530     static constexpr inline uint16_t localX(uint16_t px) {
0531       auto shift = 0;
0532       if (px > lastRowInRoc)
0533         shift += 1;
0534       if (px > numRowsInRoc)
0535         shift += 1;
0536       return px + shift;
0537     }
0538 
0539     static constexpr inline uint16_t localY(uint16_t py) {
0540       auto roc = divu52(py);
0541       auto shift = 2 * roc;
0542       auto yInRoc = py - 52 * roc;
0543       if (yInRoc > 0)
0544         shift += 1;
0545       return py + shift;
0546     }
0547   };
0548 
0549   struct HIonPhase1 : public Phase1 {
0550     // Storing here the needed constants different w.r.t. pp Phase1 topology.
0551     // All the other defined by inheritance in the HIon topology struct.
0552 
0553     using tindex_type = uint32_t;  // for tuples
0554 
0555     static constexpr uint32_t maxCellNeighbors = 90;
0556     static constexpr uint32_t maxCellTracks = 90;
0557     static constexpr uint32_t maxNumberOfTuples = 256 * 1024;
0558     static constexpr uint32_t maxNumberOfDoublets = 6 * 512 * 1024;
0559     static constexpr uint32_t maxHitsForContainers = avgHitsPerTrack * maxNumberOfTuples;
0560     static constexpr uint32_t maxNumberOfQuadruplets = maxNumberOfTuples;
0561 
0562     static constexpr uint32_t maxPixInModule = 10000;
0563 
0564     static constexpr uint32_t maxNumOfActiveDoublets =
0565         maxNumberOfDoublets / 4;  // TODO need to think a better way to avoid this duplication
0566     static constexpr uint32_t maxCellsPerHit = 256;
0567 
0568     static constexpr uint32_t maxNumClustersPerModules = phase1HIonPixelTopology::maxNumClustersPerModules;
0569     static constexpr uint32_t maxHitsInModule = phase1HIonPixelTopology::maxNumClustersPerModules;
0570 
0571     static constexpr char const *nameModifier = "HIonPhase1";
0572   };
0573 
0574   template <typename T>
0575   using isPhase1Topology = typename std::enable_if<std::is_base_of<Phase1, T>::value>::type;
0576 
0577   template <typename T>
0578   using isPhase2Topology = typename std::enable_if<std::is_base_of<Phase2, T>::value>::type;
0579 
0580 }  // namespace pixelTopology
0581 
0582 #endif  // Geometry_CommonTopologies_SimplePixelTopology_h