#include <TrackingRecHit2DHeterogeneous.h>

Public Types
using	PhiBinner = TrackingRecHit2DSOAView::PhiBinner

template<typename T >
using	unique_ptr = typename Traits::template unique_ptr< T >

Public Member Functions
auto	hitsLayerStart ()

auto	hitsModuleStart () const

cms::cuda::host::unique_ptr< uint32_t[]>	hitsModuleStartToHostAsync (cudaStream_t stream) const

auto	iphi ()

cms::cuda::host::unique_ptr< float[]>	localCoordToHostAsync (cudaStream_t stream) const

auto	nHits () const

TrackingRecHit2DHeterogeneous &	operator= (const TrackingRecHit2DHeterogeneous &)=delete

TrackingRecHit2DHeterogeneous &	operator= (TrackingRecHit2DHeterogeneous &&)=default

auto	phiBinner ()

	TrackingRecHit2DHeterogeneous ()=default

	TrackingRecHit2DHeterogeneous (const TrackingRecHit2DHeterogeneous &)=delete

	TrackingRecHit2DHeterogeneous (TrackingRecHit2DHeterogeneous &&)=default

	TrackingRecHit2DHeterogeneous (uint32_t nHits, pixelCPEforGPU::ParamsOnGPU const cpeParams, uint32_t const hitsModuleStart, cudaStream_t stream)

TrackingRecHit2DSOAView *	view ()

TrackingRecHit2DSOAView const *	view () const

	~TrackingRecHit2DHeterogeneous ()=default

Private Attributes
unique_ptr< TrackingRecHit2DSOAView::AverageGeometry >	m_AverageGeometryStore

uint32_t *	m_hitsLayerStart

uint32_t const *	m_hitsModuleStart

int16_t *	m_iphi

uint32_t	m_nHits

PhiBinner *	m_phiBinner

unique_ptr< TrackingRecHit2DSOAView::PhiBinner >	m_PhiBinnerStore

unique_ptr< uint16_t[]>	m_store16

unique_ptr< float[]>	m_store32

unique_ptr< TrackingRecHit2DSOAView >	m_view

Static Private Attributes
static constexpr uint32_t	n16 = 4

static constexpr uint32_t	n32 = 9

Detailed Description

template<typename Traits>
class TrackingRecHit2DHeterogeneous< Traits >

Definition at line 8 of file TrackingRecHit2DHeterogeneous.h.

Member Typedef Documentation

◆ PhiBinner

template<typename Traits >

using TrackingRecHit2DHeterogeneous< Traits >::PhiBinner = TrackingRecHit2DSOAView::PhiBinner

Definition at line 13 of file TrackingRecHit2DHeterogeneous.h.

◆ unique_ptr

template<typename Traits >

template<typename T >

using TrackingRecHit2DHeterogeneous< Traits >::unique_ptr = typename Traits::template unique_ptr<T>

Definition at line 11 of file TrackingRecHit2DHeterogeneous.h.

Constructor & Destructor Documentation

◆ TrackingRecHit2DHeterogeneous() [1/4]

template<typename Traits >

TrackingRecHit2DHeterogeneous< Traits >::TrackingRecHit2DHeterogeneous ( )

default

◆ TrackingRecHit2DHeterogeneous() [2/4]

template<typename Traits >

TrackingRecHit2DHeterogeneous< Traits >::TrackingRecHit2DHeterogeneous	(	uint32_t	nHits,
		pixelCPEforGPU::ParamsOnGPU const *	cpeParams,
		uint32_t const *	hitsModuleStart,
		cudaStream_t	stream
	)

explicit

