Skip to content

Commit

Permalink
Configure the number of ECAL barrel and endcap channels separately (#517
Browse files Browse the repository at this point in the history 
)

Fix memory allocation issues.

Apply come code clean up:
  - remove outdated comments;
  - replace MYMALLOC macro with a lambda;
  - reuse named values from EcalDataFrame.
  • Loading branch information
@amassiro @fwyzard
amassiro authored and fwyzard committed Oct 8, 2020
1 parent cb87890 commit 1148c56
Show file tree
Hide file tree
Showing 7 changed files with 104 additions and 117 deletions.
18 changes: 8 additions & 10 deletions EventFilter/EcalRawToDigi/plugins/DeclsForKernels.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -3,15 +3,15 @@

#include <vector>

#include "CUDADataFormats/EcalDigi/interface/DigisCollection.h"
#include "DataFormats/EcalDigi/interface/EcalDataFrame.h"
#include "EventFilter/EcalRawToDigi/interface/DCCRawDataDefinitions.h"
#include "EventFilter/EcalRawToDigi/interface/ElectronicsMappingGPU.h"
#include "HeterogeneousCore/CUDAUtilities/interface/HostAllocator.h"
#include "HeterogeneousCore/CUDAUtilities/interface/cudaCheck.h"
#include "HeterogeneousCore/CUDAUtilities/interface/device_unique_ptr.h"
#include "HeterogeneousCore/CUDAUtilities/interface/host_unique_ptr.h"

#include "CUDADataFormats/EcalDigi/interface/DigisCollection.h"

