- reorganize `SiPixelRawToDigi` as `SiPixelRawToDigiHeterogeneous` using `HeterogeneousEDProducer`
      - output a `HeterogeneousEvent`
      - use `PixelThresholdClusterizer`
      - add `SiPixelDigiHeterogeneousConverter`
      - make cabling and gain transfers asynchronous
  - reorganize `SiPixelRecHits` as `SiPixelRecHitHeterogeneous`
  - move `PixelThresholdClusterizer` (back?) to interface+src in order to use it outside of RecoLocalTracker/SiPixelClusterizer
  - replace __host__ __device__ with constexpr to avoid weird compilation failures
  - split clusters to their own converter

Adds a prototype for dealing with EventSetup data on GPUs. The prototype is applied to the ES data used by Raw2Cluster (cabling map etc, gains) and RecHits (CPE).

Now it is the `ESProduct` who owns the GPU memory. Currently each of the affected `ESProducts` have a method `getGPUProductAsync(cuda::stream_t<>&)` that will allocate the memory on the current GPU device and transfer the data there asynchronously, if the data is not there yet. The functionality of bookkeeping which devices have the data already, and necessary synchronization between multiple threads (only one thread may do the transfer per device) are abstracted to a helper template in `HeterogeneousCore/CUDACore/interface/CUDAESProduct.h`.

Technical changes:
  - `EventSetup`-based implementation for GPU cabling map, gains, etc
  - add support for multiple devices to `PixelCPEFast`
  - abstract the `EeventSetup` GPU transfer
  - move `malloc` and transfer to the lambda
  - move `cudaFree` outside of the `nullptr` check
  - move files (back) to the plusing directory
  - rename `siPixelDigisHeterogeneous` to `siPixelClustersHeterogeneous`

Provide a mechanism for a chain of modules to share a resource, that can be e.g. CUDA device memory or a CUDA stream.
Minimize data movements between the CPU and the device, and support multiple devices.
Allow the same job configuration to be used on all hardware combinations.

See HeterogeneousCore/CUDACore/README.md for a more detailed description and examples.

* Add DQM for pixel vertices

* Add pT>0.9GeV pixel track collections to MTV

* Add dzPV0p1, Pt0to1, Pt1 variants of pixel track DQM

#335)

Reimplementation of the orignal cpu-based ECAL multifit to run on Nvidia GPUs with CUDA,
for an HLT-like configuration, where regression matrices are fixed at 10x10 (no dynamic pedestal, etc...).

The main computation type is float by default, configurable at compile time, and the minimization (Cholesky + solvers, etc...) is implemented using using the Eigen library.
The timing computation is implemented, but is not run by default at HLT.

Tha implementation:
  - uses one CUDA stream per EDM stream;
  - the EventSetup conditions are updated on the GPU only they the change, via the CUDAESProduct mechanism;
  - only the per-event data is transferred for each event;
  - the results are optionally copied back to the host and synchronised by the produce() method.

A simple tool is available to validate cpu vs gpu results.

Known issues and to do list:
  - add a module to convert from new format to the legacy format;
  - make use of the CUDAService framework for the device selection, stream handling and memory allocation;
  - investigate some instabilities in the Cholesky decomposition vs original cpu version, specifically [for fnnls](https://github.com/cms-patatrack/cmssw/pull/335/files#diff-ed446c49128ac6dc6f45eeebab079613R70) causing rare, but noticeable discrepancies between cpu and gpu versions.

Extend the ECAL validation tool:
  - skip events with a mismatched number of hits, instead of aborting;
  - add 2D plots for energy and chi2 differences;
  - save all plots also in PDF format.

Move it to the .../bin directory and rename it to makeEcalGpuValidationPlots.

#361)

…389)

Replace cuda::stream_t<> with cudaStream_t in client code
Replace cuda::event_t with cudaEvent_t in the client code
Clean up BuildFiles

Rename the cudautils namespace to cms::cuda or cms::cudatest, and drop the CUDA prefix from the symbols defined there.

Always record and query the CUDA event, to minimize need for error checking in CUDAScopedContextProduce destructor.

Add comments to highlight the pieces in CachingDeviceAllocator that have been changed wrt. cub.

Various other updates and clean up:
  - enable CUDA for compute capability 3.5.
  - clean up CUDAService, CUDA tests and plugins.
  - add CUDA existence protections to BuildFiles.
  - mark thread-safe static variables with CMS_THREAD_SAFE.

Implement the ECAL unpacker running on GPU from RAW data, and update the downstream modules to use the new data format.

Fix compilation warnings, remove commented out code, and apply code formatting rules.

Move the `ecalDigis` and `ecalMultiFitUncalibRecHit` modules to
separate Tasks.

Implement an `ecalOnlyLocalRecoTask` based on the `ecalLocalRecoTask`,
without the trigger primitive-related modules.

Let EcalCPUDigisProducer optionally produce dummy ECAL integrity collections.

Implement the use of the "gpu" process modifier for ECAL-only reconstruction
workflows:
  - 10824.512: TTbar, 2018 realistic conditions, no pileup;
  - 11634.512: TTbar, 2021 realistic conditions, no pileup.

