diff --git a/CUDADataFormats/Common/BuildFile.xml b/CUDADataFormats/Common/BuildFile.xml
index b990c1295e31a..1046b76eef0f7 100644
--- a/CUDADataFormats/Common/BuildFile.xml
+++ b/CUDADataFormats/Common/BuildFile.xml
@@ -1,6 +1,7 @@
-
-
-
+
+
+
+
+
-
-
+
diff --git a/CUDADataFormats/SiPixelCluster/BuildFile.xml b/CUDADataFormats/SiPixelCluster/BuildFile.xml
new file mode 100644
index 0000000000000..d34658faa2573
--- /dev/null
+++ b/CUDADataFormats/SiPixelCluster/BuildFile.xml
@@ -0,0 +1,9 @@
+
+
+
+
+
+
+
+
+
diff --git a/CUDADataFormats/SiPixelCluster/interface/SiPixelClustersCUDA.h b/CUDADataFormats/SiPixelCluster/interface/SiPixelClustersCUDA.h
new file mode 100644
index 0000000000000..f25a8a25f0808
--- /dev/null
+++ b/CUDADataFormats/SiPixelCluster/interface/SiPixelClustersCUDA.h
@@ -0,0 +1,76 @@
+#ifndef CUDADataFormats_SiPixelCluster_interface_SiPixelClustersCUDA_h
+#define CUDADataFormats_SiPixelCluster_interface_SiPixelClustersCUDA_h
+
+#include "HeterogeneousCore/CUDAUtilities/interface/device_unique_ptr.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/host_unique_ptr.h"
+
+#include
+
+class SiPixelClustersCUDA {
+public:
+ SiPixelClustersCUDA() = default;
+ explicit SiPixelClustersCUDA(size_t maxClusters, cuda::stream_t<>& stream);
+ ~SiPixelClustersCUDA() = default;
+
+ SiPixelClustersCUDA(const SiPixelClustersCUDA&) = delete;
+ SiPixelClustersCUDA& operator=(const SiPixelClustersCUDA&) = delete;
+ SiPixelClustersCUDA(SiPixelClustersCUDA&&) = default;
+ SiPixelClustersCUDA& operator=(SiPixelClustersCUDA&&) = default;
+
+ void setNClusters(uint32_t nClusters) {
+ nClusters_h = nClusters;
+ }
+
+ uint32_t nClusters() const { return nClusters_h; }
+
+ uint32_t *moduleStart() { return moduleStart_d.get(); }
+ uint32_t *clusInModule() { return clusInModule_d.get(); }
+ uint32_t *moduleId() { return moduleId_d.get(); }
+ uint32_t *clusModuleStart() { return clusModuleStart_d.get(); }
+
+ uint32_t const *moduleStart() const { return moduleStart_d.get(); }
+ uint32_t const *clusInModule() const { return clusInModule_d.get(); }
+ uint32_t const *moduleId() const { return moduleId_d.get(); }
+ uint32_t const *clusModuleStart() const { return clusModuleStart_d.get(); }
+
+ uint32_t const *c_moduleStart() const { return moduleStart_d.get(); }
+ uint32_t const *c_clusInModule() const { return clusInModule_d.get(); }
+ uint32_t const *c_moduleId() const { return moduleId_d.get(); }
+ uint32_t const *c_clusModuleStart() const { return clusModuleStart_d.get(); }
+
+ class DeviceConstView {
+ public:
+ DeviceConstView() = default;
+
+#ifdef __CUDACC__
+ __device__ __forceinline__ uint32_t moduleStart(int i) const { return __ldg(moduleStart_+i); }
+ __device__ __forceinline__ uint32_t clusInModule(int i) const { return __ldg(clusInModule_+i); }
+ __device__ __forceinline__ uint32_t moduleId(int i) const { return __ldg(moduleId_+i); }
+ __device__ __forceinline__ uint32_t clusModuleStart(int i) const { return __ldg(clusModuleStart_+i); }
+#endif
+
+ friend SiPixelClustersCUDA;
+
+ private:
+ uint32_t const *moduleStart_;
+ uint32_t const *clusInModule_;
+ uint32_t const *moduleId_;
+ uint32_t const *clusModuleStart_;
+ };
+
+ DeviceConstView *view() const { return view_d.get(); }
+
+private:
+ cudautils::device::unique_ptr moduleStart_d; // index of the first pixel of each module
+ cudautils::device::unique_ptr clusInModule_d; // number of clusters found in each module
+ cudautils::device::unique_ptr moduleId_d; // module id of each module
+
+ // originally from rechits
+ cudautils::device::unique_ptr clusModuleStart_d;
+
+ cudautils::device::unique_ptr view_d; // "me" pointer
+
+ uint32_t nClusters_h;
+};
+
+#endif
diff --git a/CUDADataFormats/SiPixelCluster/src/SiPixelClustersCUDA.cc b/CUDADataFormats/SiPixelCluster/src/SiPixelClustersCUDA.cc
new file mode 100644
index 0000000000000..d88a1b0a6370b
--- /dev/null
+++ b/CUDADataFormats/SiPixelCluster/src/SiPixelClustersCUDA.cc
@@ -0,0 +1,23 @@
+#include "CUDADataFormats/SiPixelCluster/interface/SiPixelClustersCUDA.h"
+
+#include "FWCore/ServiceRegistry/interface/Service.h"
+#include "HeterogeneousCore/CUDAServices/interface/CUDAService.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/copyAsync.h"
+
+SiPixelClustersCUDA::SiPixelClustersCUDA(size_t maxClusters, cuda::stream_t<>& stream) {
+ edm::Service cs;
+
+ moduleStart_d = cs->make_device_unique(maxClusters+1, stream);
+ clusInModule_d = cs->make_device_unique(maxClusters, stream);
+ moduleId_d = cs->make_device_unique(maxClusters, stream);
+ clusModuleStart_d = cs->make_device_unique(maxClusters+1, stream);
+
+ auto view = cs->make_host_unique(stream);
+ view->moduleStart_ = moduleStart_d.get();
+ view->clusInModule_ = clusInModule_d.get();
+ view->moduleId_ = moduleId_d.get();
+ view->clusModuleStart_ = clusModuleStart_d.get();
+
+ view_d = cs->make_device_unique(stream);
+ cudautils::copyAsync(view_d, view, stream);
+}
diff --git a/CUDADataFormats/SiPixelCluster/src/classes.h b/CUDADataFormats/SiPixelCluster/src/classes.h
new file mode 100644
index 0000000000000..08d46244adc7d
--- /dev/null
+++ b/CUDADataFormats/SiPixelCluster/src/classes.h
@@ -0,0 +1,8 @@
+#ifndef CUDADataFormats_SiPixelCluster_classes_h
+#define CUDADataFormats_SiPixelCluster_classes_h
+
+#include "CUDADataFormats/Common/interface/CUDAProduct.h"
+#include "CUDADataFormats/SiPixelCluster/interface/SiPixelClustersCUDA.h"
+#include "DataFormats/Common/interface/Wrapper.h"
+
+#endif
diff --git a/CUDADataFormats/SiPixelCluster/src/classes_def.xml b/CUDADataFormats/SiPixelCluster/src/classes_def.xml
new file mode 100644
index 0000000000000..ba0706ac4b8aa
--- /dev/null
+++ b/CUDADataFormats/SiPixelCluster/src/classes_def.xml
@@ -0,0 +1,4 @@
+
+
+
+
diff --git a/CUDADataFormats/SiPixelDigi/BuildFile.xml b/CUDADataFormats/SiPixelDigi/BuildFile.xml
new file mode 100644
index 0000000000000..29ec13098819c
--- /dev/null
+++ b/CUDADataFormats/SiPixelDigi/BuildFile.xml
@@ -0,0 +1,9 @@
+
+
+
+
+
+
+
+
+
diff --git a/CUDADataFormats/SiPixelDigi/interface/SiPixelDigiErrorsCUDA.h b/CUDADataFormats/SiPixelDigi/interface/SiPixelDigiErrorsCUDA.h
new file mode 100644
index 0000000000000..e9c8c0f644722
--- /dev/null
+++ b/CUDADataFormats/SiPixelDigi/interface/SiPixelDigiErrorsCUDA.h
@@ -0,0 +1,40 @@
+#ifndef CUDADataFormats_SiPixelDigi_interface_SiPixelDigiErrorsCUDA_h
+#define CUDADataFormats_SiPixelDigi_interface_SiPixelDigiErrorsCUDA_h
+
+#include "DataFormats/SiPixelDigi/interface/PixelErrors.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/device_unique_ptr.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/host_unique_ptr.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/GPUSimpleVector.h"
+
+#include
+
+class SiPixelDigiErrorsCUDA {
+public:
+ SiPixelDigiErrorsCUDA() = default;
+ explicit SiPixelDigiErrorsCUDA(size_t maxFedWords, PixelFormatterErrors errors, cuda::stream_t<>& stream);
+ ~SiPixelDigiErrorsCUDA() = default;
+
+ SiPixelDigiErrorsCUDA(const SiPixelDigiErrorsCUDA&) = delete;
+ SiPixelDigiErrorsCUDA& operator=(const SiPixelDigiErrorsCUDA&) = delete;
+ SiPixelDigiErrorsCUDA(SiPixelDigiErrorsCUDA&&) = default;
+ SiPixelDigiErrorsCUDA& operator=(SiPixelDigiErrorsCUDA&&) = default;
+
+ const PixelFormatterErrors& formatterErrors() const { return formatterErrors_h; }
+
+ GPU::SimpleVector *error() { return error_d.get(); }
+ GPU::SimpleVector const *error() const { return error_d.get(); }
+ GPU::SimpleVector const *c_error() const { return error_d.get(); }
+
+ using HostDataError = std::pair, cudautils::host::unique_ptr>;
+ HostDataError dataErrorToHostAsync(cuda::stream_t<>& stream) const;
+
+ void copyErrorToHostAsync(cuda::stream_t<>& stream);
+
+private:
+ cudautils::device::unique_ptr data_d;
+ cudautils::device::unique_ptr> error_d;
+ cudautils::host::unique_ptr> error_h;
+ PixelFormatterErrors formatterErrors_h;
+};
+
+#endif
diff --git a/CUDADataFormats/SiPixelDigi/interface/SiPixelDigisCUDA.h b/CUDADataFormats/SiPixelDigi/interface/SiPixelDigisCUDA.h
new file mode 100644
index 0000000000000..6a52545483eb8
--- /dev/null
+++ b/CUDADataFormats/SiPixelDigi/interface/SiPixelDigisCUDA.h
@@ -0,0 +1,99 @@
+#ifndef CUDADataFormats_SiPixelDigi_interface_SiPixelDigisCUDA_h
+#define CUDADataFormats_SiPixelDigi_interface_SiPixelDigisCUDA_h
+
+#include "HeterogeneousCore/CUDAUtilities/interface/device_unique_ptr.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/host_unique_ptr.h"
+
+#include
+
+class SiPixelDigisCUDA {
+public:
+ SiPixelDigisCUDA() = default;
+ explicit SiPixelDigisCUDA(size_t maxFedWords, cuda::stream_t<>& stream);
+ ~SiPixelDigisCUDA() = default;
+
+ SiPixelDigisCUDA(const SiPixelDigisCUDA&) = delete;
+ SiPixelDigisCUDA& operator=(const SiPixelDigisCUDA&) = delete;
+ SiPixelDigisCUDA(SiPixelDigisCUDA&&) = default;
+ SiPixelDigisCUDA& operator=(SiPixelDigisCUDA&&) = default;
+
+ void setNModulesDigis(uint32_t nModules, uint32_t nDigis) {
+ nModules_h = nModules;
+ nDigis_h = nDigis;
+ }
+
+ uint32_t nModules() const { return nModules_h; }
+ uint32_t nDigis() const { return nDigis_h; }
+
+ uint16_t * xx() { return xx_d.get(); }
+ uint16_t * yy() { return yy_d.get(); }
+ uint16_t * adc() { return adc_d.get(); }
+ uint16_t * moduleInd() { return moduleInd_d.get(); }
+ int32_t * clus() { return clus_d.get(); }
+ uint32_t * pdigi() { return pdigi_d.get(); }
+ uint32_t * rawIdArr() { return rawIdArr_d.get(); }
+
+ uint16_t const *xx() const { return xx_d.get(); }
+ uint16_t const *yy() const { return yy_d.get(); }
+ uint16_t const *adc() const { return adc_d.get(); }
+ uint16_t const *moduleInd() const { return moduleInd_d.get(); }
+ int32_t const *clus() const { return clus_d.get(); }
+ uint32_t const *pdigi() const { return pdigi_d.get(); }
+ uint32_t const *rawIdArr() const { return rawIdArr_d.get(); }
+
+ uint16_t const *c_xx() const { return xx_d.get(); }
+ uint16_t const *c_yy() const { return yy_d.get(); }
+ uint16_t const *c_adc() const { return adc_d.get(); }
+ uint16_t const *c_moduleInd() const { return moduleInd_d.get(); }
+ int32_t const *c_clus() const { return clus_d.get(); }
+ uint32_t const *c_pdigi() const { return pdigi_d.get(); }
+ uint32_t const *c_rawIdArr() const { return rawIdArr_d.get(); }
+
+ cudautils::host::unique_ptr adcToHostAsync(cuda::stream_t<>& stream) const;
+ cudautils::host::unique_ptr< int32_t[]> clusToHostAsync(cuda::stream_t<>& stream) const;
+ cudautils::host::unique_ptr pdigiToHostAsync(cuda::stream_t<>& stream) const;
+ cudautils::host::unique_ptr rawIdArrToHostAsync(cuda::stream_t<>& stream) const;
+
+ class DeviceConstView {
+ public:
+ DeviceConstView() = default;
+
+#ifdef __CUDACC__
+ __device__ __forceinline__ uint16_t xx(int i) const { return __ldg(xx_+i); }
+ __device__ __forceinline__ uint16_t yy(int i) const { return __ldg(yy_+i); }
+ __device__ __forceinline__ uint16_t adc(int i) const { return __ldg(adc_+i); }
+ __device__ __forceinline__ uint16_t moduleInd(int i) const { return __ldg(moduleInd_+i); }
+ __device__ __forceinline__ int32_t clus(int i) const { return __ldg(clus_+i); }
+#endif
+
+ friend class SiPixelDigisCUDA;
+
+ private:
+ uint16_t const *xx_;
+ uint16_t const *yy_;
+ uint16_t const *adc_;
+ uint16_t const *moduleInd_;
+ int32_t const *clus_;
+ };
+
+ const DeviceConstView *view() const { return view_d.get(); }
+
+private:
+ // These are consumed by downstream device code
+ cudautils::device::unique_ptr xx_d; // local coordinates of each pixel
+ cudautils::device::unique_ptr yy_d; //
+ cudautils::device::unique_ptr adc_d; // ADC of each pixel
+ cudautils::device::unique_ptr moduleInd_d; // module id of each pixel
+ cudautils::device::unique_ptr clus_d; // cluster id of each pixel
+ cudautils::device::unique_ptr view_d; // "me" pointer
+
+ // These are for CPU output; should we (eventually) place them to a
+ // separate product?
+ cudautils::device::unique_ptr pdigi_d;
+ cudautils::device::unique_ptr rawIdArr_d;
+
+ uint32_t nModules_h = 0;
+ uint32_t nDigis_h = 0;
+};
+
+#endif
diff --git a/CUDADataFormats/SiPixelDigi/src/SiPixelDigiErrorsCUDA.cc b/CUDADataFormats/SiPixelDigi/src/SiPixelDigiErrorsCUDA.cc
new file mode 100644
index 0000000000000..92aab1ec9d578
--- /dev/null
+++ b/CUDADataFormats/SiPixelDigi/src/SiPixelDigiErrorsCUDA.cc
@@ -0,0 +1,44 @@
+#include "CUDADataFormats/SiPixelDigi/interface/SiPixelDigiErrorsCUDA.h"
+
+#include "FWCore/ServiceRegistry/interface/Service.h"
+#include "HeterogeneousCore/CUDAServices/interface/CUDAService.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/copyAsync.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/memsetAsync.h"
+
+SiPixelDigiErrorsCUDA::SiPixelDigiErrorsCUDA(size_t maxFedWords, PixelFormatterErrors errors, cuda::stream_t<>& stream):
+ formatterErrors_h(std::move(errors))
+{
+ edm::Service cs;
+
+ error_d = cs->make_device_unique>(stream);
+ data_d = cs->make_device_unique(maxFedWords, stream);
+
+ cudautils::memsetAsync(data_d, 0x00, maxFedWords, stream);
+
+ error_h = cs->make_host_unique>(stream);
+ GPU::make_SimpleVector(error_h.get(), maxFedWords, data_d.get());
+ assert(error_h->size() == 0);
+ assert(error_h->capacity() == static_cast(maxFedWords));
+
+ cudautils::copyAsync(error_d, error_h, stream);
+}
+
+void SiPixelDigiErrorsCUDA::copyErrorToHostAsync(cuda::stream_t<>& stream) {
+ cudautils::copyAsync(error_h, error_d, stream);
+}
+
+SiPixelDigiErrorsCUDA::HostDataError SiPixelDigiErrorsCUDA::dataErrorToHostAsync(cuda::stream_t<>& stream) const {
+ edm::Service cs;
+ // On one hand size() could be sufficient. On the other hand, if
+ // someone copies the SimpleVector<>, (s)he might expect the data
+ // buffer to actually have space for capacity() elements.
+ auto data = cs->make_host_unique(error_h->capacity(), stream);
+
+ // but transfer only the required amount
+ if(error_h->size() > 0) {
+ cudautils::copyAsync(data, data_d, error_h->size(), stream);
+ }
+ auto err = *error_h;
+ err.set_data(data.get());
+ return HostDataError(std::move(err), std::move(data));
+}
diff --git a/CUDADataFormats/SiPixelDigi/src/SiPixelDigisCUDA.cc b/CUDADataFormats/SiPixelDigi/src/SiPixelDigisCUDA.cc
new file mode 100644
index 0000000000000..ef13ed9612dbf
--- /dev/null
+++ b/CUDADataFormats/SiPixelDigi/src/SiPixelDigisCUDA.cc
@@ -0,0 +1,56 @@
+#include "CUDADataFormats/SiPixelDigi/interface/SiPixelDigisCUDA.h"
+
+#include "FWCore/ServiceRegistry/interface/Service.h"
+#include "HeterogeneousCore/CUDAServices/interface/CUDAService.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/copyAsync.h"
+
+SiPixelDigisCUDA::SiPixelDigisCUDA(size_t maxFedWords, cuda::stream_t<>& stream) {
+ edm::Service cs;
+
+ xx_d = cs->make_device_unique(maxFedWords, stream);
+ yy_d = cs->make_device_unique(maxFedWords, stream);
+ adc_d = cs->make_device_unique(maxFedWords, stream);
+ moduleInd_d = cs->make_device_unique(maxFedWords, stream);
+ clus_d = cs->make_device_unique< int32_t[]>(maxFedWords, stream);
+
+ pdigi_d = cs->make_device_unique(maxFedWords, stream);
+ rawIdArr_d = cs->make_device_unique(maxFedWords, stream);
+
+ auto view = cs->make_host_unique(stream);
+ view->xx_ = xx_d.get();
+ view->yy_ = yy_d.get();
+ view->adc_ = adc_d.get();
+ view->moduleInd_ = moduleInd_d.get();
+ view->clus_ = clus_d.get();
+
+ view_d = cs->make_device_unique(stream);
+ cudautils::copyAsync(view_d, view, stream);
+}
+
+cudautils::host::unique_ptr SiPixelDigisCUDA::adcToHostAsync(cuda::stream_t<>& stream) const {
+ edm::Service cs;
+ auto ret = cs->make_host_unique(nDigis(), stream);
+ cudautils::copyAsync(ret, adc_d, nDigis(), stream);
+ return ret;
+}
+
+cudautils::host::unique_ptr SiPixelDigisCUDA::clusToHostAsync(cuda::stream_t<>& stream) const {
+ edm::Service cs;
+ auto ret = cs->make_host_unique(nDigis(), stream);
+ cudautils::copyAsync(ret, clus_d, nDigis(), stream);
+ return ret;
+}
+
+cudautils::host::unique_ptr SiPixelDigisCUDA::pdigiToHostAsync(cuda::stream_t<>& stream) const {
+ edm::Service cs;
+ auto ret = cs->make_host_unique(nDigis(), stream);
+ cudautils::copyAsync(ret, pdigi_d, nDigis(), stream);
+ return ret;
+}
+
+cudautils::host::unique_ptr SiPixelDigisCUDA::rawIdArrToHostAsync(cuda::stream_t<>& stream) const {
+ edm::Service cs;
+ auto ret = cs->make_host_unique(nDigis(), stream);
+ cudautils::copyAsync(ret, rawIdArr_d, nDigis(), stream);
+ return ret;
+}
diff --git a/CUDADataFormats/SiPixelDigi/src/classes.h b/CUDADataFormats/SiPixelDigi/src/classes.h
new file mode 100644
index 0000000000000..41b135640b883
--- /dev/null
+++ b/CUDADataFormats/SiPixelDigi/src/classes.h
@@ -0,0 +1,9 @@
+#ifndef CUDADataFormats_SiPixelDigi_classes_h
+#define CUDADataFormats_SiPixelDigi_classes_h
+
+#include "CUDADataFormats/Common/interface/CUDAProduct.h"
+#include "CUDADataFormats/SiPixelDigi/interface/SiPixelDigisCUDA.h"
+#include "CUDADataFormats/SiPixelDigi/interface/SiPixelDigiErrorsCUDA.h"
+#include "DataFormats/Common/interface/Wrapper.h"
+
+#endif
diff --git a/CUDADataFormats/SiPixelDigi/src/classes_def.xml b/CUDADataFormats/SiPixelDigi/src/classes_def.xml
new file mode 100644
index 0000000000000..9d6816ed3b14c
--- /dev/null
+++ b/CUDADataFormats/SiPixelDigi/src/classes_def.xml
@@ -0,0 +1,7 @@
+
+
+
+
+
+
+
diff --git a/CalibTracker/Records/interface/SiPixelGainCalibrationForHLTGPURcd.h b/CalibTracker/Records/interface/SiPixelGainCalibrationForHLTGPURcd.h
new file mode 100644
index 0000000000000..afb682e5d451f
--- /dev/null
+++ b/CalibTracker/Records/interface/SiPixelGainCalibrationForHLTGPURcd.h
@@ -0,0 +1,14 @@
+#ifndef CalibTracker_Records_SiPixelGainCalibrationForHLTGPURcd_h
+#define CalibTracker_Records_SiPixelGainCalibrationForHLTGPURcd_h
+
+#include "FWCore/Framework/interface/EventSetupRecordImplementation.h"
+#include "FWCore/Framework/interface/DependentRecordImplementation.h"
+
+#include "CondFormats/DataRecord/interface/SiPixelGainCalibrationForHLTRcd.h"
+#include "Geometry/Records/interface/TrackerDigiGeometryRecord.h"
+
+#include "boost/mpl/vector.hpp"
+
+class SiPixelGainCalibrationForHLTGPURcd : public edm::eventsetup::DependentRecordImplementation > {};
+
+#endif
diff --git a/CalibTracker/Records/src/SiPixelGainCalibrationForHLTGPURcd.cc b/CalibTracker/Records/src/SiPixelGainCalibrationForHLTGPURcd.cc
new file mode 100644
index 0000000000000..e6020eca80b1f
--- /dev/null
+++ b/CalibTracker/Records/src/SiPixelGainCalibrationForHLTGPURcd.cc
@@ -0,0 +1,5 @@
+#include "CalibTracker/Records/interface/SiPixelGainCalibrationForHLTGPURcd.h"
+#include "FWCore/Framework/interface/eventsetuprecord_registration_macro.h"
+#include "FWCore/Utilities/interface/typelookup.h"
+
+EVENTSETUP_RECORD_REG(SiPixelGainCalibrationForHLTGPURcd);
diff --git a/CalibTracker/SiPixelESProducers/BuildFile.xml b/CalibTracker/SiPixelESProducers/BuildFile.xml
index 6efeef5ca0d1c..69d258da21ed1 100644
--- a/CalibTracker/SiPixelESProducers/BuildFile.xml
+++ b/CalibTracker/SiPixelESProducers/BuildFile.xml
@@ -1,10 +1,15 @@
-
-
-
-
-
-
-
+
+
+
+
+
+
+
+
+
+
+
+
-
+
diff --git a/CalibTracker/SiPixelESProducers/interface/SiPixelGainCalibrationForHLTGPU.h b/CalibTracker/SiPixelESProducers/interface/SiPixelGainCalibrationForHLTGPU.h
new file mode 100644
index 0000000000000..96989c8a2c3b2
--- /dev/null
+++ b/CalibTracker/SiPixelESProducers/interface/SiPixelGainCalibrationForHLTGPU.h
@@ -0,0 +1,32 @@
+#ifndef CalibTracker_SiPixelESProducers_SiPixelGainCalibrationForHLTGPU_H
+#define CalibTracker_SiPixelESProducers_SiPixelGainCalibrationForHLTGPU_H
+
+#include "HeterogeneousCore/CUDACore/interface/CUDAESProduct.h"
+#include "CondFormats/SiPixelObjects/interface/SiPixelGainCalibrationForHLT.h"
+
+#include
+
+class SiPixelGainCalibrationForHLT;
+class SiPixelGainForHLTonGPU;
+struct SiPixelGainForHLTonGPU_DecodingStructure;
+class TrackerGeometry;
+
+class SiPixelGainCalibrationForHLTGPU {
+public:
+ explicit SiPixelGainCalibrationForHLTGPU(const SiPixelGainCalibrationForHLT& gains, const TrackerGeometry& geom);
+ ~SiPixelGainCalibrationForHLTGPU();
+
+ const SiPixelGainForHLTonGPU *getGPUProductAsync(cuda::stream_t<>& cudaStream) const;
+
+private:
+ const SiPixelGainCalibrationForHLT *gains_ = nullptr;
+ SiPixelGainForHLTonGPU *gainForHLTonHost_ = nullptr;
+ struct GPUData {
+ ~GPUData();
+ SiPixelGainForHLTonGPU *gainForHLTonGPU = nullptr;
+ SiPixelGainForHLTonGPU_DecodingStructure *gainDataOnGPU = nullptr;
+ };
+ CUDAESProduct gpuData_;
+};
+
+#endif
diff --git a/CalibTracker/SiPixelESProducers/plugins/BuildFile.xml b/CalibTracker/SiPixelESProducers/plugins/BuildFile.xml
index 5380c9d7d346b..b33657e273036 100644
--- a/CalibTracker/SiPixelESProducers/plugins/BuildFile.xml
+++ b/CalibTracker/SiPixelESProducers/plugins/BuildFile.xml
@@ -1,13 +1,13 @@
-
-
-
-
-
-
-
-
-
-
-
-
+
+
+
+
+
+
+
+
+
+
+
+
diff --git a/CalibTracker/SiPixelESProducers/plugins/SiPixelGainCalibrationForHLTGPUESProducer.cc b/CalibTracker/SiPixelESProducers/plugins/SiPixelGainCalibrationForHLTGPUESProducer.cc
new file mode 100644
index 0000000000000..186bb2d72c3f3
--- /dev/null
+++ b/CalibTracker/SiPixelESProducers/plugins/SiPixelGainCalibrationForHLTGPUESProducer.cc
@@ -0,0 +1,47 @@
+#include "CalibTracker/SiPixelESProducers/interface/SiPixelGainCalibrationForHLTGPU.h"
+#include "CalibTracker/Records/interface/SiPixelGainCalibrationForHLTGPURcd.h"
+#include "CondFormats/SiPixelObjects/interface/SiPixelGainCalibrationForHLT.h"
+#include "CondFormats/DataRecord/interface/SiPixelGainCalibrationForHLTRcd.h"
+#include "FWCore/Framework/interface/ESProducer.h"
+#include "FWCore/Framework/interface/EventSetup.h"
+#include "FWCore/Framework/interface/ESHandle.h"
+#include "FWCore/Framework/interface/ModuleFactory.h"
+#include "FWCore/ParameterSet/interface/ParameterSet.h"
+#include "Geometry/TrackerGeometryBuilder/interface/TrackerGeometry.h"
+#include "Geometry/Records/interface/TrackerDigiGeometryRecord.h"
+
+#include
+
+class SiPixelGainCalibrationForHLTGPUESProducer: public edm::ESProducer {
+public:
+ explicit SiPixelGainCalibrationForHLTGPUESProducer(const edm::ParameterSet& iConfig);
+ std::unique_ptr produce(const SiPixelGainCalibrationForHLTGPURcd& iRecord);
+
+ static void fillDescriptions(edm::ConfigurationDescriptions& descriptions);
+private:
+};
+
+SiPixelGainCalibrationForHLTGPUESProducer::SiPixelGainCalibrationForHLTGPUESProducer(const edm::ParameterSet& iConfig) {
+ setWhatProduced(this);
+}
+
+void SiPixelGainCalibrationForHLTGPUESProducer::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
+ edm::ParameterSetDescription desc;
+ descriptions.add("siPixelGainCalibrationForHLTGPU", desc);
+}
+
+std::unique_ptr SiPixelGainCalibrationForHLTGPUESProducer::produce(const SiPixelGainCalibrationForHLTGPURcd& iRecord) {
+ edm::ESHandle gains;
+ iRecord.getRecord().get(gains);
+
+ edm::ESHandle geom;
+ iRecord.getRecord().get(geom);
+
+ return std::make_unique(*gains, *geom);
+}
+
+#include "FWCore/Framework/interface/MakerMacros.h"
+#include "FWCore/Utilities/interface/typelookup.h"
+#include "FWCore/Framework/interface/eventsetuprecord_registration_macro.h"
+
+DEFINE_FWK_EVENTSETUP_MODULE(SiPixelGainCalibrationForHLTGPUESProducer);
diff --git a/CalibTracker/SiPixelESProducers/src/ES_SiPixelGainCalibrationForHLTGPU.cc b/CalibTracker/SiPixelESProducers/src/ES_SiPixelGainCalibrationForHLTGPU.cc
new file mode 100644
index 0000000000000..80932fb468f71
--- /dev/null
+++ b/CalibTracker/SiPixelESProducers/src/ES_SiPixelGainCalibrationForHLTGPU.cc
@@ -0,0 +1,4 @@
+#include "CalibTracker/SiPixelESProducers/interface/SiPixelGainCalibrationForHLTGPU.h"
+#include "FWCore/Utilities/interface/typelookup.h"
+
+TYPELOOKUP_DATA_REG(SiPixelGainCalibrationForHLTGPU);
diff --git a/CalibTracker/SiPixelESProducers/src/SiPixelGainCalibrationForHLTGPU.cc b/CalibTracker/SiPixelESProducers/src/SiPixelGainCalibrationForHLTGPU.cc
new file mode 100644
index 0000000000000..3aef3f44c8f67
--- /dev/null
+++ b/CalibTracker/SiPixelESProducers/src/SiPixelGainCalibrationForHLTGPU.cc
@@ -0,0 +1,98 @@
+#include "CalibTracker/SiPixelESProducers/interface/SiPixelGainCalibrationForHLTGPU.h"
+#include "CondFormats/SiPixelObjects/interface/SiPixelGainCalibrationForHLT.h"
+#include "CondFormats/SiPixelObjects/interface/SiPixelGainForHLTonGPU.h"
+#include "Geometry/TrackerGeometryBuilder/interface/TrackerGeometry.h"
+#include "Geometry/CommonDetUnit/interface/GeomDetType.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/cudaCheck.h"
+
+#include
+
+SiPixelGainCalibrationForHLTGPU::SiPixelGainCalibrationForHLTGPU(const SiPixelGainCalibrationForHLT& gains, const TrackerGeometry& geom):
+ gains_(&gains)
+{
+ // bizzarre logic (looking for fist strip-det) don't ask
+ auto const & dus = geom.detUnits();
+ unsigned m_detectors = dus.size();
+ for(unsigned int i=1;i<7;++i) {
+ if(geom.offsetDU(GeomDetEnumerators::tkDetEnum[i]) != dus.size() &&
+ dus[geom.offsetDU(GeomDetEnumerators::tkDetEnum[i])]->type().isTrackerStrip()) {
+ if(geom.offsetDU(GeomDetEnumerators::tkDetEnum[i]) < m_detectors) m_detectors = geom.offsetDU(GeomDetEnumerators::tkDetEnum[i]);
+ }
+ }
+
+ /*
+ std::cout << "caching calibs for " << m_detectors << " pixel detectors of size " << gains.data().size() << std::endl;
+ std::cout << "sizes " << sizeof(char) << ' ' << sizeof(uint8_t) << ' ' << sizeof(SiPixelGainForHLTonGPU::DecodingStructure) << std::endl;
+ */
+
+ cudaCheck(cudaMallocHost((void**) & gainForHLTonHost_, sizeof(SiPixelGainForHLTonGPU)));
+ //gainForHLTonHost_->v_pedestals = gainDataOnGPU_; // how to do this?
+
+ // do not read back from the (possibly write-combined) memory buffer
+ auto minPed = gains.getPedLow();
+ auto maxPed = gains.getPedHigh();
+ auto minGain = gains.getGainLow();
+ auto maxGain = gains.getGainHigh();
+ auto nBinsToUseForEncoding = 253;
+
+ // we will simplify later (not everything is needed....)
+ gainForHLTonHost_->minPed_ = minPed;
+ gainForHLTonHost_->maxPed_ = maxPed;
+ gainForHLTonHost_->minGain_= minGain;
+ gainForHLTonHost_->maxGain_= maxGain;
+
+ gainForHLTonHost_->numberOfRowsAveragedOver_ = 80;
+ gainForHLTonHost_->nBinsToUseForEncoding_ = nBinsToUseForEncoding;
+ gainForHLTonHost_->deadFlag_ = 255;
+ gainForHLTonHost_->noisyFlag_ = 254;
+
+ gainForHLTonHost_->pedPrecision = static_cast(maxPed - minPed) / nBinsToUseForEncoding;
+ gainForHLTonHost_->gainPrecision = static_cast(maxGain - minGain) / nBinsToUseForEncoding;
+
+ /*
+ std::cout << "precisions g " << gainForHLTonHost_->pedPrecision << ' ' << gainForHLTonHost_->gainPrecision << std::endl;
+ */
+
+ // fill the index map
+ auto const & ind = gains.getIndexes();
+ /*
+ std::cout << ind.size() << " " << m_detectors << std::endl;
+ */
+
+ for (auto i=0U; igeographicalId().rawId(),SiPixelGainCalibrationForHLT::StrictWeakOrdering());
+ assert (p!=ind.end() && p->detid==dus[i]->geographicalId());
+ assert(p->iend<=gains.data().size());
+ assert(p->iend>=p->ibegin);
+ assert(0==p->ibegin%2);
+ assert(0==p->iend%2);
+ assert(p->ibegin!=p->iend);
+ assert(p->ncols>0);
+ gainForHLTonHost_->rangeAndCols[i] = std::make_pair(SiPixelGainForHLTonGPU::Range(p->ibegin,p->iend), p->ncols);
+ // if (ind[i].detid!=dus[i]->geographicalId()) std::cout << ind[i].detid<<"!="<geographicalId() << std::endl;
+ // gainForHLTonHost_->rangeAndCols[i] = std::make_pair(SiPixelGainForHLTonGPU::Range(ind[i].ibegin,ind[i].iend), ind[i].ncols);
+ }
+
+}
+
+SiPixelGainCalibrationForHLTGPU::~SiPixelGainCalibrationForHLTGPU() {
+ cudaCheck(cudaFreeHost(gainForHLTonHost_));
+}
+
+SiPixelGainCalibrationForHLTGPU::GPUData::~GPUData() {
+ cudaCheck(cudaFree(gainForHLTonGPU));
+ cudaCheck(cudaFree(gainDataOnGPU));
+}
+
+const SiPixelGainForHLTonGPU *SiPixelGainCalibrationForHLTGPU::getGPUProductAsync(cuda::stream_t<>& cudaStream) const {
+ const auto& data = gpuData_.dataForCurrentDeviceAsync(cudaStream, [this](GPUData& data, cuda::stream_t<>& stream) {
+ cudaCheck(cudaMalloc((void**) & data.gainForHLTonGPU, sizeof(SiPixelGainForHLTonGPU)));
+ cudaCheck(cudaMalloc((void**) & data.gainDataOnGPU, this->gains_->data().size())); // TODO: this could be changed to cuda::memory::device::unique_ptr<>
+ // gains.data().data() is used also for non-GPU code, we cannot allocate it on aligned and write-combined memory
+ cudaCheck(cudaMemcpyAsync(data.gainDataOnGPU, this->gains_->data().data(), this->gains_->data().size(), cudaMemcpyDefault, stream.id()));
+
+ cudaCheck(cudaMemcpyAsync(data.gainForHLTonGPU, this->gainForHLTonHost_, sizeof(SiPixelGainForHLTonGPU), cudaMemcpyDefault, stream.id()));
+ cudaCheck(cudaMemcpyAsync(&(data.gainForHLTonGPU->v_pedestals), &(data.gainDataOnGPU), sizeof(SiPixelGainForHLTonGPU_DecodingStructure*), cudaMemcpyDefault, stream.id()));
+ });
+ return data.gainForHLTonGPU;
+}
diff --git a/Calibration/LumiAlCaRecoProducers/python/ALCARECOAlCaPCCRandom_cff.py b/Calibration/LumiAlCaRecoProducers/python/ALCARECOAlCaPCCRandom_cff.py
index 8185c8cfbb089..9161b2152ade7 100644
--- a/Calibration/LumiAlCaRecoProducers/python/ALCARECOAlCaPCCRandom_cff.py
+++ b/Calibration/LumiAlCaRecoProducers/python/ALCARECOAlCaPCCRandom_cff.py
@@ -19,8 +19,8 @@
from Calibration.LumiAlCaRecoProducers.alcaPCCProducer_cfi import alcaPCCProducer
alcaPCCProducerRandom = alcaPCCProducer.clone()
-alcaPCCProducerRandom.pixelClusterLabel = cms.InputTag("siPixelClustersForLumiR")
-alcaPCCProducerRandom.trigstring = cms.untracked.string("alcaPCCRandom")
+alcaPCCProducerRandom.AlcaPCCProducerParameters.pixelClusterLabel = cms.InputTag("siPixelClustersForLumiR")
+alcaPCCProducerRandom.AlcaPCCProducerParameters.trigstring = cms.untracked.string("alcaPCCRandom")
# Sequence #
seqALCARECOAlCaPCCRandom = cms.Sequence(ALCARECORandomHLT + siPixelDigisForLumiR + siPixelClustersForLumiR + alcaPCCProducerRandom)
diff --git a/Calibration/LumiAlCaRecoProducers/python/ALCARECOAlCaPCCZeroBias_cff.py b/Calibration/LumiAlCaRecoProducers/python/ALCARECOAlCaPCCZeroBias_cff.py
index 0ef9e074cc817..6a66256ae72ca 100644
--- a/Calibration/LumiAlCaRecoProducers/python/ALCARECOAlCaPCCZeroBias_cff.py
+++ b/Calibration/LumiAlCaRecoProducers/python/ALCARECOAlCaPCCZeroBias_cff.py
@@ -19,8 +19,8 @@
from Calibration.LumiAlCaRecoProducers.alcaPCCProducer_cfi import alcaPCCProducer
alcaPCCProducerZeroBias = alcaPCCProducer.clone()
-alcaPCCProducerZeroBias.pixelClusterLabel = cms.InputTag("siPixelClustersForLumiZB")
-alcaPCCProducerZeroBias.trigstring = cms.untracked.string("alcaPCCZeroBias")
+alcaPCCProducerZeroBias.AlcaPCCProducerParameters.pixelClusterLabel = cms.InputTag("siPixelClustersForLumiZB")
+alcaPCCProducerZeroBias.AlcaPCCProducerParameters.trigstring = cms.untracked.string("alcaPCCZeroBias")
# Sequence #
seqALCARECOAlCaPCCZeroBias = cms.Sequence(ALCARECOZeroBiasHLT + siPixelDigisForLumiZB + siPixelClustersForLumiZB + alcaPCCProducerZeroBias)
diff --git a/CondFormats/SiPixelObjects/interface/SiPixelGainForHLTonGPU.h b/CondFormats/SiPixelObjects/interface/SiPixelGainForHLTonGPU.h
new file mode 100644
index 0000000000000..931ee7e65f295
--- /dev/null
+++ b/CondFormats/SiPixelObjects/interface/SiPixelGainForHLTonGPU.h
@@ -0,0 +1,73 @@
+#ifndef CondFormats_SiPixelObjects_SiPixelGainForHLTonGPU_h
+#define CondFormats_SiPixelObjects_SiPixelGainForHLTonGPU_h
+
+#include
+#include
+#include
+
+#include "HeterogeneousCore/CUDAUtilities/interface/cuda_assert.h"
+
+struct SiPixelGainForHLTonGPU_DecodingStructure{
+ uint8_t gain;
+ uint8_t ped;
+};
+
+
+// copy of SiPixelGainCalibrationForHLT
+class SiPixelGainForHLTonGPU {
+
+ public:
+
+ using DecodingStructure = SiPixelGainForHLTonGPU_DecodingStructure;
+
+ using Range = std::pair;
+
+
+ inline __host__ __device__
+ std::pair getPedAndGain(uint32_t moduleInd, int col, int row, bool& isDeadColumn, bool& isNoisyColumn ) const {
+
+
+ auto range = rangeAndCols[moduleInd].first;
+ auto nCols = rangeAndCols[moduleInd].second;
+
+ // determine what averaged data block we are in (there should be 1 or 2 of these depending on if plaquette is 1 by X or 2 by X
+ unsigned int lengthOfColumnData = (range.second-range.first)/nCols;
+ unsigned int lengthOfAveragedDataInEachColumn = 2; // we always only have two values per column averaged block
+ unsigned int numberOfDataBlocksToSkip = row / numberOfRowsAveragedOver_;
+
+
+ auto offset = range.first + col*lengthOfColumnData + lengthOfAveragedDataInEachColumn*numberOfDataBlocksToSkip;
+
+ assert(offset rangeAndCols[2000];
+
+ float minPed_, maxPed_, minGain_, maxGain_;
+
+ float pedPrecision, gainPrecision;
+
+ unsigned int numberOfRowsAveragedOver_; // this is 80!!!!
+ unsigned int nBinsToUseForEncoding_;
+ unsigned int deadFlag_;
+ unsigned int noisyFlag_;
+};
+
+#endif // CondFormats_SiPixelObjects_SiPixelGainForHLTonGPU_h
diff --git a/Configuration/StandardSequences/python/RawToDigi_cff.py b/Configuration/StandardSequences/python/RawToDigi_cff.py
index d0af52de00a47..ed10c78a40c9b 100644
--- a/Configuration/StandardSequences/python/RawToDigi_cff.py
+++ b/Configuration/StandardSequences/python/RawToDigi_cff.py
@@ -3,13 +3,14 @@
# This object is used to selectively make changes for different running
# scenarios. In this case it makes changes for Run 2.
-from EventFilter.SiPixelRawToDigi.SiPixelRawToDigi_cfi import *
+from EventFilter.SiPixelRawToDigi.siPixelDigis_cff import *
from EventFilter.SiStripRawToDigi.SiStripDigis_cfi import *
from SimCalorimetry.EcalTrigPrimProducers.ecalTriggerPrimitiveDigis_cff import *
-from EventFilter.EcalRawToDigi.ecalDigis_cff import *
+import EventFilter.EcalRawToDigi.EcalUnpackerData_cfi
+ecalDigis = EventFilter.EcalRawToDigi.EcalUnpackerData_cfi.ecalEBunpacker.clone()
import EventFilter.ESRawToDigi.esRawToDigi_cfi
ecalPreshowerDigis = EventFilter.ESRawToDigi.esRawToDigi_cfi.esRawToDigi.clone()
@@ -23,8 +24,11 @@
import EventFilter.DTRawToDigi.dtunpacker_cfi
muonDTDigis = EventFilter.DTRawToDigi.dtunpacker_cfi.muonDTDigis.clone()
-import EventFilter.RPCRawToDigi.RPCRawToDigi_cfi
-muonRPCDigis = EventFilter.RPCRawToDigi.RPCRawToDigi_cfi.muonRPCDigis.clone()
+import EventFilter.RPCRawToDigi.rpcUnpacker_cfi
+muonRPCDigis = EventFilter.RPCRawToDigi.rpcUnpacker_cfi.rpcunpacker.clone()
+
+import EventFilter.RPCRawToDigi.rpcDigiMerger_cfi
+muonRPCNewDigis = EventFilter.RPCRawToDigi.rpcDigiMerger_cfi.rpcDigiMerger.clone()
import EventFilter.GEMRawToDigi.muonGEMDigis_cfi
muonGEMDigis = EventFilter.GEMRawToDigi.muonGEMDigis_cfi.muonGEMDigis.clone()
@@ -45,9 +49,9 @@
from EventFilter.CTPPSRawToDigi.ctppsRawToDigi_cff import *
RawToDigiTask = cms.Task(L1TRawToDigiTask,
- siPixelDigis,
+ siPixelDigisTask,
siStripDigis,
- ecalDigisTask,
+ ecalDigis,
ecalPreshowerDigis,
hcalDigis,
muonCSCDigis,
@@ -60,20 +64,15 @@
)
RawToDigi = cms.Sequence(RawToDigiTask)
-RawToDigiTask_noTk = RawToDigiTask.copyAndExclude([siPixelDigis, siStripDigis])
+RawToDigiTask_noTk = RawToDigiTask.copyAndExclude([siPixelDigisTask, siStripDigis])
RawToDigi_noTk = cms.Sequence(RawToDigiTask_noTk)
-RawToDigiTask_pixelOnly = cms.Task(siPixelDigis)
+RawToDigiTask_pixelOnly = cms.Task(siPixelDigisTask)
RawToDigi_pixelOnly = cms.Sequence(RawToDigiTask_pixelOnly)
-RawToDigiTask_ecalOnly = cms.Task(ecalDigisTask, ecalPreshowerDigis, scalersRawToDigi)
-RawToDigi_ecalOnly = cms.Sequence(RawToDigiTask_ecalOnly)
-
-RawToDigiTask_hcalOnly = cms.Task(hcalDigis)
-RawToDigi_hcalOnly = cms.Sequence(RawToDigiTask_hcalOnly)
-
scalersRawToDigi.scalersInputTag = 'rawDataCollector'
-siPixelDigis.InputLabel = 'rawDataCollector'
+siPixelDigis.cpu.InputLabel = 'rawDataCollector'
+#false by default anyways ecalDigis.DoRegional = False
ecalDigis.InputLabel = 'rawDataCollector'
ecalPreshowerDigis.sourceTag = 'rawDataCollector'
hcalDigis.InputLabel = 'rawDataCollector'
@@ -117,6 +116,19 @@
from Configuration.Eras.Modifier_phase2_hgcal_cff import phase2_hgcal
phase2_hgcal.toReplaceWith(RawToDigiTask,_hgcal_RawToDigiTask)
+# RPC New Readout Validation
+from Configuration.Eras.Modifier_stage2L1Trigger_2017_cff import stage2L1Trigger_2017
+_rpc_NewReadoutVal_RawToDigiTask = RawToDigiTask.copy()
+_rpc_NewReadoutVal_RawToDigiTask_noTk = RawToDigiTask_noTk.copy()
+_rpc_NewReadoutVal_RawToDigiTask.add(muonRPCNewDigis)
+_rpc_NewReadoutVal_RawToDigiTask_noTk.add(muonRPCNewDigis)
+stage2L1Trigger_2017.toReplaceWith(RawToDigiTask, _rpc_NewReadoutVal_RawToDigiTask)
+stage2L1Trigger_2017.toReplaceWith(RawToDigiTask_noTk, _rpc_NewReadoutVal_RawToDigiTask)
+
+from Configuration.Eras.Modifier_fastSim_cff import fastSim
+fastSim.toReplaceWith(RawToDigiTask, RawToDigiTask.copyAndExclude([muonRPCNewDigis]))
+fastSim.toReplaceWith(RawToDigiTask_noTk, RawToDigiTask_noTk.copyAndExclude([muonRPCNewDigis]))
+
_hfnose_RawToDigiTask = RawToDigiTask.copy()
_hfnose_RawToDigiTask.add(hfnoseDigis)
diff --git a/DQM/Integration/python/clients/beam_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/beam_dqm_sourceclient-live_cfg.py
index 4f6ad7bd592ba..5bc3b2bf8af63 100644
--- a/DQM/Integration/python/clients/beam_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/beam_dqm_sourceclient-live_cfg.py
@@ -1,15 +1,9 @@
from __future__ import print_function
import FWCore.ParameterSet.Config as cms
+from Configuration.StandardSequences.Eras import eras
-# Define here the BeamSpotOnline record name,
-# it will be used both in BeamMonitor setup and in payload creation/upload
-BSOnlineRecordName = 'BeamSpotOnlineLegacyObjectsRcd'
-
-#from Configuration.Eras.Era_Run2_2018_cff import Run2_2018
-#process = cms.Process("BeamMonitor", Run2_2018) FIXME
-import sys
-from Configuration.Eras.Era_Run2_2018_pp_on_AA_cff import Run2_2018_pp_on_AA
-process = cms.Process("BeamMonitor", Run2_2018_pp_on_AA)
+#process = cms.Process("BeamMonitor", eras.Run2_2018) FIXME
+process = cms.Process("BeamMonitor", eras.Run2_2018_pp_on_AA)
#
process.MessageLogger = cms.Service("MessageLogger",
@@ -22,29 +16,13 @@
# switch
live = True # FIXME
-unitTest = False
-
-if 'unitTest=True' in sys.argv:
- live=False
- unitTest=True
-
-# Switch to veto the upload of the BeamSpot conditions to the DB
-# when False it performs the upload
-noDB = True
-if 'noDB=False' in sys.argv:
- noDB=False
#---------------
# Input sources
-if unitTest:
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-elif live:
+if (live):
process.load("DQM.Integration.config.inputsource_cfi")
- from DQM.Integration.config.inputsource_cfi import options
else:
process.load("DQM.Integration.config.fileinputsource_cfi")
- from DQM.Integration.config.fileinputsource_cfi import options
#--------------------------
# HLT Filter
@@ -57,9 +35,6 @@
process.load("DQM.Integration.config.environment_cfi")
process.dqmEnv.subSystemFolder = 'BeamMonitor'
process.dqmSaver.tag = 'BeamMonitor'
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = 'BeamMonitor'
-process.dqmSaverPB.runNumber = options.runNumber
process.dqmEnvPixelLess = process.dqmEnv.clone()
process.dqmEnvPixelLess.subSystemFolder = 'BeamMonitor_PixelLess'
@@ -73,7 +48,7 @@
from Configuration.AlCa.GlobalTag import GlobalTag as gtCustomise
process.GlobalTag = gtCustomise(process.GlobalTag, 'auto:run2_data', '')
# you may need to set manually the GT in the line below
- #process.GlobalTag.globaltag = '100X_upgrade2018_realistic_v10'
+ process.GlobalTag.globaltag = '100X_upgrade2018_realistic_v10'
#----------------------------
# BeamMonitor
@@ -231,7 +206,7 @@
#
process.dqmcommon = cms.Sequence(process.dqmEnv
- * process.dqmSaver*process.dqmSaverPB)
+ * process.dqmSaver)
#
process.monitor = cms.Sequence(process.dqmBeamMonitor
@@ -249,8 +224,7 @@
process.dqmBeamSpotProblemMonitor.pixelTracks = 'pixelTracks'
#
-from DQMServices.Core.DQMQualityTester import DQMQualityTester
-process.qTester = DQMQualityTester(
+process.qTester = cms.EDAnalyzer("QualityTester",
qtList = cms.untracked.FileInPath('DQM/BeamMonitor/test/BeamSpotAvailableTest.xml'),
prescaleFactor = cms.untracked.int32(1),
qtestOnEndLumi = cms.untracked.bool(True),
@@ -297,7 +271,6 @@
process.load("RecoVertex.BeamSpotProducer.BeamSpot_cfi")
process.dqmBeamMonitor.OnlineMode = True
-process.dqmBeamMonitor.recordName = BSOnlineRecordName
process.dqmBeamMonitor.resetEveryNLumi = 5 # was 10 for HI
process.dqmBeamMonitor.resetPVEveryNLumi = 5 # was 10 for HI
@@ -350,40 +323,6 @@
process.dqmBeamMonitor.hltResults = cms.InputTag("TriggerResults","","HLT")
-#---------
-# Upload BeamSpotOnlineObject (LegacyRcd) to CondDB
-process.OnlineDBOutputService = cms.Service("OnlineDBOutputService",
-
- DBParameters = cms.PSet(
- messageLevel = cms.untracked.int32(0),
- authenticationPath = cms.untracked.string('.')
- ),
-
- # Upload to CondDB
- connect = cms.string('oracle://cms_orcoff_prep/CMS_CONDITIONS'),
- preLoadConnectionString = cms.untracked.string('frontier://FrontierPrep/CMS_CONDITIONS'),
-
- runNumber = cms.untracked.uint64(options.runNumber),
- lastLumiFile = cms.untracked.string(''),
- writeTransactionDelay = cms.untracked.uint32(options.transDelay),
- latency = cms.untracked.uint32(2),
- autoCommit = cms.untracked.bool(True),
- saveLogsOnDB = cms.untracked.bool(True),
- jobName = cms.untracked.string("BeamSpotOnlineLegacyTest"), # name of the DB log record
- toPut = cms.VPSet(cms.PSet(
- record = cms.string(BSOnlineRecordName),
- tag = cms.string('BSOnlineLegacy_tag'),
- timetype = cms.untracked.string('Lumi'),
- onlyAppendUpdatePolicy = cms.untracked.bool(True)
- ))
-)
-
-# If not live or noDB: produce a (local) SQLITE file
-if not live or noDB:
- process.OnlineDBOutputService.connect = cms.string('sqlite_file:BeamSpotOnlineLegacy.db')
- process.OnlineDBOutputService.preLoadConnectionString = cms.untracked.string('sqlite_file:BeamSpotOnlineLegacy.db')
- process.OnlineDBOutputService.saveLogsOnDB = cms.untracked.bool(False)
-
#---------
# Final path
if (not process.runType.getRunType() == process.runType.hi_run):
diff --git a/DQM/Integration/python/clients/beampixel_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/beampixel_dqm_sourceclient-live_cfg.py
index 75f0545a5c5ba..7b4115f74ad84 100644
--- a/DQM/Integration/python/clients/beampixel_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/beampixel_dqm_sourceclient-live_cfg.py
@@ -1,27 +1,17 @@
from __future__ import print_function
import FWCore.ParameterSet.Config as cms
+from Configuration.StandardSequences.Eras import eras
-import sys
-from Configuration.Eras.Era_Run2_2018_cff import Run2_2018
-process = cms.Process("BeamPixel", Run2_2018)
-
-unitTest = False
-if 'unitTest=True' in sys.argv:
- unitTest = True
+process = cms.Process("BeamPixel", eras.Run2_2018)
#----------------------------
# Common for PP and HI running
#----------------------------
-if unitTest == True:
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-else:
- process.load("DQM.Integration.config.inputsource_cfi")
- from DQM.Integration.config.inputsource_cfi import options
# Use this to run locally (for testing purposes)
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
+# Otherwise use this
+process.load("DQM.Integration.config.inputsource_cfi")
#----------------------------
@@ -37,9 +27,7 @@
process.load("DQM.Integration.config.environment_cfi")
process.dqmEnv.subSystemFolder = "BeamPixel"
process.dqmSaver.tag = "BeamPixel"
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = 'BeamPixel'
-process.dqmSaverPB.runNumber = options.runNumber
+
#----------------------------
# Conditions
@@ -63,7 +51,7 @@
#----------------------------
# Define Sequences
#----------------------------
-process.dqmModules = cms.Sequence(process.dqmEnv + process.dqmSaver + process.dqmSaverPB)
+process.dqmModules = cms.Sequence(process.dqmEnv + process.dqmSaver)
process.physTrigger = cms.Sequence(process.hltTriggerTypeFilter)
diff --git a/DQM/Integration/python/clients/csc_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/csc_dqm_sourceclient-live_cfg.py
index 77b1e6b88b699..e14e1dde331c7 100644
--- a/DQM/Integration/python/clients/csc_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/csc_dqm_sourceclient-live_cfg.py
@@ -1,6 +1,5 @@
from __future__ import print_function
import FWCore.ParameterSet.Config as cms
-import sys
process = cms.Process("CSCDQMLIVE")
@@ -33,22 +32,11 @@
#----------------------------
# Event Source
#-----------------------------
-
-unitTest=False
-if 'unitTest=True' in sys.argv:
- unitTest=True
-
-if unitTest:
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-else:
- # for live online DQM in P5
- process.load("DQM.Integration.config.inputsource_cfi")
- from DQM.Integration.config.inputsource_cfi import options
+# for live online DQM in P5
+process.load("DQM.Integration.config.inputsource_cfi")
# for testing in lxplus
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
#----------------------------
# DQM Environment
@@ -61,10 +49,8 @@
process.load("DQM.Integration.config.environment_cfi")
process.dqmEnv.subSystemFolder = "CSC"
process.dqmSaver.tag = "CSC"
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = "CSC"
-process.dqmSaverPB.runNumber = options.runNumber
+process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/csc_reference.root'
#process.DQM.collectorHost = 'pccmsdqm02.cern.ch'
#process.DQM.collectorHost = 'localhost'
@@ -172,8 +158,8 @@
# Sequences
#--------------------------
-#process.p = cms.Path(process.dqmCSCClient+process.dqmEnv+process.dqmSaver+process.dqmSaverPB)
-process.p = cms.Path(process.dqmCSCClient * process.muonCSCDigis * process.csc2DRecHits * process.cscSegments * process.cscMonitor + process.dqmEnv + process.dqmSaver + process.dqmSaverPB)
+#process.p = cms.Path(process.dqmCSCClient+process.dqmEnv+process.dqmSaver)
+process.p = cms.Path(process.dqmCSCClient * process.muonCSCDigis * process.csc2DRecHits * process.cscSegments * process.cscMonitor + process.dqmEnv + process.dqmSaver)
process.castorDigis.InputLabel = cms.InputTag("rawDataCollector")
diff --git a/DQM/Integration/python/clients/fed_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/fed_dqm_sourceclient-live_cfg.py
index 6feff0263b749..db5d48aa3f106 100644
--- a/DQM/Integration/python/clients/fed_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/fed_dqm_sourceclient-live_cfg.py
@@ -1,13 +1,8 @@
import FWCore.ParameterSet.Config as cms
-import sys
# Process initialization
process = cms.Process('FED')
-unitTest = False
-if 'unitTest=True' in sys.argv:
- unitTest=True
-
# Logging:
process.MessageLogger = cms.Service(
'MessageLogger',
@@ -23,18 +18,10 @@
# Global tag:
process.load('DQM.Integration.config.FrontierCondition_GT_cfi')
# Input:
-if unitTest:
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-else:
- process.load('DQM.Integration.config.inputsource_cfi')
- from DQM.Integration.config.inputsource_cfi import options
+process.load('DQM.Integration.config.inputsource_cfi')
# Output:
process.dqmEnv.subSystemFolder = 'FED'
process.dqmSaver.tag = 'FED'
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = 'FED'
-process.dqmSaverPB.runNumber = options.runNumber
# Subsystem sequences
@@ -75,10 +62,11 @@
process.load('EventFilter.HcalRawToDigi.HcalRawToDigi_cfi')
# DT sequence:
process.load('DQM.DTMonitorModule.dtDataIntegrityTask_EvF_cff')
-process.dtDataIntegrityTask.processingMode = 'SM'
+process.DTDataIntegrityTask.processingMode = 'SM'
path = 'DT/%s/' % folder_name
-process.dtDataIntegrityTask.fedIntegrityFolder = path
-process.dtDataIntegrityTask.dtFEDlabel = 'dtunpacker'
+process.DTDataIntegrityTask.fedIntegrityFolder = path
+process.DTDataIntegrityTask.checkUros = True
+process.DTDataIntegrityTask.dtFEDlabel = 'dtunpacker'
# RPC sequence:
process.load('EventFilter.RPCRawToDigi.rpcUnpacker_cfi')
process.load('DQM.RPCMonitorClient.RPCFEDIntegrity_cfi')
@@ -136,7 +124,7 @@
+ process.hcalDigis
+ process.cscDQMEvF
+ process.dtunpacker
- + process.dtDataIntegrityTask
+ + process.DTDataIntegrityTask
+ process.rpcunpacker
+ process.rpcFEDIntegrity
@@ -147,7 +135,6 @@
process.DQMmodulesPath = cms.Path(
process.dqmEnv
+ process.dqmSaver
- + process.dqmSaverPB
)
process.schedule = cms.Schedule(
diff --git a/DQM/Integration/python/clients/l1t_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/l1t_dqm_sourceclient-live_cfg.py
index 301a79f2b2865..b50793e7a5e78 100644
--- a/DQM/Integration/python/clients/l1t_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/l1t_dqm_sourceclient-live_cfg.py
@@ -16,11 +16,9 @@
#
# for live online DQM in P5
process.load("DQM.Integration.config.inputsource_cfi")
-from DQM.Integration.config.inputsource_cfi import options
#
# for testing in lxplus
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
#----------------------------
# DQM Environment
@@ -28,12 +26,10 @@
process.load("DQM.Integration.config.environment_cfi")
process.dqmEnv.subSystemFolder = 'L1T'
process.dqmSaver.tag = 'L1T'
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = 'L1T'
-process.dqmSaverPB.runNumber = options.runNumber
#
# references needed
+process.DQMStore.referenceFileName = "/dqmdata/dqm/reference/l1t_reference.root"
# Condition for P5 cluster
process.load("DQM.Integration.config.FrontierCondition_GT_cfi")
@@ -101,8 +97,7 @@
#
process.dqmEndPath = cms.EndPath(
process.dqmEnv *
- process.dqmSaver *
- process.dqmSaverPB
+ process.dqmSaver
)
#
diff --git a/DQM/Integration/python/clients/l1temulator_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/l1temulator_dqm_sourceclient-live_cfg.py
index 9701c71d14a3c..5ddb7506f1e6e 100644
--- a/DQM/Integration/python/clients/l1temulator_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/l1temulator_dqm_sourceclient-live_cfg.py
@@ -16,11 +16,9 @@
#
# for live online DQM in P5
process.load("DQM.Integration.config.inputsource_cfi")
-from DQM.Integration.config.inputsource_cfi import options
#
# for testing in lxplus
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
#----------------------------
# DQM Environment
@@ -32,11 +30,9 @@
# for local test
process.dqmEnv.subSystemFolder = 'L1TEMU'
process.dqmSaver.tag = 'L1TEMU'
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = 'L1TEMU'
-process.dqmSaverPB.runNumber = options.runNumber
#
# no references needed
+# replace DQMStore.referenceFileName = "L1TEMU_reference.root"
#
# Condition for P5 cluster
@@ -93,7 +89,7 @@
process.l1EmulatorMonitorClientPath = cms.Path(process.l1EmulatorMonitorClient)
#
-process.l1EmulatorMonitorEndPath = cms.EndPath(process.dqmEnv*process.dqmSaver*process.dqmSaverPB)
+process.l1EmulatorMonitorEndPath = cms.EndPath(process.dqmEnv*process.dqmSaver)
#
process.valCscTriggerPrimitiveDigis.gangedME1a = cms.untracked.bool(False)
diff --git a/DQM/Integration/python/clients/l1tstage1_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/l1tstage1_dqm_sourceclient-live_cfg.py
index b8ae0bcf233b4..20371cf54497f 100644
--- a/DQM/Integration/python/clients/l1tstage1_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/l1tstage1_dqm_sourceclient-live_cfg.py
@@ -16,11 +16,9 @@
#
# for live online DQM in P5
process.load("DQM.Integration.config.inputsource_cfi")
-from DQM.Integration.config.inputsource_cfi import options
#
# for testing in lxplus
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
#----------------------------
# DQM Environment
@@ -29,12 +27,10 @@
process.load("DQM.Integration.config.environment_cfi")
process.dqmEnv.subSystemFolder = 'L1TStage1'
process.dqmSaver.tag = 'L1TStage1'
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = 'L1TStage1'
-process.dqmSaverPB.runNumber = options.runNumber
#
# references needed
+process.DQMStore.referenceFileName = "/dqmdata/dqm/reference/l1t_reference.root"
# Condition for P5 cluster
process.load("DQM.Integration.config.FrontierCondition_GT_cfi")
@@ -106,8 +102,7 @@
#
process.dqmEndPath = cms.EndPath(
process.dqmEnv *
- process.dqmSaver *
- process.dqmSaverPB
+ process.dqmSaver
)
#
diff --git a/DQM/Integration/python/clients/l1tstage1emulator_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/l1tstage1emulator_dqm_sourceclient-live_cfg.py
index 6200618c0fe44..c7b24f6137478 100644
--- a/DQM/Integration/python/clients/l1tstage1emulator_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/l1tstage1emulator_dqm_sourceclient-live_cfg.py
@@ -16,11 +16,9 @@
#
# for live online DQM in P5
process.load("DQM.Integration.config.inputsource_cfi")
-from DQM.Integration.config.inputsource_cfi import options
#
# for testing in lxplus
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
#----------------------------
# DQM Environment
@@ -30,12 +28,10 @@
# for local test
process.dqmEnv.subSystemFolder = 'L1TEMUStage1'
process.dqmSaver.tag = 'L1TEMUStage1'
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = 'L1TEMUStage1'
-process.dqmSaverPB.runNumber = options.runNumber
#
# no references needed
+# replace DQMStore.referenceFileName = "L1TEMU_reference.root"
#
# Condition for P5 cluster
@@ -93,7 +89,7 @@
process.l1EmulatorMonitorClientPath = cms.Path(process.l1EmulatorMonitorClient)
#
-process.l1EmulatorMonitorEndPath = cms.EndPath(process.dqmEnv*process.dqmSaver*process.dqmSaverPB)
+process.l1EmulatorMonitorEndPath = cms.EndPath(process.dqmEnv*process.dqmSaver)
#
diff --git a/DQM/Integration/python/clients/l1tstage2_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/l1tstage2_dqm_sourceclient-live_cfg.py
index c5788228cac1f..a6362a54ffc17 100644
--- a/DQM/Integration/python/clients/l1tstage2_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/l1tstage2_dqm_sourceclient-live_cfg.py
@@ -1,27 +1,16 @@
import FWCore.ParameterSet.Config as cms
-import sys
-from Configuration.Eras.Era_Run3_cff import Run3
-process = cms.Process("L1TStage2DQM", Run3)
-
-unitTest = False
-if 'unitTest=True' in sys.argv:
- unitTest=True
+from Configuration.StandardSequences.Eras import eras
+process = cms.Process("L1TStage2DQM", eras.Run2_2018)
#--------------------------------------------------
# Event Source and Condition
-if unitTest:
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-else:
- # Live Online DQM in P5
- process.load("DQM.Integration.config.inputsource_cfi")
- from DQM.Integration.config.inputsource_cfi import options
+# Live Online DQM in P5
+process.load("DQM.Integration.config.inputsource_cfi")
# # Testing in lxplus
# process.load("DQM.Integration.config.fileinputsource_cfi")
-# from DQM.Integration.config.fileinputsource_cfi import options
# process.load("FWCore.MessageLogger.MessageLogger_cfi")
# process.MessageLogger.cerr.FwkReport.reportEvery = 1
@@ -42,11 +31,9 @@
process.dqmEnv.subSystemFolder = "L1T"
process.dqmSaver.tag = "L1T"
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = "L1T"
-process.dqmSaverPB.runNumber = options.runNumber
+process.DQMStore.referenceFileName = "/dqmdata/dqm/reference/l1t_reference.root"
-process.dqmEndPath = cms.EndPath(process.dqmEnv * process.dqmSaver * process.dqmSaverPB)
+process.dqmEndPath = cms.EndPath(process.dqmEnv * process.dqmSaver)
#--------------------------------------------------
# Standard Unpacking Path
@@ -115,6 +102,7 @@
# Cosmic run
if (process.runType.getRunType() == process.runType.cosmic_run):
+ process.DQMStore.referenceFileName = "/dqmdata/dqm/reference/l1t_reference_cosmic.root"
# Remove Quality Tests for L1T Muon Subsystems since they are not optimized yet for cosmics
process.l1tStage2MonitorClient.remove(process.l1TStage2uGMTQualityTests)
process.l1tStage2MonitorClient.remove(process.l1TStage2EMTFQualityTests)
@@ -125,6 +113,7 @@
# Heavy-Ion run
if (process.runType.getRunType() == process.runType.hi_run):
+ process.DQMStore.referenceFileName = "/dqmdata/dqm/reference/l1t_reference_hi.root"
process.onlineMetaDataDigis.onlineMetaDataInputLabel = cms.InputTag("rawDataRepacker")
process.onlineMetaDataRawToDigi.onlineMetaDataInputLabel = cms.InputTag("rawDataRepacker")
process.castorDigis.InputLabel = cms.InputTag("rawDataRepacker")
diff --git a/DQM/Integration/python/clients/l1tstage2emulator_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/l1tstage2emulator_dqm_sourceclient-live_cfg.py
index 91f1f564366d0..614dbb3cfd59d 100644
--- a/DQM/Integration/python/clients/l1tstage2emulator_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/l1tstage2emulator_dqm_sourceclient-live_cfg.py
@@ -1,27 +1,16 @@
import FWCore.ParameterSet.Config as cms
-import sys
-from Configuration.Eras.Era_Run3_cff import Run3
-process = cms.Process("L1TStage2EmulatorDQM", Run3)
-
-unitTest = False
-if 'unitTest=True' in sys.argv:
- unitTest=True
+from Configuration.StandardSequences.Eras import eras
+process = cms.Process("L1TStage2EmulatorDQM", eras.Run2_2018)
#--------------------------------------------------
# Event Source and Condition
-if unitTest:
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-else:
- # Live Online DQM in P5
- process.load("DQM.Integration.config.inputsource_cfi")
- from DQM.Integration.config.inputsource_cfi import options
+# Live Online DQM in P5
+process.load("DQM.Integration.config.inputsource_cfi")
# Testing in lxplus
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
# Required to load Global Tag
process.load("DQM.Integration.config.FrontierCondition_GT_cfi")
@@ -38,14 +27,11 @@
process.dqmEnv.subSystemFolder = "L1TEMU"
process.dqmSaver.tag = "L1TEMU"
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = "L1TEMU"
-process.dqmSaverPB.runNumber = options.runNumber
+process.DQMStore.referenceFileName = "/dqmdata/dqm/reference/l1temu_reference.root"
process.dqmEndPath = cms.EndPath(
process.dqmEnv *
- process.dqmSaver *
- process.dqmSaverPB
+ process.dqmSaver
)
#--------------------------------------------------
diff --git a/DQM/Integration/python/clients/lumi_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/lumi_dqm_sourceclient-live_cfg.py
index 47acfcdc471f2..c38b7dd5fe9c0 100644
--- a/DQM/Integration/python/clients/lumi_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/lumi_dqm_sourceclient-live_cfg.py
@@ -8,7 +8,6 @@
# Event Source
#----------------------------
process.load("DQM.Integration.config.inputsource_cfi")
-from DQM.Integration.config.inputsource_cfi import options
#process.DQMEventStreamHttpReader.consumerName = 'DQM Luminosity Consumer'
#process.DQMEventStreamHttpReader.SelectHLTOutput = cms.untracked.string('hltOutputALCALUMIPIXELS')
@@ -18,9 +17,6 @@
process.load("DQM.Integration.config.environment_cfi")
process.dqmEnv.subSystemFolder = "Info/Lumi"
process.dqmSaver.tag = "Lumi"
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = "Lumi"
-process.dqmSaverPB.runNumber = options.runNumber
#---------------------------------------------
# Global Tag
@@ -67,8 +63,7 @@
process.dqmmodules = cms.Sequence(process.dqmEnv
+ process.expressLumiProducer
+ process.dqmLumiMonitor
- + process.dqmSaver
- + process.dqmSaverPB)
+ + process.dqmSaver)
#----------------------------
# Proton-Proton Running Stuff
#----------------------------
diff --git a/DQM/Integration/python/clients/physics_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/physics_dqm_sourceclient-live_cfg.py
index 76b6ae553949c..d7a35b425baab 100644
--- a/DQM/Integration/python/clients/physics_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/physics_dqm_sourceclient-live_cfg.py
@@ -11,11 +11,9 @@
# for live online DQM in P5
process.load("DQM.Integration.config.inputsource_cfi")
-from DQM.Integration.config.inputsource_cfi import options
# for testing in lxplus
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
#----------------------------
# DQM Environment
@@ -24,9 +22,6 @@
process.load("DQM.Integration.config.environment_cfi")
process.dqmEnv.subSystemFolder = 'Physics'
process.dqmSaver.tag = 'Physics'
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = 'Physics'
-process.dqmSaverPB.runNumber = options.runNumber
# 0=random, 1=physics, 2=calibration, 3=technical
process.hltTriggerTypeFilter = cms.EDFilter("HLTTriggerTypeFilter",
@@ -52,8 +47,7 @@
# process.dump *
process.qcdLowPtDQM *
process.dqmEnv *
- process.dqmSaver *
- process.dqmSaverPB
+ process.dqmSaver
)
process.siPixelDigis.InputLabel = cms.InputTag("rawDataCollector")
diff --git a/DQM/Integration/python/clients/pixel_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/pixel_dqm_sourceclient-live_cfg.py
index 5c45b5f46ca9c..ebaa305eb8aae 100644
--- a/DQM/Integration/python/clients/pixel_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/pixel_dqm_sourceclient-live_cfg.py
@@ -1,17 +1,11 @@
from __future__ import print_function
import FWCore.ParameterSet.Config as cms
-import sys
-from Configuration.Eras.Era_Run3_cff import Run3
-process = cms.Process("PIXELDQMLIVE", Run3)
+from Configuration.StandardSequences.Eras import eras
-live=True
-unitTest = False
-
-if 'unitTest=True' in sys.argv:
- live=False
- unitTest=True
+process = cms.Process("PIXELDQMLIVE", eras.Run2_2018_pp_on_AA)
+live=True
#set to false for lxplus offline testing
#live=False
offlineTesting=not live
@@ -32,18 +26,12 @@
#-----------------------------
# for live online DQM in P5
-if (unitTest):
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-
-elif (live):
+if (live):
process.load("DQM.Integration.config.inputsource_cfi")
- from DQM.Integration.config.inputsource_cfi import options
# for testing in lxplus
elif(offlineTesting):
process.load("DQM.Integration.config.fileinputsource_cfi")
- from DQM.Integration.config.fileinputsource_cfi import options
#-----------------------------
# DQM Environment
@@ -59,16 +47,20 @@
process.dqmEnv.subSystemFolder = TAG
process.dqmSaver.tag = TAG
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = TAG
-process.dqmSaverPB.runNumber = options.runNumber
+process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/pixel_reference_pp.root'
+#if (process.runType.getRunType() == process.runType.hi_run):
+# process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/pixel_reference_hi.root'
+
+if (process.runType.getRunType() == process.runType.cosmic_run):
+ process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/pixel_reference_cosmic.root'
+
#-----------------------------
# Magnetic Field
#-----------------------------
-process.load('Configuration.StandardSequences.MagneticField_cff')
+process.load('Configuration.StandardSequences.MagneticField_AutoFromDBCurrent_cff')
#-------------------------------------------------
# GEOMETRY
@@ -117,6 +109,7 @@
process.siPixelDigis.InputLabel = cms.InputTag("rawDataCollector")
process.siStripDigis.InputLabel = cms.InputTag("rawDataCollector")
+
## Collision Reconstruction
process.load("Configuration.StandardSequences.RawToDigi_Data_cff")
@@ -165,7 +158,7 @@
# Scheduling
#--------------------------
-process.DQMmodules = cms.Sequence(process.dqmEnv* process.dqmSaver*process.dqmSaverPB)
+process.DQMmodules = cms.Sequence(process.dqmEnv* process.dqmSaver)
process.RecoForDQM_LocalReco = cms.Sequence(process.siPixelDigis*process.siStripDigis*process.gtDigis*process.trackerlocalreco)
@@ -185,7 +178,6 @@
##### TRIGGER SELECTION #####
process.hltHighLevel*
process.scalersRawToDigi*
- process.tcdsDigis*
process.APVPhases*
process.consecutiveHEs*
process.hltTriggerTypeFilter*
@@ -231,7 +223,6 @@
process.p = cms.Path(
process.hltHighLevel #trigger selection
*process.scalersRawToDigi
- *process.tcdsDigis
*process.APVPhases
*process.consecutiveHEs
*process.Reco
diff --git a/DQM/Integration/python/clients/pixellumi_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/pixellumi_dqm_sourceclient-live_cfg.py
index bd15690ede85a..cdc9302a927d0 100644
--- a/DQM/Integration/python/clients/pixellumi_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/pixellumi_dqm_sourceclient-live_cfg.py
@@ -1,13 +1,8 @@
from __future__ import print_function
import FWCore.ParameterSet.Config as cms
-import sys
process = cms.Process("PixelLumiDQM")
-unitTest=False
-if 'unitTest=True' in sys.argv:
- unitTest=True
-
process.MessageLogger = cms.Service("MessageLogger",
debugModules = cms.untracked.vstring('siPixelDigis',
'sipixelEDAClient'),
@@ -18,18 +13,11 @@
#----------------------------
# Event Source
#-----------------------------
-
-if unitTest:
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-else:
- # for live online DQM in P5
- process.load("DQM.Integration.config.inputsource_cfi")
- from DQM.Integration.config.inputsource_cfi import options
+# for live online DQM in P5
+process.load("DQM.Integration.config.inputsource_cfi")
# for testing in lxplus
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
##
#----------------------------
@@ -43,15 +31,13 @@
process.load("DQM.Integration.config.environment_cfi")
process.dqmEnv.subSystemFolder = "PixelLumi"
process.dqmSaver.tag = "PixelLumi"
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = "PixelLumi"
-process.dqmSaverPB.runNumber = options.runNumber
-if not unitTest:
- process.source.SelectEvents = cms.untracked.vstring("HLT_ZeroBias*","HLT_L1AlwaysTrue*", "HLT_PAZeroBias*", "HLT_PAL1AlwaysTrue*")
+process.source.SelectEvents = cms.untracked.vstring("HLT_ZeroBias*","HLT_L1AlwaysTrue*", "HLT_PAZeroBias*", "HLT_PAL1AlwaysTrue*")
+#process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/pixel_reference_pp.root'
#if (process.runType.getRunType() == process.runType.hi_run):
+# process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/pixel_reference_hi.root'
-if (process.runType.getRunType() == process.runType.cosmic_run and not unitTest):
+if (process.runType.getRunType() == process.runType.cosmic_run):
process.source.SelectEvents = cms.untracked.vstring('HLT*SingleMu*')
#----------------------------
@@ -100,8 +86,7 @@
process.load('Configuration.StandardSequences.RawToDigi_Repacked_cff')
process.siPixelDigis.InputLabel = cms.InputTag("rawDataRepacker")
- if not unitTest:
- process.source.SelectEvents = cms.untracked.vstring('HLT_HIL1MinimumBiasHF2AND*')
+ process.source.SelectEvents = cms.untracked.vstring('HLT_HIL1MinimumBiasHF2AND*')
# process.DQMEventStreamHttpReader.SelectEvents = cms.untracked.PSet(
@@ -134,8 +119,7 @@
process.Reco = cms.Sequence(process.siPixelDigis*process.siPixelClusters)
process.DQMmodules = cms.Sequence(process.dqmEnv*
process.pixel_lumi_dqm*
- process.dqmSaver*
- process.dqmSaverPB)
+ process.dqmSaver)
process.p = cms.Path(process.Reco*process.DQMmodules)
diff --git a/DQM/Integration/python/clients/scal_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/scal_dqm_sourceclient-live_cfg.py
index 893443f74f1a2..95a91979e9375 100644
--- a/DQM/Integration/python/clients/scal_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/scal_dqm_sourceclient-live_cfg.py
@@ -1,27 +1,15 @@
import FWCore.ParameterSet.Config as cms
-import sys
process = cms.Process("DQM")
-unitTest = False
-if 'unitTest=True' in sys.argv:
- unitTest=True
-
#----------------------------
#### Event Source
#----------------------------
-
-if unitTest:
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-else:
- # for live online DQM in P5
- process.load("DQM.Integration.config.inputsource_cfi")
- from DQM.Integration.config.inputsource_cfi import options
+# for live online DQM in P5
+process.load("DQM.Integration.config.inputsource_cfi")
# for testing in lxplus
#process.load("DQM.Integration.config.fileinputsource_cfi")
-#from DQM.Integration.config.fileinputsource_cfi import options
#----------------------------
#### DQM Environment
@@ -29,9 +17,6 @@
process.load("DQM.Integration.config.environment_cfi")
process.dqmEnv.subSystemFolder = 'Scal'
process.dqmSaver.tag = 'Scal'
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = 'Scal'
-process.dqmSaverPB.runNumber = options.runNumber
#-----------------------------
process.load("DQMServices.Components.DQMScalInfo_cfi")
@@ -53,7 +38,7 @@
import EventFilter.L1GlobalTriggerRawToDigi.l1GtEvmUnpack_cfi
gtEvmDigis = EventFilter.L1GlobalTriggerRawToDigi.l1GtEvmUnpack_cfi.l1GtEvmUnpack.clone()
-if (process.runType.getRunType() == process.runType.pp_run and not unitTest):
+if (process.runType.getRunType() == process.runType.pp_run):
process.source.SelectEvents = cms.untracked.vstring('HLT_ZeroBias*')
process.physicsBitSelector = cms.EDFilter("PhysDecl",
@@ -74,7 +59,7 @@
process.dump = cms.EDAnalyzer('EventContentAnalyzer')
# DQM Modules
-process.dqmmodules = cms.Sequence(process.dqmEnv + process.dqmSaver + process.dqmSaverPB)
+process.dqmmodules = cms.Sequence(process.dqmEnv + process.dqmSaver)
process.evfDQMmodulesPath = cms.Path(
process.l1GtUnpack*
process.gtDigis*
diff --git a/DQM/Integration/python/clients/sistrip_dqm_sourceclient-live_cfg.py b/DQM/Integration/python/clients/sistrip_dqm_sourceclient-live_cfg.py
index 37e219bd456c2..a9f8ef0e87d00 100644
--- a/DQM/Integration/python/clients/sistrip_dqm_sourceclient-live_cfg.py
+++ b/DQM/Integration/python/clients/sistrip_dqm_sourceclient-live_cfg.py
@@ -1,9 +1,9 @@
from __future__ import print_function
import FWCore.ParameterSet.Config as cms
-import sys
-from Configuration.Eras.Era_Run3_cff import Run3
-process = cms.Process("SiStrpDQMLive", Run3)
+from Configuration.StandardSequences.Eras import eras
+
+process = cms.Process("SiStrpDQMLive", eras.Run2_2018_pp_on_AA)
process.MessageLogger = cms.Service("MessageLogger",
debugModules = cms.untracked.vstring('siStripDigis',
@@ -15,12 +15,6 @@
)
live=True
-unitTest=False
-
-if 'unitTest=True' in sys.argv:
- live=False
- unitTest=True
-
# uncomment for running on lxplus
#live=False
offlineTesting=not live
@@ -30,16 +24,11 @@
# Event Source
#-----------------------------
# for live online DQM in P5
-if (unitTest):
- process.load("DQM.Integration.config.unittestinputsource_cfi")
- from DQM.Integration.config.unittestinputsource_cfi import options
-elif (live):
+if (live):
process.load("DQM.Integration.config.inputsource_cfi")
- from DQM.Integration.config.inputsource_cfi import options
# for testing in lxplus
elif(offlineTesting):
process.load("DQM.Integration.config.fileinputsource_cfi")
- from DQM.Integration.config.fileinputsource_cfi import options
#----------------------------
# DQM Live Environment
@@ -55,9 +44,6 @@
process.dqmEnv.subSystemFolder = "SiStrip"
process.dqmSaver.tag = "SiStrip"
process.dqmSaver.backupLumiCount = 30
-process.dqmSaver.runNumber = options.runNumber
-process.dqmSaverPB.tag = "SiStrip"
-process.dqmSaverPB.runNumber = options.runNumber
from DQMServices.Core.DQMEDAnalyzer import DQMEDAnalyzer
process.dqmEnvTr = DQMEDAnalyzer('DQMEventInfo',
@@ -70,7 +56,7 @@
#-----------------------------
# Magnetic Field
#-----------------------------
-process.load('Configuration.StandardSequences.MagneticField_cff')
+process.load('Configuration.StandardSequences.MagneticField_AutoFromDBCurrent_cff')
#-------------------------------------------------
# GEOMETRY
@@ -144,8 +130,7 @@
#--------------------------
# Quality Test
#--------------------------
-from DQMServices.Core.DQMQualityTester import DQMQualityTester
-process.stripQTester = DQMQualityTester(
+process.stripQTester = cms.EDAnalyzer("QualityTester",
qtList = cms.untracked.FileInPath('DQM/SiStripMonitorClient/data/sistrip_qualitytest_config.xml'),
prescaleFactor = cms.untracked.int32(3),
getQualityTestsFromFile = cms.untracked.bool(True),
@@ -153,7 +138,7 @@
qtestOnEndRun = cms.untracked.bool(True)
)
-process.trackingQTester = DQMQualityTester(
+process.trackingQTester = cms.EDAnalyzer("QualityTester",
qtList = cms.untracked.FileInPath('DQM/TrackingMonitorClient/data/tracking_qualitytest_config.xml'),
prescaleFactor = cms.untracked.int32(3),
getQualityTestsFromFile = cms.untracked.bool(True),
@@ -214,7 +199,7 @@
# Scheduling
#--------------------------
process.SiStripSources_LocalReco = cms.Sequence(process.siStripFEDMonitor*process.SiStripMonitorDigi*process.SiStripMonitorClusterReal)
-process.DQMCommon = cms.Sequence(process.stripQTester*process.trackingQTester*process.dqmEnv*process.dqmEnvTr*process.dqmSaver*process.dqmSaverPB)
+process.DQMCommon = cms.Sequence(process.stripQTester*process.trackingQTester*process.dqmEnv*process.dqmEnvTr*process.dqmSaver)
if (process.runType.getRunType() == process.runType.hi_run):
process.RecoForDQM_LocalReco = cms.Sequence(process.siPixelDigis*process.siStripDigis*process.trackerlocalreco)
else :
@@ -234,6 +219,7 @@
# event selection for cosmic data
if ((process.runType.getRunType() == process.runType.cosmic_run) and live): process.source.SelectEvents = cms.untracked.vstring('HLT*SingleMu*','HLT_L1*')
# Reference run for cosmic
+ process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/sistrip_reference_cosmic.root'
# Source config for cosmic data
process.SiStripSources_TrkReco_cosmic = cms.Sequence(process.SiStripMonitorTrack_ckf*process.TrackMon_ckf)
# Client config for cosmic data
@@ -272,8 +258,6 @@
process.trackingQTester.qtestOnEndRun = cms.untracked.bool(True)
process.p = cms.Path(process.scalersRawToDigi*
- process.tcdsDigis*
- process.onlineMetaDataDigis*
process.APVPhases*
process.consecutiveHEs*
process.hltTriggerTypeFilter*
@@ -304,6 +288,7 @@
'HLT_PAAK*'
)
+ process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/sistrip_reference_pp.root'
# Source and Client config for pp collisions
process.SiStripMonitorDigi.UseDCSFiltering = cms.bool(False)
@@ -377,8 +362,6 @@
process.p = cms.Path(
process.scalersRawToDigi*
- process.tcdsDigis*
- process.onlineMetaDataDigis*
process.APVPhases*
process.consecutiveHEs*
process.hltTriggerTypeFilter*
@@ -414,6 +397,7 @@
# process.DQMEventStreamerReader.SelectEvents = cms.untracked.PSet(SelectEvents = cms.vstring('HLT_600Tower*','HLT_L1*','HLT_Jet*','HLT_HT*','HLT_MinBias_*','HLT_Physics*', 'HLT_ZeroBias*'))
#
+ process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/sistrip_reference_pp.root'
process.SiStripMonitorDigi.UseDCSFiltering = cms.bool(False)
process.SiStripMonitorClusterReal.UseDCSFiltering = cms.bool(False)
@@ -473,8 +457,6 @@
process.RecoForDQM_TrkReco = cms.Sequence(process.offlineBeamSpot*process.MeasurementTrackerEvent*process.siPixelClusterShapeCache*process.recopixelvertexing*process.iterTracking_FirstStep)
process.p = cms.Path(process.scalersRawToDigi*
- process.tcdsDigis*
- process.onlineMetaDataDigis*
process.APVPhases*
process.consecutiveHEs*
process.hltTriggerTypeFilter*
@@ -535,6 +517,7 @@
'HLT_HIPhysics*'
)
+ process.DQMStore.referenceFileName = '/dqmdata/dqm/reference/sistrip_reference_pp.root'
process.SiStripMonitorDigi.UseDCSFiltering = cms.bool(False)
@@ -629,8 +612,6 @@
process.p = cms.Path(
process.scalersRawToDigi*
- process.tcdsDigis*
- process.onlineMetaDataDigis*
process.APVPhases*
process.consecutiveHEs*
process.hltTriggerTypeFilter*
diff --git a/DataFormats/SiPixelCluster/interface/SiPixelCluster.h b/DataFormats/SiPixelCluster/interface/SiPixelCluster.h
index ab4ae1add2132..22f9cb1020814 100644
--- a/DataFormats/SiPixelCluster/interface/SiPixelCluster.h
+++ b/DataFormats/SiPixelCluster/interface/SiPixelCluster.h
@@ -8,7 +8,7 @@
//! Class to contain and store all the topological information of pixel clusters:
//! charge, global size, size and the barycenter in x and y
//! local directions. It builds a vector of SiPixel (which is
-//! an inner class) and a container of channels.
+//! an inner class) and a container of channels.
//!
//! March 2007: Edge methods moved to RectangularPixelTopology class (V.Chiochia)
//! Feb 2008: Modify the Pixel class from float to shorts
@@ -26,153 +26,160 @@ class PixelDigi;
class SiPixelCluster {
public:
+
class Pixel {
public:
- constexpr Pixel() : x(0), y(0), adc(0) {} // for root
- constexpr Pixel(int pix_x, int pix_y, int pix_adc) : x(pix_x), y(pix_y), adc(pix_adc) {}
- uint16_t x;
+ constexpr Pixel() : x(0), y(0), adc(0){} // for root
+ constexpr Pixel(int pix_x, int pix_y, int pix_adc) :
+ x(pix_x), y(pix_y), adc(pix_adc) {}
+ uint16_t x;
uint16_t y;
uint16_t adc;
};
-
+
//--- Integer shift in x and y directions.
class Shift {
public:
- constexpr Shift(int dx, int dy) : dx_(dx), dy_(dy) {}
+ constexpr Shift( int dx, int dy) : dx_(dx), dy_(dy) {}
constexpr Shift() : dx_(0), dy_(0) {}
- constexpr int dx() const { return dx_; }
- constexpr int dy() const { return dy_; }
-
+ constexpr int dx() const { return dx_;}
+ constexpr int dy() const { return dy_;}
private:
int dx_;
int dy_;
};
-
+
//--- Position of a SiPixel
class PixelPos {
public:
constexpr PixelPos() : row_(0), col_(0) {}
- constexpr PixelPos(int row, int col) : row_(row), col_(col) {}
- constexpr int row() const { return row_; }
- constexpr int col() const { return col_; }
- constexpr PixelPos operator+(const Shift& shift) const { return PixelPos(row() + shift.dx(), col() + shift.dy()); }
-
+ constexpr PixelPos(int row, int col) : row_(row) , col_(col) {}
+ constexpr int row() const { return row_;}
+ constexpr int col() const { return col_;}
+ constexpr PixelPos operator+( const Shift& shift) const {
+ return PixelPos( row() + shift.dx(), col() + shift.dy());
+ }
private:
int row_;
int col_;
};
-
- typedef std::vector::const_iterator PixelDigiIter;
- typedef std::pair PixelDigiRange;
-
- static constexpr unsigned int MAXSPAN = 255;
- static constexpr unsigned int MAXPOS = 2047;
-
+
+ typedef std::vector::const_iterator PixelDigiIter;
+ typedef std::pair PixelDigiRange;
+
+
+ static constexpr unsigned int MAXSPAN=255;
+ static constexpr unsigned int MAXPOS=2047;
+
/** Construct from a range of digis that form a cluster and from
* a DetID. The range is assumed to be non-empty.
*/
-
+
SiPixelCluster() {}
-
- SiPixelCluster(unsigned int isize,
- uint16_t const* adcs,
- uint16_t const* xpos,
- uint16_t const* ypos,
- uint16_t const xmin,
- uint16_t const ymin)
- : thePixelOffset(2 * isize), thePixelADC(adcs, adcs + isize) {
+
+ SiPixelCluster(unsigned int isize, uint16_t const * adcs,
+ uint16_t const * xpos, uint16_t const * ypos,
+ uint16_t const xmin, uint16_t const ymin) :
+ thePixelOffset(2*isize), thePixelADC(adcs,adcs+isize) {
uint16_t maxCol = 0;
uint16_t maxRow = 0;
- for (unsigned int i = 0; i != isize; ++i) {
- uint16_t xoffset = xpos[i] - xmin;
- uint16_t yoffset = ypos[i] - ymin;
- thePixelOffset[i * 2] = std::min(uint16_t(MAXSPAN), xoffset);
- thePixelOffset[i * 2 + 1] = std::min(uint16_t(MAXSPAN), yoffset);
- if (xoffset > maxRow)
- maxRow = xoffset;
- if (yoffset > maxCol)
- maxCol = yoffset;
+ for (unsigned int i=0; i!=isize; ++i) {
+ uint16_t xoffset = xpos[i]-xmin;
+ uint16_t yoffset = ypos[i]-ymin;
+ thePixelOffset[i*2] = std::min(uint16_t(MAXSPAN),xoffset);
+ thePixelOffset[i*2+1] = std::min(uint16_t(MAXSPAN),yoffset);
+ if (xoffset > maxRow) maxRow = xoffset;
+ if (yoffset > maxCol) maxCol = yoffset;
}
- packRow(xmin, maxRow);
- packCol(ymin, maxCol);
+ packRow(xmin,maxRow);
+ packCol(ymin,maxCol);
}
-
+
+
// obsolete (only for regression tests)
- SiPixelCluster(const PixelPos& pix, int adc);
- void add(const PixelPos& pix, int adc);
-
- // Analog linear average position (barycenter)
+ SiPixelCluster( const PixelPos& pix, int adc);
+ void add( const PixelPos& pix, int adc);
+
+ // Analog linear average position (barycenter)
float x() const {
float qm = 0.0;
int isize = thePixelADC.size();
- for (int i = 0; i < isize; ++i)
- qm += float(thePixelADC[i]) * (thePixelOffset[i * 2] + minPixelRow() + 0.5f);
- return qm / charge();
+ for (int i=0; i& pixelOffset() const { return thePixelOffset; }
- const std::vector& pixelADC() const { return thePixelADC; }
-
+ } // Return total cluster charge.
+
+ inline int minPixelRow() const { return theMinPixelRow;} // The min x index.
+ inline int maxPixelRow() const { return minPixelRow() + rowSpan();} // The max x index.
+ inline int minPixelCol() const { return theMinPixelCol;} // The min y index.
+ inline int maxPixelCol() const { return minPixelCol() + colSpan();} // The max y index.
+
+
+ const std::vector & pixelOffset() const { return thePixelOffset;}
+ const std::vector & pixelADC() const { return thePixelADC;}
+
// obsolete, use single pixel access below
const std::vector pixels() const {
std::vector oldPixVector;
int isize = thePixelADC.size();
- oldPixVector.reserve(isize);
- for (int i = 0; i < isize; ++i) {
+ oldPixVector.reserve(isize);
+ for(int i=0; i thePixelOffset;
+
+ std::vector thePixelOffset;
std::vector thePixelADC;
-
- uint16_t theMinPixelRow = MAXPOS; // Minimum pixel index in the x direction (low edge).
- uint16_t theMinPixelCol = MAXPOS; // Minimum pixel index in the y direction (left edge).
- uint8_t thePixelRowSpan = 0; // Span pixel index in the x direction (low edge).
- uint8_t thePixelColSpan = 0; // Span pixel index in the y direction (left edge).
-
- float err_x = -99999.9f;
- float err_y = -99999.9f;
+
+
+ uint16_t theMinPixelRow=MAXPOS; // Minimum pixel index in the x direction (low edge).
+ uint16_t theMinPixelCol=MAXPOS; // Minimum pixel index in the y direction (left edge).
+ uint8_t thePixelRowSpan=0; // Span pixel index in the x direction (low edge).
+ uint8_t thePixelColSpan=0; // Span pixel index in the y direction (left edge).
+
+ float err_x=-99999.9f;
+ float err_y=-99999.9f;
+
};
+
// Comparison operators (needed by DetSetVector)
-inline bool operator<(const SiPixelCluster& one, const SiPixelCluster& other) {
- if (one.minPixelRow() < other.minPixelRow()) {
+inline bool operator<( const SiPixelCluster& one, const SiPixelCluster& other) {
+ if ( one.minPixelRow() < other.minPixelRow() ) {
return true;
- } else if (one.minPixelRow() > other.minPixelRow()) {
+ } else if ( one.minPixelRow() > other.minPixelRow() ) {
return false;
- } else if (one.minPixelCol() < other.minPixelCol()) {
+ } else if ( one.minPixelCol() < other.minPixelCol() ) {
return true;
} else {
return false;
}
}
+
#include "DataFormats/Common/interface/DetSetVector.h"
#include "DataFormats/Common/interface/DetSetVectorNew.h"
#include "DataFormats/Common/interface/Ref.h"
@@ -227,4 +240,4 @@ typedef edm::RefProd SiPixelClusterRefProd;
typedef edmNew::DetSetVector SiPixelClusterCollectionNew;
typedef edm::Ref SiPixelClusterRefNew;
-#endif
+#endif
diff --git a/DataFormats/SiPixelDigi/interface/SiPixelDigisSoA.h b/DataFormats/SiPixelDigi/interface/SiPixelDigisSoA.h
new file mode 100644
index 0000000000000..df249a3790cd2
--- /dev/null
+++ b/DataFormats/SiPixelDigi/interface/SiPixelDigisSoA.h
@@ -0,0 +1,32 @@
+#ifndef DataFormats_SiPixelDigi_interface_SiPixelDigisSoA_h
+#define DataFormats_SiPixelDigi_interface_SiPixelDigisSoA_h
+
+#include
+#include
+
+class SiPixelDigisSoA {
+public:
+ SiPixelDigisSoA() = default;
+ explicit SiPixelDigisSoA(size_t nDigis, const uint32_t *pdigi, const uint32_t *rawIdArr, const uint16_t *adc, const int32_t *clus);
+ ~SiPixelDigisSoA() = default;
+
+ auto size() const { return pdigi_.size(); }
+
+ uint32_t pdigi(size_t i) const { return pdigi_[i]; }
+ uint32_t rawIdArr(size_t i) const { return rawIdArr_[i]; }
+ uint16_t adc(size_t i) const { return adc_[i]; }
+ int32_t clus(size_t i) const { return clus_[i]; }
+
+ const std::vector& pdigiVector() const { return pdigi_; }
+ const std::vector& rawIdArrVector() const { return rawIdArr_; }
+ const std::vector& adcVector() const { return adc_; }
+ const std::vector& clusVector() const { return clus_; }
+
+private:
+ std::vector pdigi_;
+ std::vector rawIdArr_;
+ std::vector adc_;
+ std::vector clus_;
+};
+
+#endif
diff --git a/DataFormats/SiPixelDigi/src/SiPixelDigisSoA.cc b/DataFormats/SiPixelDigi/src/SiPixelDigisSoA.cc
new file mode 100644
index 0000000000000..ebc8ba2055f78
--- /dev/null
+++ b/DataFormats/SiPixelDigi/src/SiPixelDigisSoA.cc
@@ -0,0 +1,12 @@
+#include "DataFormats/SiPixelDigi/interface/SiPixelDigisSoA.h"
+
+#include
+
+SiPixelDigisSoA::SiPixelDigisSoA(size_t nDigis, const uint32_t *pdigi, const uint32_t *rawIdArr, const uint16_t *adc, const int32_t *clus):
+ pdigi_(pdigi, pdigi+nDigis),
+ rawIdArr_(rawIdArr, rawIdArr+nDigis),
+ adc_(adc, adc+nDigis),
+ clus_(clus, clus+nDigis)
+{
+ assert(pdigi_.size() == nDigis);
+}
diff --git a/DataFormats/SiPixelDigi/src/classes.h b/DataFormats/SiPixelDigi/src/classes.h
index 2f36b72ca7df8..256ca41ad1867 100644
--- a/DataFormats/SiPixelDigi/src/classes.h
+++ b/DataFormats/SiPixelDigi/src/classes.h
@@ -5,9 +5,13 @@
#include "DataFormats/SiPixelDigi/interface/PixelDigiCollection.h"
#include "DataFormats/SiPixelDigi/interface/SiPixelCalibDigi.h"
#include "DataFormats/SiPixelDigi/interface/SiPixelCalibDigiError.h"
+#include "DataFormats/SiPixelDigi/interface/SiPixelDigisSoA.h"
+#include "DataFormats/SiPixelDigi/interface/SiPixelDigiErrorsSoA.h"
#include "DataFormats/Common/interface/Wrapper.h"
#include "DataFormats/Common/interface/DetSetVector.h"
#include "DataFormats/Common/interface/DetSetVectorNew.h"
+#include "boost/cstdint.hpp"
#include
-#endif // SIPIXELDIGI_CLASSES_H
+
+#endif // SIPIXELDIGI_CLASSES_H
diff --git a/DataFormats/SiPixelDigi/src/classes_def.xml b/DataFormats/SiPixelDigi/src/classes_def.xml
index de7779a5c00ea..8cabbd3f3f06e 100755
--- a/DataFormats/SiPixelDigi/src/classes_def.xml
+++ b/DataFormats/SiPixelDigi/src/classes_def.xml
@@ -49,4 +49,10 @@
+
+
+
+
+
+
diff --git a/DataFormats/SiPixelRawData/src/classes.h b/DataFormats/SiPixelRawData/src/classes.h
index 73768cc373013..ab6b5d5f11363 100644
--- a/DataFormats/SiPixelRawData/src/classes.h
+++ b/DataFormats/SiPixelRawData/src/classes.h
@@ -6,4 +6,6 @@
#include "DataFormats/Common/interface/DetSetVector.h"
#include
-#endif // SIPIXELRAWDATA_CLASSES_H
+
+#endif // SIPIXELRAWDATA_CLASSES_H
+
diff --git a/EventFilter/SiPixelRawToDigi/plugins/BuildFile.xml b/EventFilter/SiPixelRawToDigi/plugins/BuildFile.xml
index f92aa68373927..4d2b5ebf45542 100644
--- a/EventFilter/SiPixelRawToDigi/plugins/BuildFile.xml
+++ b/EventFilter/SiPixelRawToDigi/plugins/BuildFile.xml
@@ -1,4 +1,7 @@
+
+
+
diff --git a/EventFilter/SiPixelRawToDigi/plugins/SiPixelDigiErrorsFromSoA.cc b/EventFilter/SiPixelRawToDigi/plugins/SiPixelDigiErrorsFromSoA.cc
new file mode 100644
index 0000000000000..9e998b92fc403
--- /dev/null
+++ b/EventFilter/SiPixelRawToDigi/plugins/SiPixelDigiErrorsFromSoA.cc
@@ -0,0 +1,183 @@
+#include "CondFormats/DataRecord/interface/SiPixelFedCablingMapRcd.h"
+#include "CondFormats/SiPixelObjects/interface/SiPixelFedCablingMap.h"
+#include "CondFormats/SiPixelObjects/interface/SiPixelFedCablingTree.h"
+#include "DataFormats/Common/interface/DetSetVector.h"
+#include "DataFormats/Common/interface/Handle.h"
+#include "DataFormats/DetId/interface/DetIdCollection.h"
+#include "DataFormats/SiPixelDetId/interface/PixelFEDChannel.h"
+#include "DataFormats/SiPixelDigi/interface/PixelDigi.h"
+#include "DataFormats/SiPixelDigi/interface/SiPixelDigiErrorsSoA.h"
+#include "EventFilter/SiPixelRawToDigi/interface/PixelDataFormatter.h"
+#include "FWCore/Framework/interface/ESTransientHandle.h"
+#include "FWCore/Framework/interface/ESWatcher.h"
+#include "FWCore/Framework/interface/EventSetup.h"
+#include "FWCore/Framework/interface/Event.h"
+#include "FWCore/Framework/interface/MakerMacros.h"
+#include "FWCore/Framework/interface/stream/EDProducer.h"
+#include "FWCore/ParameterSet/interface/ConfigurationDescriptions.h"
+#include "FWCore/ParameterSet/interface/ParameterSetDescription.h"
+#include "FWCore/ParameterSet/interface/ParameterSet.h"
+
+#include
+
+class SiPixelDigiErrorsFromSoA: public edm::stream::EDProducer<> {
+public:
+ explicit SiPixelDigiErrorsFromSoA(const edm::ParameterSet& iConfig);
+ ~SiPixelDigiErrorsFromSoA() override = default;
+
+ static void fillDescriptions(edm::ConfigurationDescriptions& descriptions);
+
+private:
+ void produce(edm::Event& iEvent, const edm::EventSetup& iSetup) override;
+
+ edm::EDGetTokenT digiErrorSoAGetToken_;
+
+ edm::EDPutTokenT> errorPutToken_;
+ edm::EDPutTokenT tkErrorPutToken_;
+ edm::EDPutTokenT userErrorPutToken_;
+ edm::EDPutTokenT> disabledChannelPutToken_;
+
+ edm::ESWatcher cablingWatcher_;
+ std::unique_ptr cabling_;
+ const std::string cablingMapLabel_;
+
+ const std::vector tkerrorlist_;
+ const std::vector usererrorlist_;
+
+ const bool usePhase1_;
+};
+
+SiPixelDigiErrorsFromSoA::SiPixelDigiErrorsFromSoA(const edm::ParameterSet& iConfig):
+ digiErrorSoAGetToken_{consumes(iConfig.getParameter("digiErrorSoASrc"))},
+ errorPutToken_{produces>()},
+ tkErrorPutToken_{produces()},
+ userErrorPutToken_{produces("UserErrorModules")},
+ disabledChannelPutToken_{produces>()},
+ cablingMapLabel_(iConfig.getParameter("CablingMapLabel")),
+ tkerrorlist_(iConfig.getParameter>("ErrorList")),
+ usererrorlist_(iConfig.getParameter>("UserErrorList")),
+ usePhase1_(iConfig.getParameter ("UsePhase1"))
+{}
+
+void SiPixelDigiErrorsFromSoA::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
+ edm::ParameterSetDescription desc;
+ desc.add("digiErrorSoASrc", edm::InputTag("siPixelDigiErrorsSoA"));
+ desc.add("CablingMapLabel","")->setComment("CablingMap label");
+ desc.add("UsePhase1",false)->setComment("## Use phase1");
+ desc.add >("ErrorList", std::vector{29})->setComment("## ErrorList: list of error codes used by tracking to invalidate modules");
+ desc.add >("UserErrorList", std::vector{40})->setComment("## UserErrorList: list of error codes used by Pixel experts for investigation");
+ descriptions.addWithDefaultLabel(desc);
+}
+
+void SiPixelDigiErrorsFromSoA::produce(edm::Event& iEvent, const edm::EventSetup& iSetup) {
+ // pack errors into collection
+
+ // initialize cabling map or update if necessary
+ if (cablingWatcher_.check(iSetup)) {
+ // cabling map, which maps online address (fed->link->ROC->local pixel) to offline (DetId->global pixel)
+ edm::ESTransientHandle cablingMap;
+ iSetup.get().get(cablingMapLabel_, cablingMap);
+ cabling_ = cablingMap->cablingTree();
+ LogDebug("map version:")<< cabling_->version();
+ }
+
+ const auto& digiErrors = iEvent.get(digiErrorSoAGetToken_);
+
+
+ edm::DetSetVector errorcollection{};
+ DetIdCollection tkerror_detidcollection{};
+ DetIdCollection usererror_detidcollection{};
+ edmNew::DetSetVector disabled_channelcollection{};
+
+ PixelDataFormatter formatter(cabling_.get(), usePhase1_); // for phase 1 & 0
+ const PixelDataFormatter::Errors *formatterErrors = digiErrors.formatterErrors();
+ assert(formatterErrors != nullptr);
+ auto errors = *formatterErrors; // make a copy
+ PixelDataFormatter::DetErrors nodeterrors;
+
+ auto size = digiErrors.size();
+ for (auto i = 0U; i < size; i++) {
+ PixelErrorCompact err = digiErrors.error(i);
+ if (err.errorType != 0) {
+ SiPixelRawDataError error(err.word, err.errorType, err.fedId + 1200);
+ errors[err.rawId].push_back(error);
+ }
+ }
+
+ constexpr uint32_t dummydetid = 0xffffffff;
+ typedef PixelDataFormatter::Errors::iterator IE;
+ for (IE is = errors.begin(); is != errors.end(); is++) {
+
+ uint32_t errordetid = is->first;
+ if (errordetid == dummydetid) {// errors given dummy detId must be sorted by Fed
+ nodeterrors.insert( nodeterrors.end(), errors[errordetid].begin(), errors[errordetid].end() );
+ }
+ else {
+ edm::DetSet& errorDetSet = errorcollection.find_or_insert(errordetid);
+ errorDetSet.data.insert(errorDetSet.data.end(), is->second.begin(), is->second.end());
+ // Fill detid of the detectors where there is error AND the error number is listed
+ // in the configurable error list in the job option cfi.
+ // Code needs to be here, because there can be a set of errors for each
+ // entry in the for loop over PixelDataFormatter::Errors
+
+ std::vector disabledChannelsDetSet;
+
+ for (auto const& aPixelError : errorDetSet) {
+ // For the time being, we extend the error handling functionality with ErrorType 25
+ // In the future, we should sort out how the usage of tkerrorlist can be generalized
+ if (aPixelError.getType() == 25) {
+ int fedId = aPixelError.getFedId();
+ const sipixelobjects::PixelFEDCabling* fed = cabling_->fed(fedId);
+ if (fed) {
+ cms_uint32_t linkId = formatter.linkId(aPixelError.getWord32());
+ const sipixelobjects::PixelFEDLink* link = fed->link(linkId);
+ if (link) {
+ // The "offline" 0..15 numbering is fixed by definition, also, the FrameConversion depends on it
+ // in contrast, the ROC-in-channel numbering is determined by hardware --> better to use the "offline" scheme
+ PixelFEDChannel ch = {fed->id(), linkId, 25, 0};
+ for (unsigned int iRoc = 1; iRoc <= link->numberOfROCs(); iRoc++) {
+ const sipixelobjects::PixelROC * roc = link->roc(iRoc);
+ if (roc->idInDetUnit() < ch.roc_first) ch.roc_first = roc->idInDetUnit();
+ if (roc->idInDetUnit() > ch.roc_last) ch.roc_last = roc->idInDetUnit();
+ }
+ if (ch.roc_first& errorDetSet = errorcollection.find_or_insert(dummydetid);
+ errorDetSet.data = nodeterrors;
+
+ iEvent.emplace(errorPutToken_, std::move(errorcollection));
+ iEvent.emplace(tkErrorPutToken_, std::move(tkerror_detidcollection));
+ iEvent.emplace(userErrorPutToken_, std::move(usererror_detidcollection));
+ iEvent.emplace(disabledChannelPutToken_, std::move(disabled_channelcollection));
+}
+
+DEFINE_FWK_MODULE(SiPixelDigiErrorsFromSoA);
diff --git a/EventFilter/SiPixelRawToDigi/plugins/SiPixelDigiErrorsSoAFromCUDA.cc b/EventFilter/SiPixelRawToDigi/plugins/SiPixelDigiErrorsSoAFromCUDA.cc
new file mode 100644
index 0000000000000..d47542528ed86
--- /dev/null
+++ b/EventFilter/SiPixelRawToDigi/plugins/SiPixelDigiErrorsSoAFromCUDA.cc
@@ -0,0 +1,75 @@
+#include "CUDADataFormats/Common/interface/CUDAProduct.h"
+#include "CUDADataFormats/SiPixelDigi/interface/SiPixelDigiErrorsCUDA.h"
+#include "DataFormats/SiPixelDigi/interface/SiPixelDigiErrorsSoA.h"
+#include "FWCore/Framework/interface/EventSetup.h"
+#include "FWCore/Framework/interface/Event.h"
+#include "FWCore/Framework/interface/MakerMacros.h"
+#include "FWCore/Framework/interface/stream/EDProducer.h"
+#include "FWCore/ParameterSet/interface/ConfigurationDescriptions.h"
+#include "FWCore/ParameterSet/interface/ParameterSetDescription.h"
+#include "FWCore/ParameterSet/interface/ParameterSet.h"
+#include "HeterogeneousCore/CUDACore/interface/CUDAScopedContext.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/host_unique_ptr.h"
+
+class SiPixelDigiErrorsSoAFromCUDA: public edm::stream::EDProducer {
+public:
+ explicit SiPixelDigiErrorsSoAFromCUDA(const edm::ParameterSet& iConfig);
+ ~SiPixelDigiErrorsSoAFromCUDA() override = default;
+
+ static void fillDescriptions(edm::ConfigurationDescriptions& descriptions);
+
+private:
+ void acquire(const edm::Event& iEvent, const edm::EventSetup& iSetup, edm::WaitingTaskWithArenaHolder waitingTaskHolder) override;
+ void produce(edm::Event& iEvent, const edm::EventSetup& iSetup) override;
+
+ edm::EDGetTokenT> digiErrorGetToken_;
+ edm::EDPutTokenT digiErrorPutToken_;
+
+ cudautils::host::unique_ptr data_;
+ GPU::SimpleVector error_;
+ const PixelFormatterErrors *formatterErrors_ = nullptr;
+};
+
+SiPixelDigiErrorsSoAFromCUDA::SiPixelDigiErrorsSoAFromCUDA(const edm::ParameterSet& iConfig):
+ digiErrorGetToken_(consumes>(iConfig.getParameter("src"))),
+ digiErrorPutToken_(produces())
+{}
+
+void SiPixelDigiErrorsSoAFromCUDA::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
+ edm::ParameterSetDescription desc;
+ desc.add("src", edm::InputTag("siPixelClustersCUDA"));
+ descriptions.addWithDefaultLabel(desc);
+}
+
+void SiPixelDigiErrorsSoAFromCUDA::acquire(const edm::Event& iEvent, const edm::EventSetup& iSetup, edm::WaitingTaskWithArenaHolder waitingTaskHolder) {
+ // Do the transfer in a CUDA stream parallel to the computation CUDA stream
+ CUDAScopedContext ctx{iEvent.streamID(), std::move(waitingTaskHolder)};
+
+ const auto& gpuDigiErrors = ctx.get(iEvent, digiErrorGetToken_);
+
+ auto tmp = gpuDigiErrors.dataErrorToHostAsync(ctx.stream());
+ error_ = std::move(tmp.first);
+ data_ = std::move(tmp.second);
+ formatterErrors_ = &(gpuDigiErrors.formatterErrors());
+}
+
+void SiPixelDigiErrorsSoAFromCUDA::produce(edm::Event& iEvent, const edm::EventSetup& iSetup) {
+ // The following line copies the data from the pinned host memory to
+ // regular host memory. In principle that feels unnecessary (why not
+ // just use the pinned host memory?). There are a few arguments for
+ // doing it though
+ // - Now can release the pinned host memory back to the (caching) allocator
+ // * if we'd like to keep the pinned memory, we'd need to also
+ // keep the CUDA stream around as long as that, or allow pinned
+ // host memory to be allocated without a CUDA stream
+ // - What if a CPU algorithm would produce the same SoA? We can't
+ // use cudaMallocHost without a GPU...
+ iEvent.emplace(digiErrorPutToken_, error_.size(), error_.data(), formatterErrors_);
+
+ error_ = GPU::make_SimpleVector(0, nullptr);
+ data_.reset();
+ formatterErrors_ = nullptr;
+}
+
+// define as framework plugin
+DEFINE_FWK_MODULE(SiPixelDigiErrorsSoAFromCUDA);
diff --git a/EventFilter/SiPixelRawToDigi/plugins/SiPixelDigisSoAFromCUDA.cc b/EventFilter/SiPixelRawToDigi/plugins/SiPixelDigisSoAFromCUDA.cc
new file mode 100644
index 0000000000000..068701f0bcf07
--- /dev/null
+++ b/EventFilter/SiPixelRawToDigi/plugins/SiPixelDigisSoAFromCUDA.cc
@@ -0,0 +1,81 @@
+#include "CUDADataFormats/Common/interface/CUDAProduct.h"
+#include "CUDADataFormats/SiPixelDigi/interface/SiPixelDigisCUDA.h"
+#include "DataFormats/SiPixelDigi/interface/SiPixelDigisSoA.h"
+#include "FWCore/Framework/interface/EventSetup.h"
+#include "FWCore/Framework/interface/Event.h"
+#include "FWCore/Framework/interface/MakerMacros.h"
+#include "FWCore/Framework/interface/stream/EDProducer.h"
+#include "FWCore/ParameterSet/interface/ConfigurationDescriptions.h"
+#include "FWCore/ParameterSet/interface/ParameterSetDescription.h"
+#include "FWCore/ParameterSet/interface/ParameterSet.h"
+#include "HeterogeneousCore/CUDACore/interface/CUDAScopedContext.h"
+#include "HeterogeneousCore/CUDAUtilities/interface/host_unique_ptr.h"
+
+
+class SiPixelDigisSoAFromCUDA: public edm::stream::EDProducer {
+public:
+ explicit SiPixelDigisSoAFromCUDA(const edm::ParameterSet& iConfig);
+ ~SiPixelDigisSoAFromCUDA() override = default;
+
+ static void fillDescriptions(edm::ConfigurationDescriptions& descriptions);
+
+private:
+ void acquire(const edm::Event& iEvent, const edm::EventSetup& iSetup, edm::WaitingTaskWithArenaHolder waitingTaskHolder) override;
+ void produce(edm::Event& iEvent, const edm::EventSetup& iSetup) override;
+
+ edm::EDGetTokenT> digiGetToken_;
+ edm::EDPutTokenT digiPutToken_;
+
+ cudautils::host::unique_ptr pdigi_;
+ cudautils::host::unique_ptr rawIdArr_;
+ cudautils::host::unique_ptr adc_;
+ cudautils::host::unique_ptr< int32_t[]> clus_;
+
+ int nDigis_;
+};
+
+SiPixelDigisSoAFromCUDA::SiPixelDigisSoAFromCUDA(const edm::ParameterSet& iConfig):
+ digiGetToken_(consumes>(iConfig.getParameter("src"))),
+ digiPutToken_(produces())
+{}
+
+void SiPixelDigisSoAFromCUDA::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
+ edm::ParameterSetDescription desc;
+ desc.add("src", edm::InputTag("siPixelClustersCUDA"));
+ descriptions.addWithDefaultLabel(desc);
+}
+
+void SiPixelDigisSoAFromCUDA::acquire(const edm::Event& iEvent, const edm::EventSetup& iSetup, edm::WaitingTaskWithArenaHolder waitingTaskHolder) {
+ // Do the transfer in a CUDA stream parallel to the computation CUDA stream
+ CUDAScopedContext ctx{iEvent.streamID(), std::move(waitingTaskHolder)};
+
+ const auto& gpuDigis = ctx.get(iEvent, digiGetToken_);
+
+ nDigis_ = gpuDigis.nDigis();
+ pdigi_ = gpuDigis.pdigiToHostAsync(ctx.stream());
+ rawIdArr_ = gpuDigis.rawIdArrToHostAsync(ctx.stream());
+ adc_ = gpuDigis.adcToHostAsync(ctx.stream());
+ clus_ = gpuDigis.clusToHostAsync(ctx.stream());
+}
+
+void SiPixelDigisSoAFromCUDA::produce(edm::Event& iEvent, const edm::EventSetup& iSetup) {
+ // The following line copies the data from the pinned host memory to
+ // regular host memory. In principle that feels unnecessary (why not
+ // just use the pinned host memory?). There are a few arguments for
+ // doing it though
+ // - Now can release the pinned host memory back to the (caching) allocator
+ // * if we'd like to keep the pinned memory, we'd need to also
+ // keep the CUDA stream around as long as that, or allow pinned
+ // host memory to be allocated without a CUDA stream
+ // - What if a CPU algorithm would produce the same SoA? We can't
+ // use cudaMallocHost without a GPU...
+ iEvent.emplace(digiPutToken_, nDigis_, pdigi_.get(), rawIdArr_.get(), adc_.get(), clus_.get());
+
+ pdigi_.reset();
+ rawIdArr_.reset();
+ adc_.reset();
+ clus_.reset();
+}
+
+// define as framework plugin
+DEFINE_FWK_MODULE(SiPixelDigisSoAFromCUDA);
diff --git a/EventFilter/SiPixelRawToDigi/python/SiPixelRawToDigi_cfi.py b/EventFilter/SiPixelRawToDigi/python/SiPixelRawToDigi_cfi.py
index 12ff657cefd8e..50c8f0fcabd3c 100644
--- a/EventFilter/SiPixelRawToDigi/python/SiPixelRawToDigi_cfi.py
+++ b/EventFilter/SiPixelRawToDigi/python/SiPixelRawToDigi_cfi.py
@@ -1,7 +1,24 @@
import FWCore.ParameterSet.Config as cms
-import EventFilter.SiPixelRawToDigi.siPixelRawToDigi_cfi
+from EventFilter.SiPixelRawToDigi.siPixelRawToDigi_cfi import siPixelRawToDigi as _siPixelRawToDigi
-siPixelDigis = EventFilter.SiPixelRawToDigi.siPixelRawToDigi_cfi.siPixelRawToDigi.clone()
+from HeterogeneousCore.CUDACore.SwitchProducerCUDA import SwitchProducerCUDA
+siPixelDigis = SwitchProducerCUDA(
+ cpu = _siPixelRawToDigi.clone()
+)
from Configuration.Eras.Modifier_phase1Pixel_cff import phase1Pixel
-phase1Pixel.toModify(siPixelDigis, UsePhase1=True)
+phase1Pixel.toModify(siPixelDigis.cpu, UsePhase1=True)
+
+from Configuration.ProcessModifiers.gpu_cff import gpu
+gpu.toModify(siPixelDigis,
+ cuda = cms.EDAlias(
+ siPixelDigiErrors = cms.VPSet(
+ cms.PSet(type = cms.string("DetIdedmEDCollection")),
+ cms.PSet(type = cms.string("SiPixelRawDataErroredmDetSetVector")),
+ cms.PSet(type = cms.string("PixelFEDChanneledmNewDetSetVector"))
+ ),
+ siPixelDigisClustersPreSplitting = cms.VPSet(
+ cms.PSet(type = cms.string("PixelDigiedmDetSetVector"))
+ )
+ )
+)
diff --git a/EventFilter/SiPixelRawToDigi/python/siPixelDigis_cff.py b/EventFilter/SiPixelRawToDigi/python/siPixelDigis_cff.py
new file mode 100644
index 0000000000000..31ba8596bddc6
--- /dev/null
+++ b/EventFilter/SiPixelRawToDigi/python/siPixelDigis_cff.py
@@ -0,0 +1,30 @@
+import FWCore.ParameterSet.Config as cms
+
+from EventFilter.SiPixelRawToDigi.SiPixelRawToDigi_cfi import siPixelDigis
+from EventFilter.SiPixelRawToDigi.siPixelDigisSoAFromCUDA_cfi import siPixelDigisSoAFromCUDA as _siPixelDigisSoAFromCUDA
+from EventFilter.SiPixelRawToDigi.siPixelDigiErrorsSoAFromCUDA_cfi import siPixelDigiErrorsSoAFromCUDA as _siPixelDigiErrorsSoAFromCUDA
+from EventFilter.SiPixelRawToDigi.siPixelDigiErrorsFromSoA_cfi import siPixelDigiErrorsFromSoA as _siPixelDigiErrorsFromSoA
+
+siPixelDigisTask = cms.Task(siPixelDigis)
+
+siPixelDigisSoA = _siPixelDigisSoAFromCUDA.clone(
+ src = "siPixelClustersCUDAPreSplitting"
+)
+siPixelDigiErrorsSoA = _siPixelDigiErrorsSoAFromCUDA.clone(
+ src = "siPixelClustersCUDAPreSplitting"
+)
+siPixelDigiErrors = _siPixelDigiErrorsFromSoA.clone()
+
+from Configuration.Eras.Modifier_phase1Pixel_cff import phase1Pixel
+phase1Pixel.toModify(siPixelDigiErrors, UsePhase1=True)
+
+siPixelDigisTaskCUDA = cms.Task(
+ siPixelDigisSoA,
+ siPixelDigiErrorsSoA,
+ siPixelDigiErrors
+)
+
+from Configuration.ProcessModifiers.gpu_cff import gpu
+_siPixelDigisTask_gpu = siPixelDigisTask.copy()
+_siPixelDigisTask_gpu.add(siPixelDigisTaskCUDA)
+gpu.toReplaceWith(siPixelDigisTask, _siPixelDigisTask_gpu)
diff --git a/Geometry/TrackerGeometryBuilder/interface/phase1PixelTopology.h b/Geometry/TrackerGeometryBuilder/interface/phase1PixelTopology.h
index cefdbe4b3296a..05e6b01e96c24 100644
--- a/Geometry/TrackerGeometryBuilder/interface/phase1PixelTopology.h
+++ b/Geometry/TrackerGeometryBuilder/interface/phase1PixelTopology.h
@@ -2,68 +2,150 @@
#define Geometry_TrackerGeometryBuilder_phase1PixelTopology_h
#include
+#include
namespace phase1PixelTopology {
- constexpr uint16_t numRowsInRoc = 80;
- constexpr uint16_t numColsInRoc = 52;
- constexpr uint16_t lastRowInRoc = numRowsInRoc - 1;
- constexpr uint16_t lastColInRoc = numColsInRoc - 1;
+ constexpr uint16_t numRowsInRoc = 80;
+ constexpr uint16_t numColsInRoc = 52;
+ constexpr uint16_t lastRowInRoc = numRowsInRoc - 1;
+ constexpr uint16_t lastColInRoc = numColsInRoc - 1;
- constexpr uint16_t numRowsInModule = 2 * numRowsInRoc;
- constexpr uint16_t numColsInModule = 8 * numColsInRoc;
- constexpr uint16_t lastRowInModule = numRowsInModule - 1;
- constexpr uint16_t lastColInModule = numColsInModule - 1;
+ constexpr uint16_t numRowsInModule = 2 * numRowsInRoc;
+ constexpr uint16_t numColsInModule = 8 * numColsInRoc;
+ constexpr uint16_t lastRowInModule = numRowsInModule - 1;
+ constexpr uint16_t lastColInModule = numColsInModule - 1;
constexpr int16_t xOffset = -81;
- constexpr int16_t yOffset = -54 * 4;
+ constexpr int16_t yOffset = -54*4;
- constexpr uint32_t numPixsInModule = uint32_t(numRowsInModule) * uint32_t(numColsInModule);
+ constexpr uint32_t numPixsInModule = uint32_t(numRowsInModule)* uint32_t(numColsInModule);
+
+ constexpr uint32_t numberOfModules = 1856;
+ constexpr uint32_t numberOfLayers = 10;
+ constexpr uint32_t layerStart[numberOfLayers + 1] = {
+ 0, 96, 320, 672, // barrel
+ 1184, 1296, 1408, // positive endcap
+ 1520, 1632, 1744, // negative endcap
+ numberOfModules
+ };
+ constexpr char const * layerName[numberOfLayers] = {
+ "BL1", "BL2", "BL3", "BL4", // barrel
+ "E+1", "E+2", "E+3", // positive endcap
+ "E-1", "E-2", "E-3" // negative endcap
+ };
+
+
+ template
+ constexpr auto map_to_array_helper(Function f, std::index_sequence)
+ -> std::array::type, sizeof...(Indices)>
+ {
+ return {{ f(Indices)... }};
+ }
+
+ template
+ constexpr auto map_to_array(Function f)
+ -> std::array::type, N>
+ {
+ return map_to_array_helper(f, std::make_index_sequence{});
+ }
+
+
+ constexpr uint32_t findMaxModuleStride() {
+ bool go = true;
+ int n=2;
+ while (go) {
+ for (uint8_t i=1; i<11; ++i) {
+ if (layerStart[i]%n !=0) {go=false; break;}
+ }
+ if(!go) break;
+ n*=2;
+ }
+ return n/2;
+ }
+
+ constexpr uint32_t maxModuleStride = findMaxModuleStride();
+
+ constexpr uint8_t findLayer(uint32_t detId) {
+ for (uint8_t i=0; i<11; ++i) if (detId layer = map_to_array(findLayerFromCompact);
+
+ constexpr bool validateLayerIndex() {
+ bool res=true;
+ for (auto i=0U; i=layerStart[layer[j]]);
+ res &=(i> 2;
- uint16_t q = (n >> 1) + (n >> 4);
- q = q + (q >> 4) + (q >> 5);
- q = q >> 3;
- uint16_t r = n - q * 13;
+ constexpr inline
+ uint16_t divu52(uint16_t n) {
+ n = n>>2;
+ uint16_t q = (n>>1) + (n>>4);
+ q = q + (q>>4) + (q>>5); q = q >> 3;
+ uint16_t r = n - q*13;
return q + ((r + 3) >> 4);
}
- constexpr inline bool isEdgeX(uint16_t px) { return (px == 0) | (px == lastRowInModule); }
- constexpr inline bool isEdgeY(uint16_t py) { return (py == 0) | (py == lastColInModule); }
+ constexpr inline
+ bool isEdgeX(uint16_t px) { return (px==0) | (px==lastRowInModule); }
+
+ constexpr inline
+ bool isEdgeY(uint16_t py) { return (py==0) | (py==lastColInModule); }
- constexpr inline uint16_t toRocX(uint16_t px) { return (px < numRowsInRoc) ? px : px - numRowsInRoc; }
- constexpr inline uint16_t toRocY(uint16_t py) {
+ constexpr inline
+ uint16_t toRocX(uint16_t px) { return (px lastRowInRoc)
- shift += 1;
- if (px > numRowsInRoc)
- shift += 1;
- return px + shift;
+ if (px>lastRowInRoc) shift+=1;
+ if (px>numRowsInRoc) shift+=1;
+ return px+shift;
}
- constexpr inline uint16_t localY(uint16_t py) {
+ constexpr inline
+ uint16_t localY(uint16_t py) {
auto roc = divu52(py);
- auto shift = 2 * roc;
- auto yInRoc = py - 52 * roc;
- if (yInRoc > 0)
- shift += 1;
- return py + shift;
+ auto shift = 2*roc;
+ auto yInRoc = py - 52*roc;
+ if (yInRoc>0) shift+=1;
+ return py+shift;
}
-} // namespace phase1PixelTopology
+}
-#endif // Geometry_TrackerGeometryBuilder_phase1PixelTopology_h
+#endif // Geometry_TrackerGeometryBuilder_phase1PixelTopology_h
diff --git a/Geometry/TrackerGeometryBuilder/test/phase1PixelTopology_t.cpp b/Geometry/TrackerGeometryBuilder/test/phase1PixelTopology_t.cpp
index 9a00efbff9a9a..5c37dad30d73e 100644
--- a/Geometry/TrackerGeometryBuilder/test/phase1PixelTopology_t.cpp
+++ b/Geometry/TrackerGeometryBuilder/test/phase1PixelTopology_t.cpp
@@ -8,138 +8,149 @@ namespace {
// original code from CMSSW_4_4
- std::tuple localXori(int mpx) {
- const float m_pitchx = 1.f;
- int binoffx = int(mpx); // truncate to int
- float local_pitchx = m_pitchx; // defaultpitch
-
- if (binoffx > 80) { // ROC 1 - handles x on edge cluster
- binoffx = binoffx + 2;
- } else if (binoffx == 80) { // ROC 1
- binoffx = binoffx + 1;
+ std::tuple localXori(int mpx) {
+ const float m_pitchx=1.f;
+ int binoffx = int(mpx); // truncate to int
+ float local_pitchx = m_pitchx; // defaultpitch
+
+ if (binoffx>80) { // ROC 1 - handles x on edge cluster
+ binoffx=binoffx+2;
+ } else if (binoffx==80) { // ROC 1
+ binoffx=binoffx+1;
local_pitchx = 2 * m_pitchx;
- } else if (binoffx == 79) { // ROC 0
- binoffx = binoffx + 0;
+ } else if (binoffx==79) { // ROC 0
+ binoffx=binoffx+0;
local_pitchx = 2 * m_pitchx;
- } else if (binoffx >= 0) { // ROC 0
- binoffx = binoffx + 0;
+ } else if (binoffx>=0) { // ROC 0
+ binoffx=binoffx+0;
- } else { // too small
- assert("binoffx too small" == 0);
+ } else { // too small
+ assert("binoffx too small"==0);
}
- return std::make_tuple(binoffx, local_pitchx > m_pitchx);
+ return std::make_tuple(binoffx,local_pitchx>m_pitchx);
}
- std::tuple localYori(int mpy) {
- const float m_pitchy = 1.f;
- int binoffy = int(mpy); // truncate to int
- float local_pitchy = m_pitchy; // defaultpitch
+ std::tuple localYori(int mpy) {
+ const float m_pitchy=1.f;
+ int binoffy = int(mpy); // truncate to int
+ float local_pitchy = m_pitchy; // defaultpitch
- if (binoffy > 416) { // ROC 8, not real ROC
- binoffy = binoffy + 17;
- } else if (binoffy == 416) { // ROC 8
- binoffy = binoffy + 16;
+ if (binoffy>416) { // ROC 8, not real ROC
+ binoffy=binoffy+17;
+ } else if (binoffy==416) { // ROC 8
+ binoffy=binoffy+16;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy == 415) { // ROC 7, last big pixel
- binoffy = binoffy + 15;
+ } else if (binoffy==415) { // ROC 7, last big pixel
+ binoffy=binoffy+15;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy > 364) { // ROC 7
- binoffy = binoffy + 15;
- } else if (binoffy == 364) { // ROC 7
- binoffy = binoffy + 14;
+ } else if (binoffy>364) { // ROC 7
+ binoffy=binoffy+15;
+ } else if (binoffy==364) { // ROC 7
+ binoffy=binoffy+14;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy == 363) { // ROC 6
- binoffy = binoffy + 13;
+ } else if (binoffy==363) { // ROC 6
+ binoffy=binoffy+13;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy > 312) { // ROC 6
- binoffy = binoffy + 13;
- } else if (binoffy == 312) { // ROC 6
- binoffy = binoffy + 12;
+ } else if (binoffy>312) { // ROC 6
+ binoffy=binoffy+13;
+ } else if (binoffy==312) { // ROC 6
+ binoffy=binoffy+12;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy == 311) { // ROC 5
- binoffy = binoffy + 11;
+ } else if (binoffy==311) { // ROC 5
+ binoffy=binoffy+11;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy > 260) { // ROC 5
- binoffy = binoffy + 11;
- } else if (binoffy == 260) { // ROC 5
- binoffy = binoffy + 10;
+ } else if (binoffy>260) { // ROC 5
+ binoffy=binoffy+11;
+ } else if (binoffy==260) { // ROC 5
+ binoffy=binoffy+10;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy == 259) { // ROC 4
- binoffy = binoffy + 9;
+ } else if (binoffy==259) { // ROC 4
+ binoffy=binoffy+9;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy > 208) { // ROC 4
- binoffy = binoffy + 9;
- } else if (binoffy == 208) { // ROC 4
- binoffy = binoffy + 8;
+ } else if (binoffy>208) { // ROC 4
+ binoffy=binoffy+9;
+ } else if (binoffy==208) { // ROC 4
+ binoffy=binoffy+8;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy == 207) { // ROC 3
- binoffy = binoffy + 7;
+ } else if (binoffy==207) { // ROC 3
+ binoffy=binoffy+7;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy > 156) { // ROC 3
- binoffy = binoffy + 7;
- } else if (binoffy == 156) { // ROC 3
- binoffy = binoffy + 6;
+ } else if (binoffy>156) { // ROC 3
+ binoffy=binoffy+7;
+ } else if (binoffy==156) { // ROC 3
+ binoffy=binoffy+6;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy == 155) { // ROC 2
- binoffy = binoffy + 5;
+ } else if (binoffy==155) { // ROC 2
+ binoffy=binoffy+5;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy > 104) { // ROC 2
- binoffy = binoffy + 5;
- } else if (binoffy == 104) { // ROC 2
- binoffy = binoffy + 4;
+ } else if (binoffy>104) { // ROC 2
+ binoffy=binoffy+5;
+ } else if (binoffy==104) { // ROC 2
+ binoffy=binoffy+4;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy == 103) { // ROC 1
- binoffy = binoffy + 3;
+ } else if (binoffy==103) { // ROC 1
+ binoffy=binoffy+3;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy > 52) { // ROC 1
- binoffy = binoffy + 3;
- } else if (binoffy == 52) { // ROC 1
- binoffy = binoffy + 2;
+ } else if (binoffy>52) { // ROC 1
+ binoffy=binoffy+3;
+ } else if (binoffy==52) { // ROC 1
+ binoffy=binoffy+2;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy == 51) { // ROC 0
- binoffy = binoffy + 1;
+ } else if (binoffy==51) { // ROC 0
+ binoffy=binoffy+1;
local_pitchy = 2 * m_pitchy;
- } else if (binoffy > 0) { // ROC 0
- binoffy = binoffy + 1;
- } else if (binoffy == 0) { // ROC 0
- binoffy = binoffy + 0;
+ } else if (binoffy>0) { // ROC 0
+ binoffy=binoffy+1;
+ } else if (binoffy==0) { // ROC 0
+ binoffy=binoffy+0;
local_pitchy = 2 * m_pitchy;
} else {
- assert("binoffy too small" == 0);
+ assert("binoffy too small"==0);
}
- return std::make_tuple(binoffy, local_pitchy > m_pitchy);
+ return std::make_tuple(binoffy,local_pitchy>m_pitchy);
}
-} // namespace
+}
int main() {
- for (uint16_t ix = 0; ix < 80 * 2; ++ix) {
+
+ for (uint16_t ix=0; ix<80*2; ++ix) {
auto ori = localXori(ix);
auto xl = phase1PixelTopology::localX(ix);
auto bp = phase1PixelTopology::isBigPixX(ix);
- if (std::get<0>(ori) != xl)
- std::cout << "Error " << std::get<0>(ori) << "!=" << xl << std::endl;
- assert(std::get<1>(ori) == bp);
+ if (std::get<0>(ori)!=xl) std::cout << "Error " << std::get<0>(ori) << "!=" << xl << std::endl;
+ assert(std::get<1>(ori)==bp);
}
- for (uint16_t iy = 0; iy < 52 * 8; ++iy) {
+ for (uint16_t iy=0; iy<52*8; ++iy) {
auto ori = localYori(iy);
auto yl = phase1PixelTopology::localY(iy);
auto bp = phase1PixelTopology::isBigPixY(iy);
- if (std::get<0>(ori) != yl)
- std::cout << "Error " << std::get<0>(ori) << "!=" << yl << std::endl;
- assert(std::get<1>(ori) == bp);
+ if (std::get<0>(ori)!=yl) std::cout << "Error " << std::get<0>(ori) << "!=" << yl << std::endl;
+ assert(std::get<1>(ori)==bp);
+ }
+
+ for (auto i = 0U; i < phase1PixelTopology::numberOfLayers; ++i) {
+ std::cout << "layer " << i << ", \"" << phase1PixelTopology::layerName[i] << "\", [" << phase1PixelTopology::layerStart[i] << ", " << phase1PixelTopology::layerStart[i+1] << ")" << std::endl;
+ }
+
+ for (auto i = 0U; i < phase1PixelTopology::numberOfModules; ++i) {
+ int layer = phase1PixelTopology::layer[i / phase1PixelTopology::maxModuleStride];
+ //std::cout << "module " << i << ": " << "layer " << layer << ", \"" << phase1PixelTopology::layerName[layer] << "\", [" << phase1PixelTopology::layerStart[layer] << ", " << phase1PixelTopology::layerStart[layer+1] << ")" << std::endl;
+ assert(layer < 10);
+ assert(i >= phase1PixelTopology::layerStart[layer]);
+ assert(i < phase1PixelTopology::layerStart[layer+1]);
}
return 0;
diff --git a/RecoLocalTracker/Configuration/python/RecoLocalTracker_cff.py b/RecoLocalTracker/Configuration/python/RecoLocalTracker_cff.py
index 3cae176059b3b..a486a83d178f4 100644
--- a/RecoLocalTracker/Configuration/python/RecoLocalTracker_cff.py
+++ b/RecoLocalTracker/Configuration/python/RecoLocalTracker_cff.py
@@ -9,11 +9,11 @@
from RecoLocalTracker.SiStripRecHitConverter.StripCPEfromTrackAngle_cfi import *
from RecoLocalTracker.SiStripZeroSuppression.SiStripZeroSuppression_cfi import *
from RecoLocalTracker.SiStripClusterizer.SiStripClusterizer_cfi import *
-from RecoLocalTracker.SiPixelClusterizer.SiPixelClusterizerPreSplitting_cfi import *
+from RecoLocalTracker.SiPixelClusterizer.siPixelClustersPreSplitting_cff import *
from RecoLocalTracker.SiPixelRecHits.SiPixelRecHits_cfi import *
from RecoLocalTracker.SubCollectionProducers.clustersummaryproducer_cfi import *
-pixeltrackerlocalrecoTask = cms.Task(siPixelClustersPreSplitting,siPixelRecHitsPreSplitting)
+pixeltrackerlocalrecoTask = cms.Task(siPixelClustersPreSplittingTask,siPixelRecHitsPreSplitting)
striptrackerlocalrecoTask = cms.Task(siStripZeroSuppression,siStripClusters,siStripMatchedRecHits)
trackerlocalrecoTask = cms.Task(pixeltrackerlocalrecoTask,striptrackerlocalrecoTask,clusterSummaryProducer)
@@ -21,9 +21,12 @@
striptrackerlocalreco = cms.Sequence(striptrackerlocalrecoTask)
trackerlocalreco = cms.Sequence(trackerlocalrecoTask)
+from Configuration.ProcessModifiers.gpu_cff import gpu
+from RecoLocalTracker.SiPixelRecHits.siPixelRecHitHeterogeneous_cfi import siPixelRecHitHeterogeneous as _siPixelRecHitHeterogeneous
+gpu.toReplaceWith(siPixelRecHitsPreSplitting, _siPixelRecHitHeterogeneous)
+
from RecoLocalTracker.SiPhase2Clusterizer.phase2TrackerClusterizer_cfi import *
from RecoLocalTracker.Phase2TrackerRecHits.Phase2StripCPEGeometricESProducer_cfi import *
-from RecoLocalTracker.SiPhase2VectorHitBuilder.siPhase2RecHitMatcher_cfi import *
_pixeltrackerlocalrecoTask_phase2 = pixeltrackerlocalrecoTask.copy()
_pixeltrackerlocalrecoTask_phase2.add(siPhase2Clusters)
diff --git a/RecoLocalTracker/SiPixelClusterizer/BuildFile.xml b/RecoLocalTracker/SiPixelClusterizer/BuildFile.xml
new file mode 100644
index 0000000000000..74e76ab6ff3e2
--- /dev/null
+++ b/RecoLocalTracker/SiPixelClusterizer/BuildFile.xml
@@ -0,0 +1,9 @@
+
+
+
+
+
+
+
+
+
diff --git a/RecoLocalTracker/SiPixelClusterizer/plugins/BuildFile.xml b/RecoLocalTracker/SiPixelClusterizer/plugins/BuildFile.xml
index c7b16a6ef4ee2..40a489f763397 100644
--- a/RecoLocalTracker/SiPixelClusterizer/plugins/BuildFile.xml
+++ b/RecoLocalTracker/SiPixelClusterizer/plugins/BuildFile.xml
@@ -1,8 +1,20 @@
-
-
-
-
-
-
-
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
diff --git a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelClusterProducer.cc b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelClusterProducer.cc
index 15a6536ca644b..95d1e7475e33e 100644
--- a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelClusterProducer.cc
+++ b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelClusterProducer.cc
@@ -17,7 +17,7 @@
// Geometry
#include "Geometry/Records/interface/TrackerDigiGeometryRecord.h"
-#include "Geometry/CommonDetUnit/interface/PixelGeomDetUnit.h"
+#include "Geometry/TrackerGeometryBuilder/interface/PixelGeomDetUnit.h"
// Data Formats
#include "DataFormats/Common/interface/DetSetVector.h"
@@ -43,32 +43,36 @@
// MessageLogger
#include "FWCore/MessageLogger/interface/MessageLogger.h"
-//---------------------------------------------------------------------------
-//! Constructor: set the ParameterSet and defer all thinking to setupClusterizer().
-//---------------------------------------------------------------------------
-SiPixelClusterProducer::SiPixelClusterProducer(edm::ParameterSet const& conf)
- : tPutPixelClusters(produces()),
- clusterMode_(conf.getParameter("ClusterMode")),
- maxTotalClusters_(conf.getParameter("maxNumberOfClusters")) {
- if (clusterMode_ == "PixelThresholdReclusterizer")
- tPixelClusters = consumes(conf.getParameter("src"));
- else
- tPixelDigi = consumes>(conf.getParameter("src"));
-
- const auto& payloadType = conf.getParameter("payloadType");
- if (payloadType == "HLT")
- theSiPixelGainCalibration_ = std::make_unique(conf);
- else if (payloadType == "Offline")
- theSiPixelGainCalibration_ = std::make_unique(conf);
- else if (payloadType == "Full")
- theSiPixelGainCalibration_ = std::make_unique(conf);
-
- //--- Make the algorithm(s) according to what the user specified
- //--- in the ParameterSet.
- setupClusterizer(conf);
-}
-// Destructor
+ //---------------------------------------------------------------------------
+ //! Constructor: set the ParameterSet and defer all thinking to setupClusterizer().
+ //---------------------------------------------------------------------------
+ SiPixelClusterProducer::SiPixelClusterProducer(edm::ParameterSet const& conf)
+ :
+ tPutPixelClusters(produces()),
+ clusterMode_( conf.getParameter("ClusterMode") ),
+ maxTotalClusters_( conf.getParameter( "maxNumberOfClusters" ) )
+ {
+ if ( clusterMode_ == "PixelThresholdReclusterizer" )
+ tPixelClusters = consumes( conf.getParameter("src") );
+ else
+ tPixelDigi = consumes>( conf.getParameter("src") );
+
+ const auto& payloadType = conf.getParameter( "payloadType" );
+ if (payloadType == "HLT")
+ theSiPixelGainCalibration_ = std::make_unique(conf);
+ else if (payloadType == "Offline")
+ theSiPixelGainCalibration_ = std::make_unique(conf);
+ else if (payloadType == "Full")
+ theSiPixelGainCalibration_ = std::make_unique(conf);
+
+ //--- Make the algorithm(s) according to what the user specified
+ //--- in the ParameterSet.
+ setupClusterizer(conf);
+
+ }
+
+ // Destructor
SiPixelClusterProducer::~SiPixelClusterProducer() = default;
void SiPixelClusterProducer::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
@@ -77,129 +81,138 @@ void SiPixelClusterProducer::fillDescriptions(edm::ConfigurationDescriptions& de
desc.add("src", edm::InputTag("siPixelDigis"));
desc.add("ClusterMode", "PixelThresholdClusterizer");
desc.add("maxNumberOfClusters", -1)->setComment("-1 means no limit");
- desc.add("payloadType", "Offline")
- ->setComment("Options: HLT - column granularity, Offline - gain:col/ped:pix");
+ desc.add("payloadType", "Offline")->setComment("Options: HLT - column granularity, Offline - gain:col/ped:pix");
PixelThresholdClusterizer::fillPSetDescription(desc);
- SiPixelGainCalibrationServiceBase::fillPSetDescription(desc); // no-op, but in principle the structures are there...
+ SiPixelGainCalibrationServiceBase::fillPSetDescription(desc); // no-op, but in principle the structures are there...
descriptions.add("SiPixelClusterizerDefault", desc);
}
-//---------------------------------------------------------------------------
-//! The "Event" entrypoint: gets called by framework for every event
-//---------------------------------------------------------------------------
-void SiPixelClusterProducer::produce(edm::Event& e, const edm::EventSetup& es) {
- //Setup gain calibration service
- theSiPixelGainCalibration_->setESObjects(es);
-
- // Step A.1: get input data
- edm::Handle inputClusters;
- edm::Handle> inputDigi;
- if (clusterMode_ == "PixelThresholdReclusterizer")
- e.getByToken(tPixelClusters, inputClusters);
- else
- e.getByToken(tPixelDigi, inputDigi);
-
- // Step A.2: get event setup
- edm::ESHandle geom;
- es.get().get(geom);
-
- edm::ESHandle trackerTopologyHandle;
- es.get().get(trackerTopologyHandle);
- tTopo_ = trackerTopologyHandle.product();
-
- // Step B: create the final output collection
- auto output = std::make_unique();
- //FIXME: put a reserve() here
-
- // Step C: Iterate over DetIds and invoke the pixel clusterizer algorithm
- // on each DetUnit
- if (clusterMode_ == "PixelThresholdReclusterizer")
- run(*inputClusters, geom, *output);
- else
- run(*inputDigi, geom, *output);
-
- // Step D: write output to file
- output->shrink_to_fit();
- e.put(tPutPixelClusters, std::move(output));
-}
+
+ //---------------------------------------------------------------------------
+ //! The "Event" entrypoint: gets called by framework for every event
+ //---------------------------------------------------------------------------
+ void SiPixelClusterProducer::produce(edm::Event& e, const edm::EventSetup& es)
+ {
+
+ //Setup gain calibration service
+ theSiPixelGainCalibration_->setESObjects( es );
+
+ // Step A.1: get input data
+ edm::Handle< SiPixelClusterCollectionNew > inputClusters;
+ edm::Handle< edm::DetSetVector > inputDigi;
+ if ( clusterMode_ == "PixelThresholdReclusterizer" )
+ e.getByToken(tPixelClusters, inputClusters);
+ else
+ e.getByToken(tPixelDigi, inputDigi);
+
+ // Step A.2: get event setup
+ edm::ESHandle geom;
+ es.get().get( geom );
+
+ edm::ESHandle trackerTopologyHandle;
+ es.get().get(trackerTopologyHandle);
+ tTopo_ = trackerTopologyHandle.product();
+
+ // Step B: create the final output collection
+ auto output = std::make_unique< SiPixelClusterCollectionNew>();
+ //FIXME: put a reserve() here
+
+ // Step C: Iterate over DetIds and invoke the pixel clusterizer algorithm
+ // on each DetUnit
+ if ( clusterMode_ == "PixelThresholdReclusterizer" )
+ run(*inputClusters, geom, *output );
+ else
+ run(*inputDigi, geom, *output );
+
+ // Step D: write output to file
+ output->shrink_to_fit();
+ e.put(tPutPixelClusters, std::move(output));
-//---------------------------------------------------------------------------
-//! Set up the specific algorithm we are going to use.
-//! TO DO: in the future, we should allow for a different algorithm for
-//! each detector subset (e.g. barrel vs forward, per layer, etc).
-//---------------------------------------------------------------------------
-void SiPixelClusterProducer::setupClusterizer(const edm::ParameterSet& conf) {
- if (clusterMode_ == "PixelThresholdReclusterizer" || clusterMode_ == "PixelThresholdClusterizer") {
- clusterizer_ = std::make_unique(conf);
- clusterizer_->setSiPixelGainCalibrationService(theSiPixelGainCalibration_.get());
- } else {
- throw cms::Exception("Configuration") << "[SiPixelClusterProducer]:"
- << " choice " << clusterMode_ << " is invalid.\n"
- << "Possible choices:\n"
- << " PixelThresholdClusterizer";
}
-}
-//---------------------------------------------------------------------------
-//! Iterate over DetUnits, and invoke the PixelClusterizer on each.
-//---------------------------------------------------------------------------
-template
-void SiPixelClusterProducer::run(const T& input,
- const edm::ESHandle& geom,
- edmNew::DetSetVector& output) {
- int numberOfDetUnits = 0;
- int numberOfClusters = 0;
-
- // Iterate on detector units
- typename T::const_iterator DSViter = input.begin();
- for (; DSViter != input.end(); DSViter++) {
- ++numberOfDetUnits;
-
- // LogDebug takes very long time, get rid off.
- //LogDebug("SiStripClusterizer") << "[SiPixelClusterProducer::run] DetID" << DSViter->id;
-
- std::vector badChannels;
- DetId detIdObject(DSViter->detId());
-
- // Comment: At the moment the clusterizer depends on geometry
- // to access information as the pixel topology (number of columns
- // and rows in a detector module).
- // In the future the geometry service will be replaced with
- // a ES service.
- const GeomDetUnit* geoUnit = geom->idToDetUnit(detIdObject);
- const PixelGeomDetUnit* pixDet = dynamic_cast(geoUnit);
- if (!pixDet) {
- // Fatal error! TO DO: throw an exception!
- assert(0);
+ //---------------------------------------------------------------------------
+ //! Set up the specific algorithm we are going to use.
+ //! TO DO: in the future, we should allow for a different algorithm for
+ //! each detector subset (e.g. barrel vs forward, per layer, etc).
+ //---------------------------------------------------------------------------
+ void SiPixelClusterProducer::setupClusterizer(const edm::ParameterSet& conf) {
+
+ if ( clusterMode_ == "PixelThresholdReclusterizer" || clusterMode_ == "PixelThresholdClusterizer" ) {
+ clusterizer_ = std::make_unique(conf);
+ clusterizer_->setSiPixelGainCalibrationService(theSiPixelGainCalibration_.get());
+ }
+ else {
+ throw cms::Exception("Configuration") << "[SiPixelClusterProducer]:"
+ <<" choice " << clusterMode_ << " is invalid.\n"
+ << "Possible choices:\n"
+ << " PixelThresholdClusterizer";
}
- {
- // Produce clusters for this DetUnit and store them in
+ }
+
+
+ //---------------------------------------------------------------------------
+ //! Iterate over DetUnits, and invoke the PixelClusterizer on each.
+ //---------------------------------------------------------------------------
+ template
+ void SiPixelClusterProducer::run(const T & input,
+ const edm::ESHandle & geom,
+ edmNew::DetSetVector & output) {
+ int numberOfDetUnits = 0;
+ int numberOfClusters = 0;
+
+ // Iterate on detector units
+ typename T::const_iterator DSViter = input.begin();
+ for( ; DSViter != input.end(); DSViter++) {
+ ++numberOfDetUnits;
+
+ // LogDebug takes very long time, get rid off.
+ //LogDebug("SiStripClusterizer") << "[SiPixelClusterProducer::run] DetID" << DSViter->id;
+
+ std::vector badChannels;
+ DetId detIdObject(DSViter->detId());
+
+ // Comment: At the moment the clusterizer depends on geometry
+ // to access information as the pixel topology (number of columns
+ // and rows in a detector module).
+ // In the future the geometry service will be replaced with
+ // a ES service.
+ const GeomDetUnit * geoUnit = geom->idToDetUnit( detIdObject );
+ const PixelGeomDetUnit * pixDet = dynamic_cast(geoUnit);
+ if (! pixDet) {
+ // Fatal error! TO DO: throw an exception!
+ assert(0);
+ }
+ {
+ // Produce clusters for this DetUnit and store them in
// a DetSet
edmNew::DetSetVector::FastFiller spc(output, DSViter->detId());
clusterizer_->clusterizeDetUnit(*DSViter, pixDet, tTopo_, badChannels, spc);
- if (spc.empty()) {
+ if ( spc.empty() ) {
spc.abort();
} else {
- numberOfClusters += spc.size();
+ numberOfClusters += spc.size();
}
- } // spc is not deleted and detsetvector updated
- if ((maxTotalClusters_ >= 0) && (numberOfClusters > maxTotalClusters_)) {
- edm::LogError("TooManyClusters")
- << "Limit on the number of clusters exceeded. An empty cluster collection will be produced instead.\n";
- edmNew::DetSetVector empty;
- empty.swap(output);
- break;
- }
- } // end of DetUnit loop
+ } // spc is not deleted and detsetvector updated
+ if ((maxTotalClusters_ >= 0) && (numberOfClusters > maxTotalClusters_)) {
+ edm::LogError("TooManyClusters") << "Limit on the number of clusters exceeded. An empty cluster collection will be produced instead.\n";
+ edmNew::DetSetVector empty;
+ empty.swap(output);
+ break;
+ }
+ } // end of DetUnit loop
+
+ //LogDebug ("SiPixelClusterProducer") << " Executing "
+ // << clusterMode_ << " resulted in " << numberOfClusters
+ // << " SiPixelClusters in " << numberOfDetUnits << " DetUnits.";
+ }
+
+
- //LogDebug ("SiPixelClusterProducer") << " Executing "
- // << clusterMode_ << " resulted in " << numberOfClusters
- // << " SiPixelClusters in " << numberOfDetUnits << " DetUnits.";
-}
#include "FWCore/PluginManager/interface/ModuleDef.h"
#include "FWCore/Framework/interface/MakerMacros.h"
DEFINE_FWK_MODULE(SiPixelClusterProducer);
+
diff --git a/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelDigisClustersFromSoA.cc b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelDigisClustersFromSoA.cc
new file mode 100644
index 0000000000000..4c405a8c85afd
--- /dev/null
+++ b/RecoLocalTracker/SiPixelClusterizer/plugins/SiPixelDigisClustersFromSoA.cc
@@ -0,0 +1,158 @@
+#include "DataFormats/Common/interface/DetSetVector.h"
+#include "DataFormats/Common/interface/Handle.h"
+#include "DataFormats/DetId/interface/DetId.h"
+#include "DataFormats/SiPixelCluster/interface/SiPixelCluster.h"
+#include "DataFormats/SiPixelDigi/interface/PixelDigi.h"
+#include "DataFormats/SiPixelDigi/interface/SiPixelDigisSoA.h"
+#include "DataFormats/TrackerCommon/interface/TrackerTopology.h"
+#include "FWCore/Framework/interface/EventSetup.h"
+#include "FWCore/Framework/interface/Event.h"
+#include "FWCore/Framework/interface/MakerMacros.h"
+#include "FWCore/Framework/interface/global/EDProducer.h"
+#include "FWCore/ParameterSet/interface/ConfigurationDescriptions.h"
+#include "FWCore/ParameterSet/interface/ParameterSetDescription.h"
+#include "FWCore/ParameterSet/interface/ParameterSet.h"
+#include "Geometry/Records/interface/TrackerTopologyRcd.h"
+
+namespace {
+ struct AccretionCluster {
+ typedef unsigned short UShort;
+ static constexpr UShort MAXSIZE = 256;
+ UShort adc[MAXSIZE];
+ UShort x[MAXSIZE];
+ UShort y[MAXSIZE];
+ UShort xmin=16000;
+ UShort ymin=16000;
+ unsigned int isize=0;
+ int charge=0;
+
+ void clear() {
+ isize=0;
+ charge=0;
+ xmin=16000;
+ ymin=16000;
+ }
+
+ bool add(SiPixelCluster::PixelPos const & p, UShort const iadc) {
+ if (isize==MAXSIZE) return false;
+ xmin=std::min(xmin,(unsigned short)(p.row()));
+ ymin=std::min(ymin,(unsigned short)(p.col()));
+ adc[isize]=iadc;
+ x[isize]=p.row();
+ y[isize++]=p.col();
+ charge+=iadc;
+ return true;
+ }
+ };
+
+ constexpr uint32_t dummydetid = 0xffffffff;
+}
+
+class SiPixelDigisClustersFromSoA: public edm::global::EDProducer<> {
+public:
+ explicit SiPixelDigisClustersFromSoA(const edm::ParameterSet& iConfig);
+ ~SiPixelDigisClustersFromSoA() override = default;
+
+ static void fillDescriptions(edm::ConfigurationDescriptions& descriptions);
+
+private:
+ void produce(edm::StreamID, edm::Event& iEvent, const edm::EventSetup& iSetup) const override;
+
+ edm::EDGetTokenT digiGetToken_;
+
+ edm::EDPutTokenT> digiPutToken_;
+ edm::EDPutTokenT clusterPutToken_;
+
+};
+
+SiPixelDigisClustersFromSoA::SiPixelDigisClustersFromSoA(const edm::ParameterSet& iConfig):
+ digiGetToken_(consumes(iConfig.getParameter("src"))),
+ digiPutToken_(produces>()),
+ clusterPutToken_(produces())
+{}
+
+void SiPixelDigisClustersFromSoA::fillDescriptions(edm::ConfigurationDescriptions& descriptions) {
+ edm::ParameterSetDescription desc;
+ desc.add("src", edm::InputTag("siPixelDigisSoA"));
+ descriptions.addWithDefaultLabel(desc);
+}
+
+void SiPixelDigisClustersFromSoA::produce(edm::StreamID, edm::Event& iEvent, const edm::EventSetup& iSetup) const {
+ const auto& digis = iEvent.get(digiGetToken_);
+
+ edm::ESHandle trackerTopologyHandle;
+ iSetup.get().get(trackerTopologyHandle);
+ const auto& ttopo = *trackerTopologyHandle;
+
+ auto collection = std::make_unique>();
+ auto outputClusters = std::make_unique();
+
+ const uint32_t nDigis = digis.size();
+ edm::DetSet * detDigis=nullptr;
+ for (uint32_t i = 0; i < nDigis; i++) {
+ if (digis.pdigi(i)==0) continue;
+ detDigis = &collection->find_or_insert(digis.rawIdArr(i));
+ if ( (*detDigis).empty() ) (*detDigis).data.reserve(32); // avoid the first relocations
+ break;
+ }
+
+ int32_t nclus=-1;
+ std::vector aclusters(1024);
+ auto totCluseFilled=0;
+
+ auto fillClusters = [&](uint32_t detId){
+ if (nclus<0) return; // this in reality should never happen
+ edmNew::DetSetVector