namespace ecal {
namespace raw {

Expand All @@ -26,7 +26,8 @@ namespace ecal {
};

struct ConfigurationParameters {
uint32_t maxChannels;
uint32_t maxChannelsEE;
uint32_t maxChannelsEB;
};

struct OutputDataCPU {
Expand All @@ -37,14 +38,11 @@ namespace ecal {
struct OutputDataGPU {
DigisCollection<::calo::common::DevStoragePolicy> digisEB, digisEE;

// FIXME: we should separate max channels parameter for eb and ee
// FIXME: replace hardcoded values
void allocate(ConfigurationParameters const &config, cudaStream_t cudaStream) {
digisEB.data = cms::cuda::make_device_unique<uint16_t[]>(config.maxChannels, cudaStream);
digisEE.data = cms::cuda::make_device_unique<uint16_t[]>(config.maxChannels, cudaStream);

digisEB.ids = cms::cuda::make_device_unique<uint32_t[]>(config.maxChannels, cudaStream);
digisEE.ids = cms::cuda::make_device_unique<uint32_t[]>(config.maxChannels, cudaStream);
digisEB.data = cms::cuda::make_device_unique<uint16_t[]>(config.maxChannelsEB * EcalDataFrame::MAXSAMPLES, cudaStream);
digisEE.data = cms::cuda::make_device_unique<uint16_t[]>(config.maxChannelsEE * EcalDataFrame::MAXSAMPLES, cudaStream);
digisEB.ids = cms::cuda::make_device_unique<uint32_t[]>(config.maxChannelsEB, cudaStream);
digisEE.ids = cms::cuda::make_device_unique<uint32_t[]>(config.maxChannelsEE, cudaStream);
}
};

Expand Down
6 changes: 3 additions & 3 deletions EventFilter/EcalRawToDigi/plugins/EcalCPUDigisProducer.cc
View file
Original file line number Diff line number Diff line change
Expand Up @@ -3,6 +3,7 @@
#include "CUDADataFormats/EcalDigi/interface/DigisCollection.h"
#include "CondFormats/DataRecord/interface/EcalMappingElectronicsRcd.h"
#include "DataFormats/EcalDetId/interface/EcalDetIdCollections.h"
#include "DataFormats/EcalDigi/interface/EcalDataFrame.h"
#include "DataFormats/EcalDigi/interface/EcalDigiCollections.h"
#include "DataFormats/EcalDigi/interface/EcalDigiCollections.h"
#include "DataFormats/FEDRawData/interface/FEDRawDataCollection.h"
Expand Down Expand Up @@ -134,11 +135,10 @@ void EcalCPUDigisProducer::acquire(edm::Event const& event,
auto const& eedigis = ctx.get(eedigisProduct);

// resize tmp buffers
// FIXME remove hardcoded values
dataebtmp.resize(ebdigis.size * EcalDataFrame::MAXSAMPLES);
dataeetmp.resize(eedigis.size * EcalDataFrame::MAXSAMPLES);
idsebtmp.resize(ebdigis.size);
dataebtmp.resize(ebdigis.size * 10);
idseetmp.resize(eedigis.size);
dataeetmp.resize(eedigis.size * 10);

// enqeue transfers
cudaCheck(cudaMemcpyAsync(
Expand Down
6 changes: 4 additions & 2 deletions EventFilter/EcalRawToDigi/plugins/EcalRawToDigiGPU.cc
View file
Original file line number Diff line number Diff line change
Expand Up @@ -48,7 +48,8 @@ void EcalRawToDigiGPU::fillDescriptions(edm::ConfigurationDescriptions& confDesc
for (uint32_t i = 0; i < 54; ++i)
feds[i] = i + 601;
desc.add<std::vector<int>>("FEDs", feds);
desc.add<uint32_t>("maxChannels", 20000);
desc.add<uint32_t>("maxChannelsEB", 61200);
desc.add<uint32_t>("maxChannelsEE", 14648);
desc.add<std::string>("digisLabelEB", "ebDigisGPU");
desc.add<std::string>("digisLabelEE", "eeDigisGPU");

Expand All @@ -61,7 +62,8 @@ EcalRawToDigiGPU::EcalRawToDigiGPU(const edm::ParameterSet& ps)
digisEBToken_{produces<OutputProduct>(ps.getParameter<std::string>("digisLabelEB"))},
digisEEToken_{produces<OutputProduct>(ps.getParameter<std::string>("digisLabelEE"))},
fedsToUnpack_{ps.getParameter<std::vector<int>>("FEDs")} {
config_.maxChannels = ps.getParameter<uint32_t>("maxChannels");
config_.maxChannelsEB = ps.getParameter<uint32_t>("maxChannelsEB");
config_.maxChannelsEE = ps.getParameter<uint32_t>("maxChannelsEE");
}

EcalRawToDigiGPU::~EcalRawToDigiGPU() {}
Expand Down
163 changes: 73 additions & 90 deletions RecoLocalCalo/EcalRecProducers/plugins/DeclsForKernels.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -91,41 +91,43 @@ namespace ecal {

bool shouldRunTimingComputation;

uint32_t maxNumberHits;
uint32_t maxNumberHitsEB;
uint32_t maxNumberHitsEE;
};

struct EventOutputDataGPU {
UncalibratedRecHit<::calo::common::DevStoragePolicy> recHitsEB, recHitsEE;

void allocate(ConfigurationParameters const& configParameters, cudaStream_t cudaStream) {
auto const size = configParameters.maxNumberHits;
auto const sizeEB = configParameters.maxNumberHitsEB;
recHitsEB.amplitudesAll =
cms::cuda::make_device_unique<reco::ComputationScalarType[]>(size * EcalDataFrame::MAXSAMPLES, cudaStream);
recHitsEB.amplitude = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
recHitsEB.chi2 = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
recHitsEB.pedestal = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
cms::cuda::make_device_unique<reco::ComputationScalarType[]>(sizeEB * EcalDataFrame::MAXSAMPLES, cudaStream);
recHitsEB.amplitude = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEB, cudaStream);
recHitsEB.chi2 = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEB, cudaStream);
recHitsEB.pedestal = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEB, cudaStream);

if (configParameters.shouldRunTimingComputation) {
recHitsEB.jitter = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
recHitsEB.jitterError = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
recHitsEB.jitter = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEB, cudaStream);
recHitsEB.jitterError = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEB, cudaStream);
}

recHitsEB.did = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);
recHitsEB.flags = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);
recHitsEB.did = cms::cuda::make_device_unique<uint32_t[]>(sizeEB, cudaStream);
recHitsEB.flags = cms::cuda::make_device_unique<uint32_t[]>(sizeEB, cudaStream);

auto const sizeEE = configParameters.maxNumberHitsEE;
recHitsEE.amplitudesAll =
cms::cuda::make_device_unique<reco::ComputationScalarType[]>(size * EcalDataFrame::MAXSAMPLES, cudaStream);
recHitsEE.amplitude = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
recHitsEE.chi2 = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
recHitsEE.pedestal = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
cms::cuda::make_device_unique<reco::ComputationScalarType[]>(sizeEE * EcalDataFrame::MAXSAMPLES, cudaStream);
recHitsEE.amplitude = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEE, cudaStream);
recHitsEE.chi2 = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEE, cudaStream);
recHitsEE.pedestal = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEE, cudaStream);

