From 21137905baa83066295755511015c86175b590c9 Mon Sep 17 00:00:00 2001 From: Viktor Khristenko Date: Wed, 30 May 2018 11:35:35 +0200 Subject: [PATCH 1/4] Make HcalDetId and HcalDigi dataformats constexpr Similar to DataFormats/DetId updates for HcalDetid and HcalDigi - move the implementation to the header files - mark methods as `constexpr` --- DataFormats/Common/interface/DataFrame.h | 22 +- DataFormats/HcalDetId/interface/HcalDetId.h | 259 ++++++++++++++++-- .../HcalDetId/interface/HcalElectronicsId.h | 78 ++++-- DataFormats/HcalDetId/src/HcalDetId.cc | 246 ----------------- .../HcalDetId/src/HcalElectronicsId.cc | 35 --- .../HcalDigi/interface/HBHEDataFrame.h | 78 ++++-- DataFormats/HcalDigi/interface/HFDataFrame.h | 79 ++++-- DataFormats/HcalDigi/interface/HODataFrame.h | 76 +++-- .../HcalDigi/interface/HcalQIESample.h | 65 ++++- .../HcalDigi/interface/QIE10DataFrame.h | 99 ++++--- .../HcalDigi/interface/QIE11DataFrame.h | 76 +++-- DataFormats/HcalDigi/src/HBHEDataFrame.cc | 62 ----- DataFormats/HcalDigi/src/HFDataFrame.cc | 58 ---- DataFormats/HcalDigi/src/HODataFrame.cc | 60 ---- DataFormats/HcalDigi/src/HcalQIESample.cc | 31 --- DataFormats/HcalDigi/src/QIE10DataFrame.cc | 50 ---- DataFormats/HcalDigi/src/QIE11DataFrame.cc | 32 --- 17 files changed, 630 insertions(+), 776 deletions(-) diff --git a/DataFormats/Common/interface/DataFrame.h b/DataFormats/Common/interface/DataFrame.h index 63c0fa1153f35..25383636cd4c6 100644 --- a/DataFormats/Common/interface/DataFrame.h +++ b/DataFormats/Common/interface/DataFrame.h @@ -22,9 +22,9 @@ namespace edm { typedef data_type const * const_iterator; - inline + constexpr inline DataFrame() : m_id(0), m_data(nullptr), m_size(0){} - inline + constexpr inline DataFrame(id_type i, data_type const * idata, size_type isize) : m_id(i), m_data(idata), m_size(isize) {} @@ -34,40 +34,40 @@ namespace edm { inline void set(DataFrameContainer const & icont, size_type i); - inline + constexpr inline data_type & operator[](size_type i) { return data()[i]; } - inline + constexpr inline data_type operator[](size_type i) const { return m_data[i]; } - inline + constexpr inline iterator begin() { return data();} - inline + constexpr inline iterator end() { return data()+m_size;} - inline + constexpr inline const_iterator begin() const { return m_data;} - inline + constexpr inline const_iterator end() const { return m_data+m_size;} - inline + constexpr inline id_type id() const { return m_id;} - inline + constexpr inline size_type size() const { return m_size; } private: //for testing friend class ::TestDataFrame; - data_type * data() { + constexpr data_type * data() { return const_cast(m_data); } diff --git a/DataFormats/HcalDetId/interface/HcalDetId.h b/DataFormats/HcalDetId/interface/HcalDetId.h index f8fc122fdabb9..d43ca602990fe 100644 --- a/DataFormats/HcalDetId/interface/HcalDetId.h +++ b/DataFormats/HcalDetId/interface/HcalDetId.h @@ -4,6 +4,7 @@ #include #include "DataFormats/DetId/interface/DetId.h" #include "DataFormats/HcalDetId/interface/HcalSubdetector.h" +#include "FWCore/Utilities/interface/Exception.h" /** \class HcalDetId @@ -31,63 +32,261 @@ class HcalDetId : public DetId { public: /** Create a null cellid*/ - HcalDetId(); + constexpr HcalDetId() : DetId() { + } /** Create cellid from raw id (0=invalid tower id) */ - HcalDetId(uint32_t rawid); + constexpr HcalDetId(uint32_t rawid) { + if ((DetId::Detector(rawid>>DetId::kDetOffset)&DetId::kDetMask) != Hcal) { + id_ = rawid; + } else { + HcalSubdetector subdet = (HcalSubdetector)((rawid>>DetId::kSubdetOffset)&DetId::kSubdetMask); + if ((subdet==HcalBarrel) || (subdet==HcalEndcap) || + (subdet==HcalOuter) || (subdet==HcalForward)) { + id_ = newForm(rawid); + } else { + id_ = rawid; + } + } + } + /** Constructor from subdetector, signed tower ieta,iphi,and depth */ - HcalDetId(HcalSubdetector subdet, int tower_ieta, int tower_iphi, int depth); + constexpr HcalDetId(HcalSubdetector subdet, int tower_ieta, int tower_iphi, int depth) : DetId(Hcal,subdet) { + // (no checking at this point!) + id_ |= (kHcalIdFormat2) | ((depth&kHcalDepthMask2)<0)?(kHcalZsideMask2|(tower_ieta<0)?(kHcalZsideMask1|(eta<0)?(kHcalZsideMask2|(eta<>kHcalEtaOffset1)&kHcalEtaMask1; + else return (id_>>kHcalEtaOffset2)&kHcalEtaMask2; + } /// get the cell ieta - int ieta() const { return zside()*ietaAbs(); } + constexpr int ieta() const { return zside()*ietaAbs(); } /// get the cell iphi - int iphi() const; + constexpr int iphi() const { + if (oldFormat()) return id_&kHcalPhiMask1; + else return id_&kHcalPhiMask2; + } /// get the tower depth - int depth() const; + constexpr int depth() const { + if (oldFormat()) return (id_>>kHcalDepthOffset1)&kHcalDepthMask1; + else return (id_>>kHcalDepthOffset2)&kHcalDepthMask2; + } /// get full depth information for HF - int hfdepth() const; + constexpr int hfdepth() const { + int dep = depth(); + if (subdet() == HcalForward) { + if (dep > 2) dep -= 2; + } + return dep; + } /// get the tower depth - uint32_t maskDepth() const; + constexpr uint32_t maskDepth() const { + if (oldFormat()) return (id_|kHcalDepthSet1); + else return (id_|kHcalDepthSet2); + } /// change format - uint32_t otherForm() const; - void changeForm(); - uint32_t newForm() const; - static uint32_t newForm(const uint32_t&); + constexpr uint32_t otherForm() const { + uint32_t rawid = (id_&kHcalIdMask); + if (oldFormat()) { + rawid = newForm(id_); + } else { + rawid |= ((depth()&kHcalDepthMask1)<0)?(kHcalZsideMask1|(ieta()<0)?(kHcalZsideMask2|(eta< 2) dep -= 2; + bool result = ((zsid==zside()) && (eta==ietaAbs()) && (phi==iphi()) && + (dep==hfdepth())); + return result; + } + constexpr HcalDetId baseDetId() const { + if (subdet() != HcalForward || depth() <= 2) { + return HcalDetId(id_); + } else { + int zsid{0}, eta{0}, phi{0}, dep{0}; + unpackId(id_, zsid, eta, phi, dep); + dep -= 2; + uint32_t rawid = id_&kHcalIdMask; + rawid |= (kHcalIdFormat2) | ((dep&kHcalDepthMask2)<0)?(kHcalZsideMask2|(eta< 2) { + return HcalDetId(id_); + } else { + int zsid{0}, eta{0}, phi{0}, dep{0}; + unpackId(id_, zsid, eta, phi, dep); + dep += 2; + uint32_t rawid = id_&kHcalIdMask; + rawid |= (kHcalIdFormat2) | ((dep&kHcalDepthMask2)<0)?(kHcalZsideMask2|(eta<360)?(simple_iphi-360):(simple_iphi)); + } /// get the largest crystal_iphi of the crystal in front of this tower (HB and HE tower 17 only) - int crystal_iphi_high() const; + constexpr int crystal_iphi_high() const { + int simple_iphi=((iphi()-1)*5)+5; + simple_iphi+=10; + return ((simple_iphi>360)?(simple_iphi-360):(simple_iphi)); + } static const HcalDetId Undefined; private: - void newFromOld(const uint32_t&); - static void unpackId(const uint32_t&, int&, int&, int&, int&); + constexpr void newFromOld(const uint32_t& rawid) { + id_ = newForm(rawid); + } + + constexpr static void unpackId(const uint32_t& rawid, int& zsid, int& eta, int& phi, + int& dep) { + if ((rawid&kHcalIdFormat2)==0) { + zsid = (rawid&kHcalZsideMask1)?(1):(-1); + eta = (rawid>>kHcalEtaOffset1)&kHcalEtaMask1; + phi = rawid&kHcalPhiMask1; + dep = (rawid>>kHcalDepthOffset1)&kHcalDepthMask1; + } else { + zsid = (rawid&kHcalZsideMask2)?(1):(-1); + eta = (rawid>>kHcalEtaOffset2)&kHcalEtaMask2; + phi = rawid&kHcalPhiMask2; + dep = (rawid>>kHcalDepthOffset2)&kHcalDepthMask2; + } + } }; std::ostream& operator<<(std::ostream&,const HcalDetId& id); diff --git a/DataFormats/HcalDetId/interface/HcalElectronicsId.h b/DataFormats/HcalDetId/interface/HcalElectronicsId.h index 6fa7c2401cb55..727062df2f4b0 100644 --- a/DataFormats/HcalDetId/interface/HcalElectronicsId.h +++ b/DataFormats/HcalDetId/interface/HcalElectronicsId.h @@ -32,70 +32,92 @@ class HcalElectronicsId { public: /** Default constructor -- invalid value */ - HcalElectronicsId(); + constexpr HcalElectronicsId() + : hcalElectronicsId_{0xffffffffu} + {} /** from raw */ - HcalElectronicsId(uint32_t); + constexpr HcalElectronicsId(uint32_t id) + : hcalElectronicsId_{id} + {} /** VME Constructor from fiberchan,fiber index,spigot,dccid */ - HcalElectronicsId(int fiberChan, int fiberIndex, int spigot, int dccid); + constexpr HcalElectronicsId(int fiberChan, int fiberIndex, int spigot, int dccid) + : hcalElectronicsId_((uint32_t)((fiberChan&0x3) | (((fiberIndex-1)&0x7)<<2) | + ((spigot&0xF)<<5) | ((dccid&0x1F)<<9))) + {} /** VME Constructor from slb channel,slb site,spigot,dccid */ - HcalElectronicsId(int slbChan, int slbSite, int spigot, int dccid, int crate, int slot, int tb); + constexpr HcalElectronicsId(int slbChan, int slbSite, int spigot, + int dccid, int crate, int slot, int tb) + : hcalElectronicsId_((uint32_t)((slbChan&0x3) | (((slbSite)&0x7)<<2) | + ((spigot&0xF)<<5) | ((dccid&0x1F)<<9))) { + hcalElectronicsId_|=((tb&0x1)<<19) | ((slot&0x1f)<<14) | ((crate&0x3f)<<20); + hcalElectronicsId_|=0x02000000; + } /** uTCA constructor */ - HcalElectronicsId(int crate, int slot, int fiber, int fiberchan, bool isTrigger); + constexpr HcalElectronicsId(int crate, int slot, int fiber, int fc, bool isTrigger) + : hcalElectronicsId_((int)((fc&0xF) | (((fiber)&0x1F)<<4) | + ((slot&0xF)<<9) | ((crate&0x3F)<<13))) { + if (isTrigger) hcalElectronicsId_|=0x02000000; + hcalElectronicsId_|=0x04000000; + } - uint32_t operator()() { return hcalElectronicsId_; } + constexpr uint32_t operator()() { return hcalElectronicsId_; } - uint32_t rawId() const { return hcalElectronicsId_; } + constexpr uint32_t rawId() const { return hcalElectronicsId_; } - bool isVMEid() const { return (hcalElectronicsId_&0x04000000)==0; } - bool isUTCAid() const { return (hcalElectronicsId_&0x04000000)!=0; } - bool isTriggerChainId() const { return (hcalElectronicsId_&0x02000000)!=0; } + constexpr bool isVMEid() const { return (hcalElectronicsId_&0x04000000)==0; } + constexpr bool isUTCAid() const { return (hcalElectronicsId_&0x04000000)!=0; } + constexpr bool isTriggerChainId() const { return (hcalElectronicsId_&0x02000000)!=0; } /** Set the VME htr-related information 1=top, 0=bottom*/ - void setHTR(int crate, int slot, int tb); + constexpr void setHTR(int crate, int slot, int tb) { + if (isUTCAid()) return; // cannot do this for uTCA + hcalElectronicsId_&=0x3FFF; // keep the readout chain info + hcalElectronicsId_|=((tb&0x1)<<19) | ((slot&0x1f)<<14) | ((crate&0x3f)<<20); + } /// get subtype for this channel (valid for uTCA only) - int subtype() const { return (isUTCAid())?((hcalElectronicsId_>>21)&0x1F):(-1); } + constexpr int subtype() const { return (isUTCAid())?((hcalElectronicsId_>>21)&0x1F):(-1); } /// get the fiber channel id (which of channels on a fiber) - int fiberChanId() const { return (isVMEid())?(hcalElectronicsId_&0x3):(hcalElectronicsId_&0xF); } + constexpr int fiberChanId() const { return (isVMEid())?(hcalElectronicsId_&0x3):(hcalElectronicsId_&0xF); } /// get the fiber index. For VME 1-8 (which of eight fibers carried by a spigot), for uTCA fibers are zero-based - int fiberIndex() const { return (isVMEid())?(((hcalElectronicsId_>>2)&0x7)+1):((hcalElectronicsId_>>4)&0x1F); } + constexpr int fiberIndex() const { return (isVMEid())?(((hcalElectronicsId_>>2)&0x7)+1):((hcalElectronicsId_>>4)&0x1F); } /// get the SLB channel index (valid only for VME trigger-chain ids) - int slbChannelIndex() const { return hcalElectronicsId_&0x3; } + constexpr int slbChannelIndex() const { return hcalElectronicsId_&0x3; } /// get the SLB site number (valid only for VME trigger-chain ids) - int slbSiteNumber() const { return ((hcalElectronicsId_>>2)&0x7); } + constexpr int slbSiteNumber() const { return ((hcalElectronicsId_>>2)&0x7); } /// get the HTR channel id (1-24) - int htrChanId() const { return isVMEid()?((fiberChanId()+1)+((fiberIndex()-1)*3)):(0); } + constexpr int htrChanId() const { return isVMEid()?((fiberChanId()+1)+((fiberIndex()-1)*3)):(0); } /// get the HTR-wide slb channel code (letter plus number) std::string slbChannelCode() const; /// get the spigot (input number on DCC, AMC card number for uTCA) - int spigot() const { return (isVMEid())?((hcalElectronicsId_>>5)&0xF):slot(); } + constexpr int spigot() const { return (isVMEid())?((hcalElectronicsId_>>5)&0xF):slot(); } /// get the (Hcal local) DCC id for VME, crate number for uTCA - int dccid() const { return (isVMEid())?((hcalElectronicsId_>>9)&0x1F):crateId(); } + constexpr int dccid() const { return (isVMEid())?((hcalElectronicsId_>>9)&0x1F):crateId(); } /// get the htr slot - int htrSlot() const { return slot(); } + constexpr int htrSlot() const { return slot(); } /// get the htr or uHTR slot - int slot() const { return (isVMEid())?((hcalElectronicsId_>>14)&0x1F):((hcalElectronicsId_>>9)&0xF); } + constexpr int slot() const { return (isVMEid())?((hcalElectronicsId_>>14)&0x1F):((hcalElectronicsId_>>9)&0xF); } /// get the htr top/bottom (1=top/0=bottom), valid for VME - int htrTopBottom() const { return (isVMEid())?((hcalElectronicsId_>>19)&0x1):(-1); } + constexpr int htrTopBottom() const { return (isVMEid())?((hcalElectronicsId_>>19)&0x1):(-1); } /// get the readout VME crate number - int readoutVMECrateId() const { return crateId(); } + constexpr int readoutVMECrateId() const { return crateId(); } /// get the readout VME crate number - int crateId() const { return (isVMEid())?((hcalElectronicsId_>>20)&0x1F):((hcalElectronicsId_>>13)&0x3F); } + constexpr int crateId() const { return (isVMEid())?((hcalElectronicsId_>>20)&0x1F):((hcalElectronicsId_>>13)&0x3F); } /// get a fast, compact, unique index for linear lookups - int linearIndex() const { return (isVMEid())?((hcalElectronicsId_)&0x3FFF):((hcalElectronicsId_)&0x7FFFF); } + constexpr int linearIndex() const { return (isVMEid())?((hcalElectronicsId_)&0x3FFF):((hcalElectronicsId_)&0x7FFFF); } static const int maxLinearIndex = 0x7FFFF; // static const int maxDCCId = 31; /** Equality operator */ - int operator==(const HcalElectronicsId& id) const { return id.hcalElectronicsId_==hcalElectronicsId_; } + constexpr int operator==(const HcalElectronicsId& id) const { return id.hcalElectronicsId_==hcalElectronicsId_; } /** Non-Equality operator */ - int operator!=(const HcalElectronicsId& id) const { return id.hcalElectronicsId_!=hcalElectronicsId_; } + constexpr int operator!=(const HcalElectronicsId& id) const { return id.hcalElectronicsId_!=hcalElectronicsId_; } /// Compare the id to another id for use in a map - int operator<(const HcalElectronicsId& id) const { return hcalElectronicsId_ #include const HcalDetId HcalDetId::Undefined(HcalEmpty,0,0,0); -HcalDetId::HcalDetId() : DetId() { -} - -HcalDetId::HcalDetId(uint32_t rawid) { - if ((DetId::Detector(rawid>>DetId::kDetOffset)&DetId::kDetMask) != Hcal) { - id_ = rawid; - } else { - HcalSubdetector subdet = (HcalSubdetector)((rawid>>DetId::kSubdetOffset)&DetId::kSubdetMask); - if ((subdet==HcalBarrel) || (subdet==HcalEndcap) || - (subdet==HcalOuter) || (subdet==HcalForward)) { - id_ = newForm(rawid); - } else { - id_ = rawid; - } - } -} - -HcalDetId::HcalDetId(HcalSubdetector subdet, int tower_ieta, int tower_iphi, int depth) : DetId(Hcal,subdet) { - // (no checking at this point!) - id_ |= (kHcalIdFormat2) | ((depth&kHcalDepthMask2)<0)?(kHcalZsideMask2|(tower_ieta<0)?(kHcalZsideMask1|(eta<0)?(kHcalZsideMask2|(eta<>kHcalEtaOffset1)&kHcalEtaMask1; - else return (id_>>kHcalEtaOffset2)&kHcalEtaMask2; -} - -int HcalDetId::iphi() const { - if (oldFormat()) return id_&kHcalPhiMask1; - else return id_&kHcalPhiMask2; -} - -int HcalDetId::depth() const { - if (oldFormat()) return (id_>>kHcalDepthOffset1)&kHcalDepthMask1; - else return (id_>>kHcalDepthOffset2)&kHcalDepthMask2; -} - -int HcalDetId::hfdepth() const { - int dep = depth(); - if (subdet() == HcalForward) { - if (dep > 2) dep -= 2; - } - return dep; -} - -uint32_t HcalDetId::maskDepth() const { - if (oldFormat()) return (id_|kHcalDepthSet1); - else return (id_|kHcalDepthSet2); -} - -uint32_t HcalDetId::otherForm() const { - uint32_t rawid = (id_&kHcalIdMask); - if (oldFormat()) { - rawid = newForm(id_); - } else { - rawid |= ((depth()&kHcalDepthMask1)<0)?(kHcalZsideMask1|(ieta()<0)?(kHcalZsideMask2|(eta< 2) dep -= 2; - bool result = ((zsid==zside()) && (eta==ietaAbs()) && (phi==iphi()) && - (dep==hfdepth())); - return result; -} - -HcalDetId HcalDetId::baseDetId() const { - if (subdet() != HcalForward || depth() <= 2) { - return HcalDetId(id_); - } else { - int zsid, eta, phi, dep; - unpackId(id_, zsid, eta, phi, dep); - dep -= 2; - uint32_t rawid = id_&kHcalIdMask; - rawid |= (kHcalIdFormat2) | ((dep&kHcalDepthMask2)<0)?(kHcalZsideMask2|(eta< 2) { - return HcalDetId(id_); - } else { - int zsid, eta, phi, dep; - unpackId(id_, zsid, eta, phi, dep); - dep += 2; - uint32_t rawid = id_&kHcalIdMask; - rawid |= (kHcalIdFormat2) | ((dep&kHcalDepthMask2)<0)?(kHcalZsideMask2|(eta<360)?(simple_iphi-360):(simple_iphi)); -} - -int HcalDetId::crystal_iphi_high() const { - int simple_iphi=((iphi()-1)*5)+5; - simple_iphi+=10; - return ((simple_iphi>360)?(simple_iphi-360):(simple_iphi)); -} - -void HcalDetId::newFromOld(const uint32_t& rawid) { - id_ = newForm(rawid); -} - -void HcalDetId::unpackId(const uint32_t& rawid, int& zsid, int& eta, int& phi, - int& dep) { - if ((rawid&kHcalIdFormat2)==0) { - zsid = (rawid&kHcalZsideMask1)?(1):(-1); - eta = (rawid>>kHcalEtaOffset1)&kHcalEtaMask1; - phi = rawid&kHcalPhiMask1; - dep = (rawid>>kHcalDepthOffset1)&kHcalDepthMask1; - } else { - zsid = (rawid&kHcalZsideMask2)?(1):(-1); - eta = (rawid>>kHcalEtaOffset2)&kHcalEtaMask2; - phi = rawid&kHcalPhiMask2; - dep = (rawid>>kHcalDepthOffset2)&kHcalDepthMask2; - } -} - std::ostream& operator<<(std::ostream& s,const HcalDetId& id) { switch (id.subdet()) { case(HcalBarrel) : return s << "(HB " << id.ieta() << ',' << id.iphi() << ',' << id.depth() << ')'; diff --git a/DataFormats/HcalDetId/src/HcalElectronicsId.cc b/DataFormats/HcalDetId/src/HcalElectronicsId.cc index 4c267643e5885..0e7f912b9cf3d 100644 --- a/DataFormats/HcalDetId/src/HcalElectronicsId.cc +++ b/DataFormats/HcalDetId/src/HcalElectronicsId.cc @@ -1,32 +1,5 @@ #include "DataFormats/HcalDetId/interface/HcalElectronicsId.h" -HcalElectronicsId::HcalElectronicsId() { - hcalElectronicsId_=0xffffffffu; -} - -HcalElectronicsId::HcalElectronicsId(uint32_t id) { - hcalElectronicsId_=id; -} - -HcalElectronicsId::HcalElectronicsId(int fiberChan, int fiberIndex, int spigot, int dccid) { - hcalElectronicsId_=(fiberChan&0x3) | (((fiberIndex-1)&0x7)<<2) | - ((spigot&0xF)<<5) | ((dccid&0x1F)<<9); -} - -HcalElectronicsId::HcalElectronicsId(int slbChan, int slbSite, int spigot, int dccid, int crate, int slot, int tb) { - hcalElectronicsId_=(slbChan&0x3) | (((slbSite)&0x7)<<2) | - ((spigot&0xF)<<5) | ((dccid&0x1F)<<9); - hcalElectronicsId_|=((tb&0x1)<<19) | ((slot&0x1f)<<14) | ((crate&0x3f)<<20); - hcalElectronicsId_|=0x02000000; -} - -HcalElectronicsId::HcalElectronicsId(int crate, int slot, int fiber, int fc, bool isTrigger) { - hcalElectronicsId_=(fc&0xF) | (((fiber)&0x1F)<<4) | - ((slot&0xF)<<9) | ((crate&0x3F)<<13); - if (isTrigger) hcalElectronicsId_|=0x02000000; - hcalElectronicsId_|=0x04000000; -} - std::string HcalElectronicsId::slbChannelCode() const { std::string retval; if (isTriggerChainId() && isVMEid()) { @@ -49,12 +22,6 @@ std::string HcalElectronicsId::slbChannelCode() const { return retval; } -void HcalElectronicsId::setHTR(int crate, int slot, int tb) { - if (isUTCAid()) return; // cannot do this for uTCA - hcalElectronicsId_&=0x3FFF; // keep the readout chain info - hcalElectronicsId_|=((tb&0x1)<<19) | ((slot&0x1f)<<14) | ((crate&0x3f)<<20); -} - std::ostream& operator<<(std::ostream& os,const HcalElectronicsId& id) { if (id.isUTCAid()) { if (id.isTriggerChainId()) os << "UTCA(trigger): "; @@ -71,5 +38,3 @@ std::ostream& operator<<(std::ostream& os,const HcalElectronicsId& id) { } } } - - diff --git a/DataFormats/HcalDigi/interface/HBHEDataFrame.h b/DataFormats/HcalDigi/interface/HBHEDataFrame.h index 9afc182d87731..d4d392595ad52 100644 --- a/DataFormats/HcalDigi/interface/HBHEDataFrame.h +++ b/DataFormats/HcalDigi/interface/HBHEDataFrame.h @@ -4,7 +4,6 @@ #include "DataFormats/HcalDetId/interface/HcalDetId.h" #include "DataFormats/HcalDetId/interface/HcalElectronicsId.h" #include "DataFormats/HcalDigi/interface/HcalQIESample.h" -#include #include /** \class HBHEDataFrame @@ -16,40 +15,79 @@ class HBHEDataFrame { public: typedef HcalDetId key_type; ///< For the sorted collection - HBHEDataFrame(); // for persistence - explicit HBHEDataFrame(const HcalDetId& id); + constexpr HBHEDataFrame() + : id_(0), size_(0), hcalPresamples_(0) + {} + constexpr explicit HBHEDataFrame(const HcalDetId& id) : + id_(id), size_(0), hcalPresamples_(0) + { + // TODO : test id for HcalBarrel or HcalEndcap + } - const HcalDetId& id() const { return id_; } - const HcalElectronicsId& elecId() const { return electronicsId_; } + constexpr const HcalDetId& id() const { return id_; } + constexpr const HcalElectronicsId& elecId() const { return electronicsId_; } /// total number of samples in the digi - int size() const { return size_&0xF; } + constexpr int size() const { return size_&0xF; } /// number of samples before the sample from the triggered beam crossing (according to the hardware) - int presamples() const { return hcalPresamples_&0xF; } + constexpr int presamples() const { return hcalPresamples_&0xF; } /// was ZS MarkAndPass? - bool zsMarkAndPass() const { return (hcalPresamples_&0x10); } + constexpr bool zsMarkAndPass() const { return (hcalPresamples_&0x10); } /// was ZS unsuppressed? - bool zsUnsuppressed() const { return (hcalPresamples_&0x20); } + constexpr bool zsUnsuppressed() const { return (hcalPresamples_&0x20); } /// zs crossing mask (which sums considered) - uint32_t zsCrossingMask() const { return (hcalPresamples_&0x3FF000)>>12; } + constexpr uint32_t zsCrossingMask() const { return (hcalPresamples_&0x3FF000)>>12; } /// access a sample - const HcalQIESample& operator[](int i) const { return data_[i]; } + constexpr HcalQIESample const& operator[](int i) const { return data_[i]; } /// access a sample - const HcalQIESample& sample(int i) const { return data_[i]; } + constexpr HcalQIESample const& sample(int i) const { return data_[i]; } /// offset of bunch number for this channel relative to nominal set in the unpacker (range is +7->-7. -1000 indicates the data is invalid/unavailable) - int fiberIdleOffset() const; + constexpr int fiberIdleOffset() const { + int val=(hcalPresamples_&0xF00)>>8; + return (val==0)?(-1000):(((val&0x8)==0)?(-(val&0x7)):(val&0x7)); + } /// validate appropriate DV and ER bits as well as capid rotation for the specified samples (default is all) - bool validate(int firstSample=0, int nSamples=100) const; + constexpr bool validate(int firstSample, int nSamples) const { + int capid=-1; + bool ok=true; + for (int i=0; ok && iMAXSAMPLES) size_=MAXSAMPLES; + else if (size<=0) size_=0; + else size_=size; + } + constexpr void setPresamples(int ps) { + hcalPresamples_|=ps&0xF; + } + constexpr void setZSInfo(bool unsuppressed, bool markAndPass, + uint32_t crossingMask) { + hcalPresamples_&=0x7FC00F0F; // preserve actual presamples and fiber idle offset + if (markAndPass) hcalPresamples_|=0x10; + if (unsuppressed) hcalPresamples_|=0x20; + hcalPresamples_|=(crossingMask&0x3FF)<<12; + } + constexpr void setSample(int i, const HcalQIESample& sam) { data_[i]=sam; } + constexpr void setReadoutIds(const HcalElectronicsId& eid) { + electronicsId_=eid; + } + constexpr void setFiberIdleOffset(int offset) { + hcalPresamples_&=0x7FFFF0FF; + if (offset>=7) hcalPresamples_|=0xF00; + else if (offset>=0) hcalPresamples_|=(0x800)|(offset<<8); + else if (offset>=-7) hcalPresamples_|=((-offset)<<8); + else hcalPresamples_|=0x700; + } static const int MAXSAMPLES = 10; private: diff --git a/DataFormats/HcalDigi/interface/HFDataFrame.h b/DataFormats/HcalDigi/interface/HFDataFrame.h index 40bf0b749a844..41b9117fee35e 100644 --- a/DataFormats/HcalDigi/interface/HFDataFrame.h +++ b/DataFormats/HcalDigi/interface/HFDataFrame.h @@ -4,7 +4,6 @@ #include "DataFormats/HcalDetId/interface/HcalDetId.h" #include "DataFormats/HcalDetId/interface/HcalElectronicsId.h" #include "DataFormats/HcalDigi/interface/HcalQIESample.h" -#include #include /** \class HFDataFrame @@ -16,40 +15,78 @@ class HFDataFrame { public: typedef HcalDetId key_type; ///< For the sorted collection - HFDataFrame(); // for persistence - explicit HFDataFrame(const HcalDetId& id); - - const HcalDetId& id() const { return id_; } - const HcalElectronicsId& elecId() const { return electronicsId_; } + constexpr HFDataFrame() + : id_(0), size_(0), hcalPresamples_(0) + {} + constexpr explicit HFDataFrame(const HcalDetId& id) : + id_(id), size_(0), hcalPresamples_(0) + {// TODO : test id for HcalForward + } + + constexpr HcalDetId const& id() const { return id_; } + constexpr HcalElectronicsId const& elecId() const { return electronicsId_; } /// total number of samples in the digi - int size() const { return size_&0xF; } + constexpr int size() const { return size_&0xF; } /// number of samples before the sample from the triggered beam crossing (according to the hardware) - int presamples() const { return hcalPresamples_&0xF; } + constexpr int presamples() const { return hcalPresamples_&0xF; } /// was ZS MarkAndPass? - bool zsMarkAndPass() const { return (hcalPresamples_&0x10); } + constexpr bool zsMarkAndPass() const { return (hcalPresamples_&0x10); } /// was ZS unsuppressed? - bool zsUnsuppressed() const { return (hcalPresamples_&0x20); } + constexpr bool zsUnsuppressed() const { return (hcalPresamples_&0x20); } /// zs crossing mask (which sums considered) - uint32_t zsCrossingMask() const { return (hcalPresamples_&0x3FF000)>>12; } + constexpr uint32_t zsCrossingMask() const { return (hcalPresamples_&0x3FF000)>>12; } /// access a sample - const HcalQIESample& operator[](int i) const { return data_[i]; } + constexpr HcalQIESample const& operator[](int i) const { return data_[i]; } /// access a sample - const HcalQIESample& sample(int i) const { return data_[i]; } + constexpr HcalQIESample const& sample(int i) const { return data_[i]; } /// offset of bunch number for this channel relative to nominal set in the unpacker (range is +7->-7. -1000 indicates the data is invalid/unavailable) - int fiberIdleOffset() const; + constexpr int fiberIdleOffset() const { + int val=(hcalPresamples_&0xF00)>>8; + return (val==0)?(-1000):(((val&0x8)==0)?(-(val&0x7)):(val&0x7)); + } /// validate appropriate DV and ER bits as well as capid rotation for the specified samples (default is all) - bool validate(int firstSample=0, int nSamples=100) const; + constexpr bool validate(int firstSample, int nSamples) const { + int capid=-1; + bool ok=true; + for (int i=0; ok && iMAXSAMPLES) size_=MAXSAMPLES; + else if (size<=0) size_=0; + else size_=size; + } + constexpr void setPresamples(int ps) { + hcalPresamples_|=ps&0xF; + } + constexpr void setZSInfo(bool unsuppressed, bool markAndPass, + uint32_t crossingMask) { + hcalPresamples_&=0x7FC00F0F; // preserve actual presamples and fiber idle offset + if (markAndPass) hcalPresamples_|=0x10; + if (unsuppressed) hcalPresamples_|=0x20; + hcalPresamples_|=(crossingMask&0x3FF)<<12; + } + constexpr void setSample(int i, const HcalQIESample& sam) { data_[i]=sam; } + constexpr void setReadoutIds(const HcalElectronicsId& eid) { + electronicsId_=eid; + } + constexpr void setFiberIdleOffset(int offset) { + hcalPresamples_&=0x7FFFF0FF; + if (offset>=7) hcalPresamples_|=0xF00; + else if (offset>=0) hcalPresamples_|=(0x800)|(offset<<8); + else if (offset>=-7) hcalPresamples_|=((-offset)<<8); + else hcalPresamples_|=0x700; + } static const int MAXSAMPLES = 10; private: diff --git a/DataFormats/HcalDigi/interface/HODataFrame.h b/DataFormats/HcalDigi/interface/HODataFrame.h index 0c131c1588a4a..4a7b96d0662c6 100644 --- a/DataFormats/HcalDigi/interface/HODataFrame.h +++ b/DataFormats/HcalDigi/interface/HODataFrame.h @@ -4,7 +4,6 @@ #include "DataFormats/HcalDetId/interface/HcalDetId.h" #include "DataFormats/HcalDetId/interface/HcalElectronicsId.h" #include "DataFormats/HcalDigi/interface/HcalQIESample.h" -#include #include @@ -17,40 +16,77 @@ class HODataFrame { public: typedef HcalDetId key_type; ///< For the sorted collection - HODataFrame(); // for persistence - explicit HODataFrame(const HcalDetId& id); + HODataFrame() + : id_(0), size_(0), hcalPresamples_(0) + {} + explicit constexpr HODataFrame(const HcalDetId& id) + : id_(id), size_(0), hcalPresamples_(0) + {// TODO : test id for HcalOuter + } - const HcalDetId& id() const { return id_; } - const HcalElectronicsId& elecId() const { return electronicsId_; } + constexpr HcalDetId const& id() const { return id_; } + constexpr HcalElectronicsId const& elecId() const { return electronicsId_; } /// total number of samples in the digi - int size() const { return size_&0xF; } + constexpr int size() const { return size_&0xF; } /// number of samples before the sample from the triggered beam crossing (according to the hardware) - int presamples() const { return hcalPresamples_&0xF; } + constexpr int presamples() const { return hcalPresamples_&0xF; } /// was ZS MarkAndPass? - bool zsMarkAndPass() const { return (hcalPresamples_&0x10); } + constexpr bool zsMarkAndPass() const { return (hcalPresamples_&0x10); } /// was ZS unsuppressed? - bool zsUnsuppressed() const { return (hcalPresamples_&0x20); } + constexpr bool zsUnsuppressed() const { return (hcalPresamples_&0x20); } /// zs crossing mask (which sums considered) - uint32_t zsCrossingMask() const { return (hcalPresamples_&0x3FF000)>>12; } + constexpr uint32_t zsCrossingMask() const { return (hcalPresamples_&0x3FF000)>>12; } /// access a sample - const HcalQIESample& operator[](int i) const { return data_[i]; } + constexpr const HcalQIESample& operator[](int i) const { return data_[i]; } /// access a sample - const HcalQIESample& sample(int i) const { return data_[i]; } + constexpr const HcalQIESample& sample(int i) const { return data_[i]; } /// offset of bunch number for this channel relative to nominal set in the unpacker (range is +7->-7. -1000 indicates the data is invalid/unavailable) - int fiberIdleOffset() const; + constexpr int fiberIdleOffset() const { + int val=(hcalPresamples_&0xF00)>>8; + return (val==0)?(-1000):(((val&0x8)==0)?(-(val&0x7)):(val&0x7)); + } /// validate appropriate DV and ER bits as well as capid rotation for the specified samples (default is all) - bool validate(int firstSample=0, int nSamples=100) const; + constexpr bool validate(int firstSample, int nSamples) const { + int capid=-1; + bool ok=true; + for (int i=0; ok && iMAXSAMPLES) size_=MAXSAMPLES; + else if (size<=0) size_=0; + else size_=size; + } + constexpr void setPresamples(int ps) { + hcalPresamples_|=ps&0xF; + } + constexpr void setZSInfo(bool unsuppressed, bool markAndPass, uint32_t crossingMask) { + hcalPresamples_&=0x7FC00F0F; // preserve actual presamples and fiber idle offset + if (markAndPass) hcalPresamples_|=0x10; + if (unsuppressed) hcalPresamples_|=0x20; + hcalPresamples_|=(crossingMask&0x3FF)<<12; + } + constexpr void setSample(int i, const HcalQIESample& sam) { data_[i]=sam; } + constexpr void setReadoutIds(const HcalElectronicsId& eid) { + electronicsId_=eid; + } + constexpr void setFiberIdleOffset(int offset) { + hcalPresamples_&=0xFFFF0FF; + if (offset>=7) hcalPresamples_|=0xF00; + else if (offset>=0) hcalPresamples_|=(0x800)|(offset<<8); + else if (offset>=-7) hcalPresamples_|=((-offset)<<8); + else hcalPresamples_|=0x700; + } static const int MAXSAMPLES = 10; private: diff --git a/DataFormats/HcalDigi/interface/HcalQIESample.h b/DataFormats/HcalDigi/interface/HcalQIESample.h index c26c1b0c4ce9c..2de7450016f61 100644 --- a/DataFormats/HcalDigi/interface/HcalQIESample.h +++ b/DataFormats/HcalDigi/interface/HcalQIESample.h @@ -4,6 +4,34 @@ #include #include +static +#ifdef __CUDA_ARCH__ + __constant__ +#else + constexpr +#endif +float +nominal_adc2fc[128] = {-0.5f,0.5f,1.5f,2.5f,3.5f,4.5f,5.5f,6.5f,7.5f,8.5f,9.5f,10.5f,11.5f,12.5f,13.5f, + 15.0f,17.0f,19.0f,21.0f,23.0f,25.0f,27.0f, + 29.5f,32.5f,35.5f,38.5f, + 42.0f,46.0f,50.0f, + 54.5f,59.5f,64.5f, + 59.5f,64.5f,69.5f,74.5f,79.5f,84.5f,89.5f,94.5f,99.5f,104.5f,109.5f,114.5f,119.5f,124.5f,129.5f, + 137.0f,147.0f,157.0f,167.0f,177.0f,187.0f,197.0f, + 209.5f,224.5f,239.5f,254.5f, + 272.0f,292.0f,312.0f, + 334.5f,359.5f,384.5f, + 359.5f,384.5f,409.5f,434.5f,459.5f,484.5f,509.5f,534.5f,559.5f,584.5f,609.5f,634.5f,659.5f,684.5f,709.5f, + 747.0f,797.0f,847.0f,897.0f,947.0f,997.0f,1047.0f, + 1109.5f,1184.5f,1259.5f,1334.5f, + 1422.0f,1522.0f,1622.0f, + 1734.5f,1859.5f,1984.5f, + 1859.5f,1984.5f,2109.5f,2234.5f,2359.5f,2484.5f,2609.5f,2734.5f,2859.5f,2984.5f,3109.5f,3234.5f,3359.5f,3484.5f,3609.5f, + 3797.0f,4047.0f,4297.0f,4547.0f,4797.0f,5047.0f,5297.0f, + 5609.5f,5984.5f,6359.5f,6734.5f, + 7172.0f,7672.0f,8172.0f, + 8734.5f,9359.5f,9984.5f}; + /** \class HcalQIESample * Simple container packer/unpacker for a single QIE data word * @@ -12,31 +40,42 @@ */ class HcalQIESample { public: - HcalQIESample() { theSample=0; } - HcalQIESample(uint16_t data) { theSample=data; } - HcalQIESample(int adc, int capid, int fiber, int fiberchan, bool dv=true, bool er=false); + constexpr HcalQIESample() + : theSample{0} + {} + constexpr HcalQIESample(uint16_t data) + : theSample{data} + {} + constexpr HcalQIESample(int adc, int capid, int fiber, + int fiberchan, bool dv, bool er) + : theSample((uint16_t)((adc&0x7f) | ((capid&0x3)<<7) | + (((fiber-1)&0x7)<<13) | ((fiberchan&0x3)<<11) | + ((dv)?(0x0200):(0)) | ((er)?(0x0400):(0)))) + {} /// get the raw word - uint16_t raw() const { return theSample; } + constexpr uint16_t raw() const { return theSample; } /// get the ADC sample - int adc() const { return theSample&0x7F; } + constexpr int adc() const { return theSample&0x7F; } /// get the nominal FC (no calibrations applied) - double nominal_fC() const; + constexpr double nominal_fC() const { + return nominal_adc2fc[adc()]; + } /// get the Capacitor id - int capid() const { return (theSample>>7)&0x3; } + constexpr int capid() const { return (theSample>>7)&0x3; } /// is the Data Valid bit set? - bool dv() const { return (theSample&0x0200)!=0; } + constexpr bool dv() const { return (theSample&0x0200)!=0; } /// is the error bit set? - bool er() const { return (theSample&0x0400)!=0; } + constexpr bool er() const { return (theSample&0x0400)!=0; } /// get the fiber number - int fiber() const { return ((theSample>>13)&0x7)+1; } + constexpr int fiber() const { return ((theSample>>13)&0x7)+1; } /// get the fiber channel number - int fiberChan() const { return (theSample>>11)&0x3; } + constexpr int fiberChan() const { return (theSample>>11)&0x3; } /// get the id channel - int fiberAndChan() const { return (theSample>>11)&0x1F; } + constexpr int fiberAndChan() const { return (theSample>>11)&0x1F; } /// for streaming - uint16_t operator()() { return theSample; } + constexpr uint16_t operator()() { return theSample; } private: uint16_t theSample; diff --git a/DataFormats/HcalDigi/interface/QIE10DataFrame.h b/DataFormats/HcalDigi/interface/QIE10DataFrame.h index ffbfdf27b14f6..0109fe452fefe 100644 --- a/DataFormats/HcalDigi/interface/QIE10DataFrame.h +++ b/DataFormats/HcalDigi/interface/QIE10DataFrame.h @@ -15,16 +15,24 @@ class QIE10DataFrame { static const int HEADER_WORDS = 1; static const int FLAG_WORDS = 1; - QIE10DataFrame() { } - QIE10DataFrame(edm::DataFrame const & df) : m_data(df) { } + constexpr QIE10DataFrame() { } + constexpr QIE10DataFrame(edm::DataFrame const & df) : m_data(df) { } class Sample { public: typedef uint32_t wide_type; - Sample(const edm::DataFrame& frame, edm::DataFrame::size_type i) : word1_(frame[i]), word2_(frame[i+1]) { } - Sample(const edm::DataFrame::data_type& word1, const edm::DataFrame::data_type& word2) : word1_(word1), word2_(word2) {} - explicit Sample(wide_type wide); + constexpr Sample(const edm::DataFrame& frame, edm::DataFrame::size_type i) : word1_(frame[i]), word2_(frame[i+1]) { } + constexpr Sample(const edm::DataFrame::data_type& word1, const edm::DataFrame::data_type& word2) : word1_(word1), word2_(word2) {} + explicit Sample(const wide_type wide) + : word1_{0}, word2_{0} { + static_assert(sizeof(wide) == 2*sizeof(word1_), + "The wide input type must be able to contain two words"); + const edm::DataFrame::data_type* ptr = + reinterpret_cast(&wide); + word1_ = ptr[0]; + word2_ = ptr[1]; + } static const int MASK_ADC = 0xFF; static const int MASK_LE_TDC = 0x3F; @@ -35,57 +43,86 @@ class QIE10DataFrame { static const int MASK_CAPID = 0x3; static const int OFFSET_CAPID = 12; - inline int adc() const { return word1_&MASK_ADC; } - inline int le_tdc() const { return word2_&MASK_LE_TDC; } - inline int te_tdc() const { return (word2_>>OFFSET_TE_TDC)&MASK_TE_TDC; } - inline bool ok() const { return word1_&MASK_OK; } - inline bool soi() const { return word1_&MASK_SOI; } - inline int capid() const { return (word2_>>OFFSET_CAPID)&MASK_CAPID; } - inline edm::DataFrame::data_type raw(edm::DataFrame::size_type i) const + constexpr inline int adc() const { return word1_&MASK_ADC; } + constexpr inline int le_tdc() const { return word2_&MASK_LE_TDC; } + constexpr inline int te_tdc() const { return (word2_>>OFFSET_TE_TDC)&MASK_TE_TDC; } + constexpr inline bool ok() const { return word1_&MASK_OK; } + constexpr inline bool soi() const { return word1_&MASK_SOI; } + constexpr inline int capid() const { return (word2_>>OFFSET_CAPID)&MASK_CAPID; } + constexpr inline edm::DataFrame::data_type raw(edm::DataFrame::size_type i) const { return (i > WORDS_PER_SAMPLE) ? 0 : ( (i==1) ? word2_ : word1_ ); } - wide_type wideRaw() const; + QIE10DataFrame::Sample::wide_type wideRaw() const { + static_assert(sizeof(QIE10DataFrame::Sample::wide_type) == 2*sizeof(word1_), + "The wide result type must be able to contain two words"); + wide_type result = 0; + edm::DataFrame::data_type* ptr = + reinterpret_cast(&result); + ptr[0] = word1_; + ptr[1] = word2_; + return result; + } private: edm::DataFrame::data_type word1_; edm::DataFrame::data_type word2_; }; - void copyContent(const QIE10DataFrame& src); + constexpr void copyContent(const QIE10DataFrame& digi) { + for (edm::DataFrame::size_type i=0; i>OFFSET_FLAVOR)&MASK_FLAVOR); } + constexpr int flavor() const { return ((m_data[0]>>OFFSET_FLAVOR)&MASK_FLAVOR); } /// was there a link error? static const int MASK_LINKERROR = 0x800; - bool linkError() const { return m_data[0]&MASK_LINKERROR; } + constexpr bool linkError() const { return m_data[0]&MASK_LINKERROR; } /// was this a mark-and-pass ZS event? static const int MASK_MARKPASS = 0x100; - bool zsMarkAndPass() const {return m_data[0]&MASK_MARKPASS; } + constexpr bool zsMarkAndPass() const {return m_data[0]&MASK_MARKPASS; } /// set ZS params - void setZSInfo(bool markAndPass); + constexpr void setZSInfo(bool markAndPass) { + if(markAndPass) m_data[0] |= MASK_MARKPASS; + } /// get the sample - inline Sample operator[](edm::DataFrame::size_type i) const { return Sample(m_data,i*WORDS_PER_SAMPLE+HEADER_WORDS); } + constexpr inline Sample operator[](edm::DataFrame::size_type i) const { return Sample(m_data,i*WORDS_PER_SAMPLE+HEADER_WORDS); } /// set the sample contents - void setSample(edm::DataFrame::size_type isample, int adc, int le_tdc, int te_tdc, int capid, bool soi=false, bool ok=true); + constexpr void setSample(edm::DataFrame::size_type isample, int adc, + int le_tdc, int te_tdc, int capid, bool soi=false, + bool ok=true) { + if (isample>=size()) return; + m_data[isample*WORDS_PER_SAMPLE+HEADER_WORDS]=(adc&Sample::MASK_ADC)|(soi?(Sample::MASK_SOI):(0))|(ok?(Sample::MASK_OK):(0)); + m_data[isample*WORDS_PER_SAMPLE+HEADER_WORDS+1]=(le_tdc&Sample::MASK_LE_TDC)|((te_tdc&Sample::MASK_TE_TDC)<>OFFSET_TDC)&MASK_TDC; } - bool soi() const { return frame_[i_]&MASK_SOI; } - int capid() const { return ((((frame_[0]>>OFFSET_CAPID)&MASK_CAPID)+i_-HEADER_WORDS)&MASK_CAPID); } + constexpr int adc() const { return frame_[i_]&MASK_ADC; } + constexpr int tdc() const { return (frame_[i_]>>OFFSET_TDC)&MASK_TDC; } + constexpr bool soi() const { return frame_[i_]&MASK_SOI; } + constexpr int capid() const { return ((((frame_[0]>>OFFSET_CAPID)&MASK_CAPID)+i_-HEADER_WORDS)&MASK_CAPID); } private: const edm::DataFrame& frame_; edm::DataFrame::size_type i_; }; - void copyContent(const QIE11DataFrame&); + constexpr void copyContent(const QIE11DataFrame& digi) { + for (edm::DataFrame::size_type i=0; i>OFFSET_FLAVOR)&MASK_FLAVOR); } + constexpr int flavor() const { return ((m_data[0]>>OFFSET_FLAVOR)&MASK_FLAVOR); } /// was there a link error? static const int MASK_LINKERROR = 0x800; - bool linkError() const { return m_data[0]&MASK_LINKERROR; } + constexpr bool linkError() const { return m_data[0]&MASK_LINKERROR; } /// was there a capid rotation error? static const int MASK_CAPIDERROR = 0x400; - bool capidError() const { return m_data[0]&MASK_CAPIDERROR; } + constexpr bool capidError() const { return m_data[0]&MASK_CAPIDERROR; } /// was this a mark-and-pass ZS event? - bool zsMarkAndPass() const {return (flavor()==1); } + constexpr bool zsMarkAndPass() const {return (flavor()==1); } /// set ZS params - void setZSInfo(bool markAndPass); + constexpr void setZSInfo(bool markAndPass) { + if(markAndPass) m_data[0] |= (markAndPass&MASK_FLAVOR)<=size()) return; + m_data[isample+1]=(adc&Sample::MASK_ADC)|(soi?(Sample::MASK_SOI):(0))|((tdc&Sample::MASK_TDC)<MAXSAMPLES) size_=MAXSAMPLES; - else if (size<=0) size_=0; - else size_=size; -} -void HBHEDataFrame::setPresamples(int ps) { - hcalPresamples_|=ps&0xF; -} -void HBHEDataFrame::setReadoutIds(const HcalElectronicsId& eid) { - electronicsId_=eid; -} - -bool HBHEDataFrame::validate(int firstSample, int nSamples) const { - int capid=-1; - bool ok=true; - for (int i=0; ok && i>8; - return (val==0)?(-1000):(((val&0x8)==0)?(-(val&0x7)):(val&0x7)); -} - -void HBHEDataFrame::setFiberIdleOffset(int offset) { - hcalPresamples_&=0x7FFFF0FF; - if (offset>=7) hcalPresamples_|=0xF00; - else if (offset>=0) hcalPresamples_|=(0x800)|(offset<<8); - else if (offset>=-7) hcalPresamples_|=((-offset)<<8); - else hcalPresamples_|=0x700; -} - - std::ostream& operator<<(std::ostream& s, const HBHEDataFrame& digi) { s << digi.id() << " " << digi.size() << " samples " << digi.presamples() << " presamples "; if (digi.zsUnsuppressed()) s << " zsUS"; @@ -73,5 +13,3 @@ std::ostream& operator<<(std::ostream& s, const HBHEDataFrame& digi) { s << " " << digi.sample(i) << std::endl; return s; } - - diff --git a/DataFormats/HcalDigi/src/HFDataFrame.cc b/DataFormats/HcalDigi/src/HFDataFrame.cc index 631ae1c85fdec..c2cd07316a561 100644 --- a/DataFormats/HcalDigi/src/HFDataFrame.cc +++ b/DataFormats/HcalDigi/src/HFDataFrame.cc @@ -1,63 +1,5 @@ #include "DataFormats/HcalDigi/interface/HFDataFrame.h" -HFDataFrame::HFDataFrame() : id_(0), - size_(0), - hcalPresamples_(0) -{ -} - -HFDataFrame::HFDataFrame(const HcalDetId& id) : - id_(id), - size_(0), - hcalPresamples_(0) -{ - // TODO : test id for HcalForward -} - -void HFDataFrame::setSize(int size) { - if (size>MAXSAMPLES) size_=MAXSAMPLES; - else if (size<=0) size_=0; - else size_=size; -} -void HFDataFrame::setPresamples(int ps) { - hcalPresamples_|=ps&0xF; -} -void HFDataFrame::setReadoutIds(const HcalElectronicsId& eid) { - electronicsId_=eid; -} - -bool HFDataFrame::validate(int firstSample, int nSamples) const { - int capid=-1; - bool ok=true; - for (int i=0; ok && i>8; - return (val==0)?(-1000):(((val&0x8)==0)?(-(val&0x7)):(val&0x7)); -} - -void HFDataFrame::setFiberIdleOffset(int offset) { - hcalPresamples_&=0x7FFFF0FF; - if (offset>=7) hcalPresamples_|=0xF00; - else if (offset>=0) hcalPresamples_|=(0x800)|(offset<<8); - else if (offset>=-7) hcalPresamples_|=((-offset)<<8); - else hcalPresamples_|=0x700; -} - std::ostream& operator<<(std::ostream& s, const HFDataFrame& digi) { s << digi.id() << " " << digi.size() << " samples " << digi.presamples() << " presamples "; if (digi.zsUnsuppressed()) s << " zsUS "; diff --git a/DataFormats/HcalDigi/src/HODataFrame.cc b/DataFormats/HcalDigi/src/HODataFrame.cc index 9fa8b9f348db9..1604a5a919fe9 100644 --- a/DataFormats/HcalDigi/src/HODataFrame.cc +++ b/DataFormats/HcalDigi/src/HODataFrame.cc @@ -1,65 +1,5 @@ #include "DataFormats/HcalDigi/interface/HODataFrame.h" - -HODataFrame::HODataFrame() : id_(0), - size_(0), - hcalPresamples_(0) -{ -} - -HODataFrame::HODataFrame(const HcalDetId& id) : - id_(id), - size_(0), - hcalPresamples_(0) -{ - // TODO : test id for HcalOuter -} - -void HODataFrame::setSize(int size) { - if (size>MAXSAMPLES) size_=MAXSAMPLES; - else if (size<=0) size_=0; - else size_=size; -} -void HODataFrame::setPresamples(int ps) { - hcalPresamples_|=ps&0xF; -} -void HODataFrame::setReadoutIds(const HcalElectronicsId& eid) { - electronicsId_=eid; -} - -bool HODataFrame::validate(int firstSample, int nSamples) const { - int capid=-1; - bool ok=true; - for (int i=0; ok && i>8; - return (val==0)?(-1000):(((val&0x8)==0)?(-(val&0x7)):(val&0x7)); -} - -void HODataFrame::setFiberIdleOffset(int offset) { - hcalPresamples_&=0xFFFF0FF; - if (offset>=7) hcalPresamples_|=0xF00; - else if (offset>=0) hcalPresamples_|=(0x800)|(offset<<8); - else if (offset>=-7) hcalPresamples_|=((-offset)<<8); - else hcalPresamples_|=0x700; -} - std::ostream& operator<<(std::ostream& s, const HODataFrame& digi) { s << digi.id() << " " << digi.size() << " samples " << digi.presamples() << " presamples "; if (digi.zsUnsuppressed()) s << " zsUS "; diff --git a/DataFormats/HcalDigi/src/HcalQIESample.cc b/DataFormats/HcalDigi/src/HcalQIESample.cc index 93ba46825433a..6408b6d69c329 100644 --- a/DataFormats/HcalDigi/src/HcalQIESample.cc +++ b/DataFormats/HcalDigi/src/HcalQIESample.cc @@ -1,36 +1,5 @@ #include "DataFormats/HcalDigi/interface/HcalQIESample.h" -static const float nominal_adc2fc[128] = {-0.5f,0.5f,1.5f,2.5f,3.5f,4.5f,5.5f,6.5f,7.5f,8.5f,9.5f,10.5f,11.5f,12.5f,13.5f, - 15.0f,17.0f,19.0f,21.0f,23.0f,25.0f,27.0f, - 29.5f,32.5f,35.5f,38.5f, - 42.0f,46.0f,50.0f, - 54.5f,59.5f,64.5f, - 59.5f,64.5f,69.5f,74.5f,79.5f,84.5f,89.5f,94.5f,99.5f,104.5f,109.5f,114.5f,119.5f,124.5f,129.5f, - 137.0f,147.0f,157.0f,167.0f,177.0f,187.0f,197.0f, - 209.5f,224.5f,239.5f,254.5f, - 272.0f,292.0f,312.0f, - 334.5f,359.5f,384.5f, - 359.5f,384.5f,409.5f,434.5f,459.5f,484.5f,509.5f,534.5f,559.5f,584.5f,609.5f,634.5f,659.5f,684.5f,709.5f, - 747.0f,797.0f,847.0f,897.0f,947.0f,997.0f,1047.0f, - 1109.5f,1184.5f,1259.5f,1334.5f, - 1422.0f,1522.0f,1622.0f, - 1734.5f,1859.5f,1984.5f, - 1859.5f,1984.5f,2109.5f,2234.5f,2359.5f,2484.5f,2609.5f,2734.5f,2859.5f,2984.5f,3109.5f,3234.5f,3359.5f,3484.5f,3609.5f, - 3797.0f,4047.0f,4297.0f,4547.0f,4797.0f,5047.0f,5297.0f, - 5609.5f,5984.5f,6359.5f,6734.5f, - 7172.0f,7672.0f,8172.0f, - 8734.5f,9359.5f,9984.5f}; - -HcalQIESample::HcalQIESample(int adc, int capid, int fiber, int fiberchan, bool dv, bool er) { - theSample=(adc&0x7f) | ((capid&0x3)<<7) | - (((fiber-1)&0x7)<<13) | ((fiberchan&0x3)<<11) | - ((dv)?(0x0200):(0)) | ((er)?(0x0400):(0)); -} - -double HcalQIESample::nominal_fC() const { - return nominal_adc2fc[adc()]; -} - std::ostream& operator<<(std::ostream& s, const HcalQIESample& samp) { s << "ADC=" << samp.adc() << ", capid=" << samp.capid(); if (samp.er()) s << ", ER"; diff --git a/DataFormats/HcalDigi/src/QIE10DataFrame.cc b/DataFormats/HcalDigi/src/QIE10DataFrame.cc index 74a6fb46a2e29..85bc9643c384b 100644 --- a/DataFormats/HcalDigi/src/QIE10DataFrame.cc +++ b/DataFormats/HcalDigi/src/QIE10DataFrame.cc @@ -1,56 +1,6 @@ -#include - #include "DataFormats/HcalDigi/interface/QIE10DataFrame.h" #include "DataFormats/HcalDetId/interface/HcalGenericDetId.h" -void QIE10DataFrame::setSample(edm::DataFrame::size_type isample, int adc, int le_tdc, int te_tdc, int capid, bool soi, bool ok) { - if (isample>=size()) return; - m_data[isample*WORDS_PER_SAMPLE+HEADER_WORDS]=(adc&Sample::MASK_ADC)|(soi?(Sample::MASK_SOI):(0))|(ok?(Sample::MASK_OK):(0)); - m_data[isample*WORDS_PER_SAMPLE+HEADER_WORDS+1]=(le_tdc&Sample::MASK_LE_TDC)|((te_tdc&Sample::MASK_TE_TDC)<(&wide); - word1_ = ptr[0]; - word2_ = ptr[1]; -} - -QIE10DataFrame::Sample::wide_type QIE10DataFrame::Sample::wideRaw() const { - static_assert(sizeof(QIE10DataFrame::Sample::wide_type) == 2*sizeof(word1_), - "The wide result type must be able to contain two words"); - wide_type result = 0; - edm::DataFrame::data_type* ptr = - reinterpret_cast(&result); - ptr[0] = word1_; - ptr[1] = word2_; - return result; -} - std::ostream& operator<<(std::ostream& s, const QIE10DataFrame& digi) { if (digi.detid().det()==DetId::Hcal) { s << HcalGenericDetId(digi.detid()); diff --git a/DataFormats/HcalDigi/src/QIE11DataFrame.cc b/DataFormats/HcalDigi/src/QIE11DataFrame.cc index 8d9d4882a2995..95ab0575e1efb 100644 --- a/DataFormats/HcalDigi/src/QIE11DataFrame.cc +++ b/DataFormats/HcalDigi/src/QIE11DataFrame.cc @@ -1,38 +1,6 @@ #include "DataFormats/HcalDigi/interface/QIE11DataFrame.h" #include "DataFormats/HcalDetId/interface/HcalGenericDetId.h" -void QIE11DataFrame::setCapid0(int cap0) { - m_data[0]&=0xFCFF; // inversion of the capid0 mask - m_data[0]|=((cap0&Sample::MASK_CAPID)<=size()) return; - m_data[isample+1]=(adc&Sample::MASK_ADC)|(soi?(Sample::MASK_SOI):(0))|((tdc&Sample::MASK_TDC)< Date: Wed, 30 May 2018 11:36:22 +0200 Subject: [PATCH 2/4] Add CUDA tests for HcalDetId and HcalDigi dataformats The final test is: - create an `HBHEDigiCollection`/`HFDigiCollection`/`HODigiCollection` on the host side and fill it - allocate space on the gpu - transfer to the gpu - run the kernel to compute some func per channel - transfer the results back to host - assert for each channel --- DataFormats/HcalDetId/test/BuildFile.xml | 4 + DataFormats/HcalDetId/test/test_hcal_detid.cu | 79 ++++++++ DataFormats/HcalDigi/test/BuildFile.xml | 14 +- DataFormats/HcalDigi/test/test_hcal_digi.cu | 185 ++++++++++++++++++ 4 files changed, 278 insertions(+), 4 deletions(-) create mode 100644 DataFormats/HcalDetId/test/BuildFile.xml create mode 100644 DataFormats/HcalDetId/test/test_hcal_detid.cu create mode 100644 DataFormats/HcalDigi/test/test_hcal_digi.cu diff --git a/DataFormats/HcalDetId/test/BuildFile.xml b/DataFormats/HcalDetId/test/BuildFile.xml new file mode 100644 index 0000000000000..a876d1c99fc57 --- /dev/null +++ b/DataFormats/HcalDetId/test/BuildFile.xml @@ -0,0 +1,4 @@ + + + + diff --git a/DataFormats/HcalDetId/test/test_hcal_detid.cu b/DataFormats/HcalDetId/test/test_hcal_detid.cu new file mode 100644 index 0000000000000..ea2099ef2971f --- /dev/null +++ b/DataFormats/HcalDetId/test/test_hcal_detid.cu @@ -0,0 +1,79 @@ +#include +#include + +#include +#include +#include "DataFormats/DetId/interface/DetId.h" +#include "DataFormats/HcalDetId/interface/HcalDetId.h" + +__global__ void test_gen_detid(DetId* id) { + DetId did; + *id = did; +} + +__global__ void test_gen_hcal_detid(HcalDetId *id) { + HcalDetId did(HcalBarrel, 5, 5, 0); + *id = did; + + // trigger functions on the device + did.iphi(); + did.ieta(); + did.zside(); + did.subdet(); + did.ietaAbs(); + did.depth(); + did.hfdepth(); + did.maskDepth(); + did.baseDetId(); + did.secondAnodeId(); + did.crystal_ieta_low(); + did.crystal_ieta_high(); + did.crystal_iphi_low(); + did.crystal_iphi_high(); +} + +void test_detid() { + // test det ids + DetId h_id, h_id_test; + DetId h_test0{1}; + DetId *d_id; + + cudaMalloc((void**)&d_id, sizeof(DetId)); + cudaMemcpy(d_id, &h_id, sizeof(DetId), cudaMemcpyHostToDevice); + test_gen_detid<<<1,1>>>(d_id); + cudaMemcpy(&h_id_test, d_id, sizeof(DetId), cudaMemcpyDeviceToHost); + + assert(h_id_test == h_id); + assert(h_id != h_test0); +} + +void test_hcal_detid() { + HcalDetId h_id; + HcalDetId h_id_test0{HcalBarrel, 5, 5, 0}; + HcalDetId *d_id; + + cudaMalloc((void**)&d_id, sizeof(HcalDetId)); + cudaMemcpy(d_id, &h_id, sizeof(HcalDetId), cudaMemcpyHostToDevice); + test_gen_hcal_detid<<<1,1>>>(d_id); + cudaMemcpy(&h_id, d_id, sizeof(HcalDetId), cudaMemcpyDeviceToHost); + + std::cout << h_id_test0 << std::endl; + std::cout << h_id << std::endl; + assert(h_id_test0 == h_id); +} + +int main(int argc, char** argv) { + int nDevices; + cudaGetDeviceCount(&nDevices); + std::cout << "nDevices = " << nDevices << std::endl; + + // test det id functionality + if (nDevices>0) + test_detid(); + + // test hcal det ids + if (nDevices>0) + test_hcal_detid(); + + return 0; +} diff --git a/DataFormats/HcalDigi/test/BuildFile.xml b/DataFormats/HcalDigi/test/BuildFile.xml index 4764adcafaeea..6d65f320caedd 100644 --- a/DataFormats/HcalDigi/test/BuildFile.xml +++ b/DataFormats/HcalDigi/test/BuildFile.xml @@ -1,7 +1,13 @@ - - - - + + + + + + + + + + diff --git a/DataFormats/HcalDigi/test/test_hcal_digi.cu b/DataFormats/HcalDigi/test/test_hcal_digi.cu new file mode 100644 index 0000000000000..fafa52cbc6ee4 --- /dev/null +++ b/DataFormats/HcalDigi/test/test_hcal_digi.cu @@ -0,0 +1,185 @@ +#include +#include + +#include +#include +#include "DataFormats/DetId/interface/DetId.h" +#include "DataFormats/HcalDetId/interface/HcalDetId.h" +#include "DataFormats/HcalDigi/interface/HBHEDataFrame.h" +#include "DataFormats/HcalDigi/interface/QIE10DataFrame.h" +#include "DataFormats/HcalDigi/interface/QIE11DataFrame.h" +#include "DataFormats/HcalDigi/interface/HcalDigiCollections.h" +#include "DataFormats/Common/interface/DataFrame.h" + +__global__ void kernel_test_hcal_qiesample(HcalQIESample* sample, uint16_t value) { + printf("kernel: testing hcal qie sampel\n"); + printf("%f %f %f\n", nominal_adc2fc[0], nominal_adc2fc[1], nominal_adc2fc[2]); + + HcalQIESample tmp{value}; + *sample = tmp; +} + +__global__ void kernel_test_hcal_qie8_hbhedf(HBHEDataFrame *df) { + printf("kernel: testing hcal hbhe dataframe\n"); + df->setSize(10); + for (auto i=0; i<10; i++) + df->setSample(i, HcalQIESample(100)); + df->setReadoutIds(HcalElectronicsId(100)); +} + +void test_hcal_qiesample() { + HcalQIESample h_sample, h_test_sample0{100}, h_test_sample1; + HcalQIESample *d_sample; + + cudaMalloc((void**)&d_sample, sizeof(HcalQIESample)); + cudaMemcpy(d_sample, &h_sample, sizeof(HcalQIESample), cudaMemcpyHostToDevice); + kernel_test_hcal_qiesample<<<1,1>>>(d_sample, 100); + cudaMemcpy(&h_sample, d_sample, sizeof(HcalQIESample), cudaMemcpyDeviceToHost); + + assert(h_sample() == h_test_sample0()); + assert(h_sample() != h_test_sample1()); +} + +template +__global__ void kernel_test_hcal_qie8_digis(TDF *pdfs, uint32_t* out) { + int id = threadIdx.x; + uint32_t sum = 0; + for (auto i=0; i<10; i++) + sum += pdfs[id].sample(i).raw(); + out[id] = sum; +} + +template +__global__ void kernel_test_hcal_qie1011_digis(uint16_t* pdfs, uint32_t* out, int samples) { + printf("kernel: testing hcal qie1011 df\n"); + int id = threadIdx.x; + uint32_t sum=0; + int nwords = TDF::WORDS_PER_SAMPLE*samples + TDF::HEADER_WORDS + TDF::FLAG_WORDS; + TDF df(edm::DataFrame(0, pdfs + id*nwords, nwords)); + for (auto i=0; i< df.samples(); i++) { + sum += df[i].adc(); + } + + out[id] = sum; +} + +template +void test_hcal_qie1011_digis() { + constexpr int size = 10; + constexpr int samples = 10; + constexpr int detid = 2; + HcalDataFrameContainer coll{samples, detid}; + TDF *d_dfs; + uint16_t *d_data; + uint32_t *d_out; + uint32_t h_out[size], h_test_out[size]; + for (auto i=0; i<<<1, size>>>(d_data, d_out, samples); + cudaDeviceSynchronize(); + auto code = cudaGetLastError(); + if (code != cudaSuccess) + std::cout << cudaGetErrorString(code); + cudaMemcpy(&h_out, d_out, size * sizeof(uint32_t), cudaMemcpyDeviceToHost); + + // comparison + for (auto i=0; i +void test_hcal_qie8_digis() { + constexpr int n = 10; + edm::SortedCollection coll{n}; + TDF *d_dfs; + uint32_t *d_out; + uint32_t h_out[n], h_test_out[n]; + for (auto i=0; i>>(d_dfs, d_out); + cudaMemcpy(&h_out, d_out, n * sizeof(uint32_t), cudaMemcpyDeviceToHost); + + std::cout << "collection size = " << coll.size() << std::endl; + + // comparison + for (auto i=0; i>>(d_df); + cudaMemcpy(&h_df, d_df, sizeof(HBHEDataFrame), cudaMemcpyDeviceToHost); + + assert(h_df.size() == h_test_df.size()); + assert(h_df.elecId() == h_test_df.elecId()); + for (auto i=0; i<10; i++) + assert(h_df[i].raw() == h_test_df[i].raw()); +} + +int main(int argc, char** argv) { + int nDevices; + cudaGetDeviceCount(&nDevices); + std::cout << "nDevices = " << nDevices << std::endl; + + if (nDevices > 0) { + // qie8 + test_hcal_qiesample(); + test_hcal_qie8_hbhedf(); + test_hcal_qie8_digis(); + test_hcal_qie8_digis(); + test_hcal_qie8_digis(); + + // qie1011 + test_hcal_qie1011_digis(); + test_hcal_qie1011_digis(); + } + + return 0; +} From dc81c738843dc28d36ce1dff077dc2269668ed18 Mon Sep 17 00:00:00 2001 From: Viktor Khristenko Date: Wed, 30 May 2018 14:14:16 +0200 Subject: [PATCH 3/4] fixing default parameters for qie sample --- DataFormats/HcalDigi/interface/HcalQIESample.h | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/DataFormats/HcalDigi/interface/HcalQIESample.h b/DataFormats/HcalDigi/interface/HcalQIESample.h index 2de7450016f61..c05ab44e9a31c 100644 --- a/DataFormats/HcalDigi/interface/HcalQIESample.h +++ b/DataFormats/HcalDigi/interface/HcalQIESample.h @@ -47,7 +47,7 @@ class HcalQIESample { : theSample{data} {} constexpr HcalQIESample(int adc, int capid, int fiber, - int fiberchan, bool dv, bool er) + int fiberchan, bool dv=true, bool er=false) : theSample((uint16_t)((adc&0x7f) | ((capid&0x3)<<7) | (((fiber-1)&0x7)<<13) | ((fiberchan&0x3)<<11) | ((dv)?(0x0200):(0)) | ((er)?(0x0400):(0)))) From 6cc610f2a8692054d0ff176e1dda85b41fe4225c Mon Sep 17 00:00:00 2001 From: Viktor Khristenko Date: Wed, 30 May 2018 15:55:02 +0200 Subject: [PATCH 4/4] fixing default parameters for all qie8 data frames --- DataFormats/HcalDigi/interface/HBHEDataFrame.h | 4 ++-- DataFormats/HcalDigi/interface/HFDataFrame.h | 4 ++-- DataFormats/HcalDigi/interface/HODataFrame.h | 4 ++-- 3 files changed, 6 insertions(+), 6 deletions(-) diff --git a/DataFormats/HcalDigi/interface/HBHEDataFrame.h b/DataFormats/HcalDigi/interface/HBHEDataFrame.h index d4d392595ad52..5ff9ef96ea233 100644 --- a/DataFormats/HcalDigi/interface/HBHEDataFrame.h +++ b/DataFormats/HcalDigi/interface/HBHEDataFrame.h @@ -50,7 +50,7 @@ class HBHEDataFrame { } /// validate appropriate DV and ER bits as well as capid rotation for the specified samples (default is all) - constexpr bool validate(int firstSample, int nSamples) const { + constexpr bool validate(int firstSample=0, int nSamples=100) const { int capid=-1; bool ok=true; for (int i=0; ok && i