…rameter (#463)

customizeEcalOnlyForProfilingGPUOnly:
  Customise the ECAL-only reconstruction to run on GPU.
  Currently, this means running only the unpacker and multifit,
  up to the uncalbrated rechits.

customizeEcalOnlyForProfilingGPUWithHostCopy:
  Customise the ECAL-only reconstruction to run on GPU, and copy
  the data to the host.
  Currently, this means running only the unpacker and multifit,
  up to the uncalbrated rechits.

customizeEcalOnlyForProfiling:
  Customise the ECAL-only reconstruction to run on GPU, copy the
  data to the host, and convert them to legacy format.
  Currently, this means running only the unpacker and multifit,
  up to the uncalbrated rechits, on the GPU - and the rechits
  producer on the CPU.
  The same customisation can be also used on the CPU workflow,
  running up to the rechits on CPU.

Include standalone executables for the validation of ECAL uncalibrated
and calibrated rechits.

EcalLinearCorrectionsGPU:
  - add missing cudaFree()

EcalRecHitBuilderKernels:
  - fix warnings due to unused variables
  - fix incorrectly skipping some channels due to the use of "return" in the kernel loop
  - add default values for dead/invalid channels

Extend the ECAL HLT cutomisation to run the ECAL rechit producer on GPU.

Add an edm::LogError in case of too many channels for the rechits and uncalibrated rechits producers.

Move ECAL and HCAL CUDA code to plugins.
General cleanup: remove unused code, apply clang-format and various include changes.
Fix product labels for HCAL rechits on CPU.

Co-authored-by: Andrea Bocci <[email protected]>

Comment instead of removing because there is commented code using eps_diff

Remove dictionary definitions for classes already defined in CUDADataFormats/StdDictionaries.

Co-authored-by: Andrea Bocci <[email protected]>

…sw#30350)

Backport cms-sw#30105: add ECAL-only workflows for data.
Backport cms-sw#30136: add HCAL-only workflows for MC and data.

Use caching allocators for host and device CUDA memory.
Use dedicated ESProducers to make part of the modules' configuration available on all GPUs.
Rename hcal and hcal::common namespaces to to calo::common.

Add missing ESProducers for ECAL and HCAL GPU modules: add to the
offline workflows and to the HLT customisations the ESProducers required
to complement the configuration of the ECAL and HCAL GPU modules.

Fix fillDescriptions() for EcalRecHitParametersGPUESProducer.

)

Fix memory allocation issues.

Apply come code clean up:
  - remove outdated comments;
  - replace MYMALLOC macro with a lambda;
  - reuse named values from EcalDataFrame.

Move duplicated Eigen code to a common file, and use it for both ECAL and HCAL.
Move HCAL general reconstruction code from the hcal::multifit to the hcal::reconstruction namespace.

Minor bug fixes:
  - fix a typo in EventFilter/EcalRawToDigi/plugins/BuildFile.xml .

Clean up:
  - remove obsolete ArrayShadow class;
  - remove obsolete profiling functions.

Move multifit/MAHI common code to DataFormats/CaloRecHit/interface/MultifitComputations.h .
Improve naming and description of fnnls parameters.
Use Eigen preprocessor symbols instead of explicit CUDA keywords, and CUDA preprocessor symbols to protect CUDA-only functions.

Co-authored-by: Andrea Bocci <[email protected]>

Adjust include guards.
Adjust comments on preprocessor macros.

Move HCAL constants to a separate file, and update CPU and GPU code accordingly.

Add an infinite-IOV record for GPU modules configuration: as an interim approach,
some modules are using the ventSetup approach to copy complex configurations to
the GPUs; the new "JobConfigurationGPURecord" should be used for those, to make
it easier both to highlight the intent, and clean up the client code when a
better solution is found.

Replace ECAL and HCAL job configuration records with JobConfigurationGPURecord.

Clean up ecalRawDecodingAndMultifit.py .
Rename ecalMultiFitUncalibRecHitGPUcfi.py to ecalMultiFitUncalibRecHitGPU_cfi.py .

…cies)T (#569)

Move few similar implementations of templated ESProducers
  - EcalESProducerGPU
  - EcalRawESProducerGPU
  - HcalESProducerGPU
  - HcalESProducerGPUWithDependencies
  - HcalRawESProducerGPU
to a common implementation under HeterogeneousCore/CUDACore/ .

Adapt all client code accordingly.

Do not use transient handles to avoid ESProducers taking references to
transient memory objects.

Factor out the chi2 computation from the ECAL multifit and HCAL MAHI code,
and move it to MultifitComputations.

Clean up ECAL local reconstruction code:
  - improve comments for documentation
  - remove commented out code, unused functions and unused header files
  - introduce better variable names
  - migrates to esConsumes
  - store ParameterSet by value instead of by reference
  - move __syncthreads() outside of the if blocks to avoid possibly blocked threads
  - peplaces ecal::abs() with std::abs()

Co-authored-by: amassiro <[email protected]>

Update the default configuration of "EcalRawToDigiGPU" and "EcalUncalibRecHitProducerGPU":
  - enable the ECAL pulse timing computation in the offline workflow;
  - change the ECAL GPU digis label to match the CPU ones.

Other clean up:
  - remove an unused cfi file;
  - rename "EcalUncalibRecHitMultiFitAlgo_gpu_new" to "EcalUncalibRecHitMultiFitAlgoGPU".