if (configParameters.shouldRunTimingComputation) {
recHitsEE.jitter = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
recHitsEE.jitterError = cms::cuda::make_device_unique<reco::StorageScalarType[]>(size, cudaStream);
recHitsEE.jitter = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEE, cudaStream);
recHitsEE.jitterError = cms::cuda::make_device_unique<reco::StorageScalarType[]>(sizeEE, cudaStream);
}

recHitsEE.did = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);
recHitsEE.flags = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);
recHitsEE.did = cms::cuda::make_device_unique<uint32_t[]>(sizeEE, cudaStream);
recHitsEE.flags = cms::cuda::make_device_unique<uint32_t[]>(sizeEE, cudaStream);
}
};

Expand Down Expand Up @@ -169,64 +171,43 @@ namespace ecal {
constexpr auto smlength = getLength<SampleMatrix>();
constexpr auto pmlength = getLength<PulseMatrixType>();
constexpr auto bxvlength = getLength<BXVectorType>();
auto const size = configParameters.maxNumberHits;

#define MYMALLOC(var, size) var = cms::cuda::make_device_unique<decltype(var)::element_type[]>(size, cudaStream)
MYMALLOC(samples, size * svlength);
//cudaCheck(cudaMalloc((void**)&samples, size * sizeof(SampleVector)));
MYMALLOC(gainsNoise, size * sgvlength);
//cudaCheck(cudaMalloc((void**)&gainsNoise, size * sizeof(SampleGainVector)));

MYMALLOC(noisecov, size * smlength);
//cudaCheck(cudaMalloc((void**)&noisecov, size * sizeof(SampleMatrix)));
MYMALLOC(pulse_matrix, size * pmlength);
//cudaCheck(cudaMalloc((void**)&pulse_matrix, size * sizeof(PulseMatrixType)));
MYMALLOC(activeBXs, size * bxvlength);
//cudaCheck(cudaMalloc((void**)&activeBXs, size * sizeof(BXVectorType)));
MYMALLOC(acState, size);
//cudaCheck(cudaMalloc((void**)&acState, size * sizeof(char)));

MYMALLOC(hasSwitchToGain6, size);
//cudaCheck(cudaMalloc((void**)&hasSwitchToGain6, size * sizeof(bool)));
MYMALLOC(hasSwitchToGain1, size);
//cudaCheck(cudaMalloc((void**)&hasSwitchToGain1, size * sizeof(bool)));
MYMALLOC(isSaturated, size);
//cudaCheck(cudaMalloc((void**)&isSaturated, size * sizeof(bool)));
auto const size = configParameters.maxNumberHitsEB + configParameters.maxNumberHitsEE;

auto alloc = [cudaStream](auto& var, uint32_t size) {
using element_type = typename std::remove_reference_t<decltype(var)>::element_type;
var = cms::cuda::make_device_unique<element_type[]>(size, cudaStream);
};

alloc(samples, size * svlength);
alloc(gainsNoise, size * sgvlength);

alloc(noisecov, size * smlength);
alloc(pulse_matrix, size * pmlength);
alloc(activeBXs, size * bxvlength);
alloc(acState, size);

alloc(hasSwitchToGain6, size);
alloc(hasSwitchToGain1, size);
alloc(isSaturated, size);

if (configParameters.shouldRunTimingComputation) {
MYMALLOC(sample_values, size * svlength);
//cudaCheck(cudaMalloc((void**)&sample_values, size * sizeof(SampleVector)));
MYMALLOC(sample_value_errors, size * svlength);
//cudaCheck(cudaMalloc((void**)&sample_value_errors, size * sizeof(SampleVector)));
MYMALLOC(useless_sample_values, size * EcalDataFrame::MAXSAMPLES);
//cudaCheck(cudaMalloc((void**)&useless_sample_values, size * sizeof(bool) * EcalDataFrame::MAXSAMPLES));
MYMALLOC(chi2sNullHypot, size);
//cudaCheck(cudaMalloc((void**)&chi2sNullHypot, size * sizeof(SampleVector::Scalar)));
MYMALLOC(sum0sNullHypot, size);
//cudaCheck(cudaMalloc((void**)&sum0sNullHypot, size * sizeof(SampleVector::Scalar)));
MYMALLOC(sumAAsNullHypot, size);
//cudaCheck(cudaMalloc((void**)&sumAAsNullHypot, size * sizeof(SampleVector::Scalar)));
MYMALLOC(pedestal_nums, size);
//cudaCheck(cudaMalloc((void**)&pedestal_nums, size * sizeof(char)));

MYMALLOC(tMaxAlphaBetas, size);
//cudaCheck(cudaMalloc((void**)&tMaxAlphaBetas, size * sizeof(SampleVector::Scalar)));
MYMALLOC(tMaxErrorAlphaBetas, size);
//cudaCheck(cudaMalloc((void**)&tMaxErrorAlphaBetas, size * sizeof(SampleVector::Scalar)));
MYMALLOC(accTimeMax, size);
//cudaCheck(cudaMalloc((void**)&accTimeMax, size * sizeof(SampleVector::Scalar)));
MYMALLOC(accTimeWgt, size);
//cudaCheck(cudaMalloc((void**)&accTimeWgt, size * sizeof(SampleVector::Scalar)));
MYMALLOC(ampMaxAlphaBeta, size);
//cudaCheck(cudaMalloc((void**)&ampMaxAlphaBeta, size * sizeof(SampleVector::Scalar)));
MYMALLOC(ampMaxError, size);
//cudaCheck(cudaMalloc((void**)&ampMaxError, size * sizeof(SampleVector::Scalar)));
MYMALLOC(timeMax, size);
//cudaCheck(cudaMalloc((void**)&timeMax, size * sizeof(SampleVector::Scalar)));
MYMALLOC(timeError, size);
//cudaCheck(cudaMalloc((void**)&timeError, size * sizeof(SampleVector::Scalar)));
MYMALLOC(tcState, size);
//cudaCheck(cudaMalloc((void**)&tcState, size * sizeof(TimeComputationState)));
alloc(sample_values, size * svlength);
alloc(sample_value_errors, size * svlength);
alloc(useless_sample_values, size * EcalDataFrame::MAXSAMPLES);
alloc(chi2sNullHypot, size);
alloc(sum0sNullHypot, size);
alloc(sumAAsNullHypot, size);
alloc(pedestal_nums, size);

alloc(tMaxAlphaBetas, size);
alloc(tMaxErrorAlphaBetas, size);
alloc(accTimeMax, size);
alloc(accTimeWgt, size);
alloc(ampMaxAlphaBeta, size);
alloc(ampMaxError, size);
alloc(timeMax, size);
alloc(timeError, size);
alloc(tcState, size);
}
}
};
Expand Down Expand Up @@ -295,8 +276,9 @@ namespace ecal {
uint32_t expanded_v_DB_reco_flagsSize;

uint32_t flagmask;
uint32_t maxNumberHits;

uint32_t maxNumberHitsEB;
uint32_t maxNumberHitsEE;

//
// bool shouldRunTimingComputation;
};
Expand All @@ -307,20 +289,21 @@ namespace ecal {
void allocate(ConfigurationParameters const& configParameters, cudaStream_t cudaStream) {
// void allocate(uint32_t size) {
//---- configParameters -> needed only to decide if to save the timing information or not
auto const size = configParameters.maxNumberHits;
recHitsEB.energy = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(size, cudaStream);
recHitsEB.time = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(size, cudaStream);
recHitsEB.chi2 = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(size, cudaStream);
recHitsEB.flagBits = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);
recHitsEB.extra = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);
recHitsEB.did = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);