Definition at line 70 of file TrackingRecHit2DHeterogeneous.h.

     : m_nHits(nHits), m_hitsModuleStart(hitsModuleStart) {
   auto view = Traits::template make_host_unique<TrackingRecHit2DSOAView>(stream);
  
   view->m_nHits = nHits;
   m_view = Traits::template make_device_unique<TrackingRecHit2DSOAView>(stream);
   m_AverageGeometryStore = Traits::template make_device_unique<TrackingRecHit2DSOAView::AverageGeometry>(stream);
   view->m_averageGeometry = m_AverageGeometryStore.get();
   view->m_cpeParams = cpeParams;
   view->m_hitsModuleStart = hitsModuleStart;
  
   // if empy do not bother
   if (0 == nHits) {
     if constexpr (std::is_same<Traits, cms::cudacompat::GPUTraits>::value) {
       cms::cuda::copyAsync(m_view, view, stream);
     } else {
       m_view.reset(view.release());  // NOLINT: std::move() breaks CUDA version
     }
     return;
   }
  
   // the single arrays are not 128 bit alligned...
   // the hits are actually accessed in order only in building
   // if ordering is relevant they may have to be stored phi-ordered by layer or so
   // this will break 1to1 correspondence with cluster and module locality
   // so unless proven VERY inefficient we keep it ordered as generated
   m_store16 = Traits::template make_device_unique<uint16_t[]>(nHits * n16, stream);
   m_store32 = Traits::template make_device_unique<float[]>(nHits * n32 + 11, stream);
   m_PhiBinnerStore = Traits::template make_device_unique<TrackingRecHit2DSOAView::PhiBinner>(stream);
  
   auto get16 = [&](int i) { return m_store16.get() + i * nHits; };
   auto get32 = [&](int i) { return m_store32.get() + i * nHits; };
  
   // copy all the pointers
   m_phiBinner = view->m_phiBinner = m_PhiBinnerStore.get();
  
   view->m_xl = get32(0);
   view->m_yl = get32(1);
   view->m_xerr = get32(2);
   view->m_yerr = get32(3);
  
   view->m_xg = get32(4);
   view->m_yg = get32(5);
   view->m_zg = get32(6);
   view->m_rg = get32(7);
  
   m_iphi = view->m_iphi = reinterpret_cast<int16_t*>(get16(0));
  
   view->m_charge = reinterpret_cast<int32_t*>(get32(8));
   view->m_xsize = reinterpret_cast<int16_t*>(get16(2));
   view->m_ysize = reinterpret_cast<int16_t*>(get16(3));
   view->m_detInd = get16(1);
  
   m_hitsLayerStart = view->m_hitsLayerStart = reinterpret_cast<uint32_t*>(get32(n32));
  
   // transfer view
   if constexpr (std::is_same<Traits, cms::cudacompat::GPUTraits>::value) {
     cms::cuda::copyAsync(m_view, view, stream);
   } else {
     m_view.reset(view.release());  // NOLINT: std::move() breaks CUDA version
   }
 }

◆ ~TrackingRecHit2DHeterogeneous()

template<typename Traits >

TrackingRecHit2DHeterogeneous< Traits >::~TrackingRecHit2DHeterogeneous ( )

default

◆ TrackingRecHit2DHeterogeneous() [3/4]

template<typename Traits >

TrackingRecHit2DHeterogeneous< Traits >::TrackingRecHit2DHeterogeneous ( const TrackingRecHit2DHeterogeneous< Traits > & )

delete

◆ TrackingRecHit2DHeterogeneous() [4/4]

template<typename Traits >

TrackingRecHit2DHeterogeneous< Traits >::TrackingRecHit2DHeterogeneous ( TrackingRecHit2DHeterogeneous< Traits > && )

default

Member Function Documentation

◆ hitsLayerStart()

template<typename Traits >

auto TrackingRecHit2DHeterogeneous< Traits >::hitsLayerStart ( )

inline

Definition at line 35 of file TrackingRecHit2DHeterogeneous.h.

35 { return m_hitsLayerStart; }

References TrackingRecHit2DHeterogeneous< Traits >::m_hitsLayerStart.

◆ hitsModuleStart()

template<typename Traits >

auto TrackingRecHit2DHeterogeneous< Traits >::hitsModuleStart ( ) const

inline

Definition at line 34 of file TrackingRecHit2DHeterogeneous.h.

34 { return m_hitsModuleStart; }

References TrackingRecHit2DHeterogeneous< Traits >::m_hitsModuleStart.

Referenced by TrackingRecHit2DHeterogeneous< Traits >::TrackingRecHit2DHeterogeneous().

◆ hitsModuleStartToHostAsync()

cms::cuda::host::unique_ptr< uint32_t[]> TrackingRecHit2DCUDA::hitsModuleStartToHostAsync ( cudaStream_t stream ) const

Definition at line 15 of file TrackingRecHit2DHeterogeneous.cc.

                                                                                                             {
   auto ret = cms::cuda::make_host_unique<uint32_t[]>(gpuClustering::maxNumModules + 1, stream);
   cudaCheck(cudaMemcpyAsync(
       ret.get(), m_hitsModuleStart, sizeof(uint32_t) * (gpuClustering::maxNumModules + 1), cudaMemcpyDefault, stream));
   return ret;
 }

References cudaCheck, TrackingRecHit2DHeterogeneous< Traits >::m_hitsModuleStart, gpuClustering::maxNumModules, runTheMatrix::ret, and cms::cuda::stream.

◆ iphi()

template<typename Traits >

auto TrackingRecHit2DHeterogeneous< Traits >::iphi ( )

inline

Definition at line 37 of file TrackingRecHit2DHeterogeneous.h.

37 { return m_iphi; }

References TrackingRecHit2DHeterogeneous< Traits >::m_iphi.

◆ localCoordToHostAsync()

cms::cuda::host::unique_ptr< float[]> TrackingRecHit2DCUDA::localCoordToHostAsync ( cudaStream_t stream ) const

Definition at line 8 of file TrackingRecHit2DHeterogeneous.cc.

                                                                                                     {
   auto ret = cms::cuda::make_host_unique<float[]>(4 * nHits(), stream);
   cms::cuda::copyAsync(ret, m_store32, 4 * nHits(), stream);
   return ret;
 }