Skip to content
Snippets Groups Projects

Allen RICH decoding

Merged Daniel Hugo Campora Perez requested to merge cnca_rich_decoding into master

This MR introduces RICH decoding, concretely producing RichSmartIDs from RICH1 and RICH2 raw banks.

  • Created algorithm rich_decoding_t, which performs the decoding by calculating the number of hits, prefix sum and finally decoding the hits.
  • Added simplified copies of RichSmartID, PackedFrameSizes, Tel40CableMapping and PDMDBDecodeMapping adjusted to the requirements of GPU computing, namely no dynamic memory allocation, no std::string elements and __host__ __device__ attributes.
  • Created Allen::RawEvent and Allen::RawBank which replace CaloRawEvent and CaloRawBank. Ideally this should be used in other subdetectors as well due to the raw bank format being mostly identical.
  • Added configurations rich and hlt1_pp_rich_no_ut.
  • Added throughput test of hlt1_pp_rich_no_ut.
  • Added two geometry files corresponding to Tel40CableMapping and PDMDBDecodeMapping, alongside constants and the necessary scaffolding.

Goes with LHCb!4001 (merged) and Moore!2171 (merged)

Done in collaboration with @caschbur .

Edited by Daniel Hugo Campora Perez

Merge request reports

Loading
Loading

Activity

Filter activity
  • Approvals
  • Assignees & reviewers
  • Comments (from bots)
  • Comments (from users)
  • Commits & branches
  • Edits
  • Labels
  • Lock status
  • Mentions
  • Merge request status
  • Tracking
  • added 1 commit

    Compare with previous version

  • added 30 commits

    • d00bff51...190c1589 - 2 commits from branch master
    • 0f9be3ce - Test Rich 1 decoding.
    • 54b67cf5 - Allow compilation.
    • b22420e9 - Message for loading bin files.
    • dbfc47fe - First attempt at decoding.
    • 67103b8e - Added kernel to calculate number of hits.
    • 580c28c0 - Decoding with prefix sum and calculate hits.
    • cab045b7 - First seemingly working correctly decoding of RICH.
    • f7e211e6 - Added the necessary declaration specifiers to make the code compile with CUDA
    • ef3cd531 - Fix formatting.
    • 9960499f - Refactor RawEvent and RawBank to a common place, used by RICH and Calo.
    • b9a2ae4b - Now runs correctly on GPU and generates the correct ids
    • bc0e34dd - Add rich geometry files, point to those in init.
    • 794997d1 - Use consumers for Rich geometries.
    • f4213c36 - Move temporary to allen buffer.
    • 674d8300 - Compile with Gaudi-Allen.
    • ef406d18 - Solve warning.
    • b548afef - Add RICH2 decoding.
    • da04c0f6 - Use class for RichSmartID.
    • e45e99f7 - Add host device attributes to RichSmartID constructors.
    • eb3b2af5 - Added Rich no ut configuration.
    • 4c12d46d - Fix CaloDecode logic to using 32-bit raw bank pointer.
    • c418c2ac - Add hlt1_pp_rich_no_ut throughput job.
    • 144fbe0c - Fixed formatting
    • a621865a - Add Apache2 license.
    • 027a39d7 - Uncomment build sequences.
    • 1e6d6b77 - Modify test to run with correct geometry.
    • 407afeff - Move hlt1 rich configuration.
    • c0c3738a - Refactor code.

    Compare with previous version

  • added 1 commit

    • c3aa3a67 - Add property with number of threads in block in RICH decoding.

    Compare with previous version

  • added 1 commit

    • 3c3ab19c - Use [] instead of at. Add verbosity print.

    Compare with previous version

  • added 35 commits

    • 3c3ab19c...3d32e14a - 5 commits from branch master
    • db054a7c - Test Rich 1 decoding.
    • e6f42c27 - Allow compilation.
    • 66db852f - Message for loading bin files.
    • d25cb346 - First attempt at decoding.
    • ccf2c7a4 - Added kernel to calculate number of hits.
    • f1020b75 - Decoding with prefix sum and calculate hits.
    • 223aaca0 - First seemingly working correctly decoding of RICH.
    • e10e6b4c - Added the necessary declaration specifiers to make the code compile with CUDA
    • 1b941f38 - Fix formatting.
    • c99714ed - Refactor RawEvent and RawBank to a common place, used by RICH and Calo.
    • 84fd7fcd - Now runs correctly on GPU and generates the correct ids
    • 9cdfd953 - Add rich geometry files, point to those in init.
    • 61f36d8a - Use consumers for Rich geometries.
    • 63ab03f7 - Move temporary to allen buffer.
    • fb1eb560 - Compile with Gaudi-Allen.
    • 54424b39 - Solve warning.
    • e98d6c1a - Add RICH2 decoding.
    • 359c3698 - Use class for RichSmartID.
    • d5954982 - Add host device attributes to RichSmartID constructors.
    • 176ad909 - Added Rich no ut configuration.
    • e0222a4e - Fix CaloDecode logic to using 32-bit raw bank pointer.
    • abbd08bd - Add hlt1_pp_rich_no_ut throughput job.
    • 693d3c26 - Fixed formatting
    • 078fc5e9 - Add Apache2 license.
    • ac56c6d1 - Uncomment build sequences.
    • ad6ddecd - Modify test to run with correct geometry.
    • 1d1e28cd - Move hlt1 rich configuration.
    • 6f0e8546 - Refactor code.
    • 1dcfaf80 - Add property with number of threads in block in RICH decoding.
    • 78411a38 - Use [] instead of at. Add verbosity print.

    Compare with previous version

  • added 1 commit

    Compare with previous version

  • Daniel Hugo Campora Perez changed the description

    changed the description

  • Loading
  • Loading
  • Loading
  • Loading
  • Loading
  • Loading
  • Loading
  • Loading
  • Loading
  • Loading
  • Please register or sign in to reply
    Loading