recHitsEE.energy = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(size, cudaStream);
recHitsEE.time = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(size, cudaStream);
recHitsEE.chi2 = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(size, cudaStream);
recHitsEE.flagBits = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);
recHitsEE.extra = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);
recHitsEE.did = cms::cuda::make_device_unique<uint32_t[]>(size, cudaStream);
auto const sizeEB = configParameters.maxNumberHitsEB;
recHitsEB.energy = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(sizeEB, cudaStream);
recHitsEB.time = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(sizeEB, cudaStream);
recHitsEB.chi2 = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(sizeEB, cudaStream);
recHitsEB.flagBits = cms::cuda::make_device_unique<uint32_t[]>(sizeEB, cudaStream);
recHitsEB.extra = cms::cuda::make_device_unique<uint32_t[]>(sizeEB, cudaStream);
recHitsEB.did = cms::cuda::make_device_unique<uint32_t[]>(sizeEB, cudaStream);

auto const sizeEE = configParameters.maxNumberHitsEE;
recHitsEE.energy = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(sizeEE, cudaStream);
recHitsEE.time = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(sizeEE, cudaStream);
recHitsEE.chi2 = cms::cuda::make_device_unique<::ecal::reco::StorageScalarType[]>(sizeEE, cudaStream);
recHitsEE.flagBits = cms::cuda::make_device_unique<uint32_t[]>(sizeEE, cudaStream);
recHitsEE.extra = cms::cuda::make_device_unique<uint32_t[]>(sizeEE, cudaStream);
recHitsEE.did = cms::cuda::make_device_unique<uint32_t[]>(sizeEE, cudaStream);
}
};

Expand Down
12 changes: 7 additions & 5 deletions RecoLocalCalo/EcalRecProducers/plugins/EcalRecHitProducerGPU.cc
View file
Original file line number Diff line number Diff line change
Expand Up @@ -100,7 +100,8 @@ void EcalRecHitProducerGPU::fillDescriptions(edm::ConfigurationDescriptions& con
desc.add<double>("EBLaserMAX", 30.0);
desc.add<double>("EELaserMAX", 30.0);

desc.add<uint32_t>("maxNumberHits", 20000);
desc.add<uint32_t>("maxNumberHitsEB", 61200);
desc.add<uint32_t>("maxNumberHitsEE", 14648);
}

EcalRecHitProducerGPU::EcalRecHitProducerGPU(const edm::ParameterSet& ps) {
Expand All @@ -121,8 +122,9 @@ EcalRecHitProducerGPU::EcalRecHitProducerGPU(const edm::ParameterSet& ps) {
configParameters_.EELaserMAX = ps.getParameter<double>("EELaserMAX");

// max number of digis to allocate for
configParameters_.maxNumberHits = ps.getParameter<uint32_t>("maxNumberHits");

configParameters_.maxNumberHitsEB = ps.getParameter<uint32_t>("maxNumberHitsEB");
configParameters_.maxNumberHitsEE = ps.getParameter<uint32_t>("maxNumberHitsEE");

flagmask_ = 0;
flagmask_ |= 0x1 << EcalRecHit::kNeighboursRecovered;
flagmask_ |= 0x1 << EcalRecHit::kTowerRecovered;
Expand Down Expand Up @@ -163,8 +165,8 @@ void EcalRecHitProducerGPU::acquire(edm::Event const& event,
nee_ = eeUncalibRecHits.size;
// std::cout << " [EcalRecHitProducerGPU::acquire] neb_:nee_ = " << neb_ << " : " << nee_ << std::endl;

if ((neb_ + nee_) > configParameters_.maxNumberHits) {
edm::LogError("EcalRecHitProducerGPU") << "max number of channels exceeded. See options 'maxNumberHits' ";
if ((neb_ > configParameters_.maxNumberHitsEB) || (nee_ > configParameters_.maxNumberHitsEE)) {
edm::LogError("EcalRecHitProducerGPU") << "max number of channels exceeded. See options 'maxNumberHitsEB and maxNumberHitsEE' ";
}

int nchannelsEB = ebUncalibRecHits.size; // --> offsetForInput, first EB and then EE
Expand Down
12 changes: 7 additions & 5 deletions RecoLocalCalo/EcalRecProducers/plugins/EcalUncalibRecHitProducerGPU.cc
View file
Original file line number Diff line number Diff line change
Expand Up @@ -99,7 +99,8 @@ void EcalUncalibRecHitProducerGPU::fillDescriptions(edm::ConfigurationDescriptio
desc.add<double>("outOfTimeThresholdGain61mEE", 1000);
desc.add<double>("amplitudeThresholdEB", 10);
desc.add<double>("amplitudeThresholdEE", 10);
desc.add<uint32_t>("maxNumberHits", 20000); //---- AM TEST
desc.add<uint32_t>("maxNumberHitsEB", 61200);
desc.add<uint32_t>("maxNumberHitsEE", 14648);
desc.add<std::vector<uint32_t>>("kernelMinimizeThreads", {32, 1, 1});
// ---- default false or true? It was set to true, but at HLT it is false
desc.add<bool>("shouldRunTimingComputation", false);
Expand Down Expand Up @@ -145,8 +146,9 @@ EcalUncalibRecHitProducerGPU::EcalUncalibRecHitProducerGPU(const edm::ParameterS
auto amplitudeThreshEE = ps.getParameter<double>("amplitudeThresholdEE");

// max number of digis to allocate for
configParameters_.maxNumberHits = ps.getParameter<uint32_t>("maxNumberHits");

configParameters_.maxNumberHitsEB = ps.getParameter<uint32_t>("maxNumberHitsEB");
configParameters_.maxNumberHitsEE = ps.getParameter<uint32_t>("maxNumberHitsEE");

// switch to run timing computation kernels
configParameters_.shouldRunTimingComputation = ps.getParameter<bool>("shouldRunTimingComputation");

Expand Down Expand Up @@ -210,8 +212,8 @@ void EcalUncalibRecHitProducerGPU::acquire(edm::Event const& event,
neb_ = ebDigis.size;
nee_ = eeDigis.size;

if ((neb_ + nee_) > configParameters_.maxNumberHits) {
edm::LogError("EcalUncalibRecHitProducerGPU") << "max number of channels exceeded. See options 'maxNumberHits' ";
if ((neb_ > configParameters_.maxNumberHitsEB) || (nee_ > configParameters_.maxNumberHitsEE)) {
edm::LogError("EcalUncalibRecHitProducerGPU") << "max number of channels exceeded. See options 'maxNumberHitsEB and maxNumberHitsEE' ";
}

// conditions
Expand Down
Loading

0 comments on commit 1148c56

Please sign in to comment.