Inheritance diagram for TestCUDAProducerGPUEWTask:

Public Member Functions
void	acquire (edm::Event const &iEvent, edm::EventSetup const &iSetup, edm::WaitingTaskWithArenaHolder waitingTaskHolder) override

void	produce (edm::Event &iEvent, const edm::EventSetup &iSetup) override

	TestCUDAProducerGPUEWTask (edm::ParameterSet const &iConfig)

	~TestCUDAProducerGPUEWTask () override=default

Public Member Functions inherited from edm::stream::EDProducer< edm::ExternalWork >
	EDProducer ()=default

bool	hasAbilityToProduceInBeginLumis () const final

bool	hasAbilityToProduceInBeginRuns () const final

bool	hasAbilityToProduceInEndLumis () const final

bool	hasAbilityToProduceInEndRuns () const final

Static Public Member Functions
static void	fillDescriptions (edm::ConfigurationDescriptions &descriptions)

Private Member Functions
void	addSimpleWork (edm::EventNumber_t eventID, edm::StreamID streamID, cms::cuda::ScopedContextTask &ctx)

Private Attributes
cms::cuda::ContextState	ctxState_

cms::cuda::device::unique_ptr< float[]>	devicePtr_

const edm::EDPutTokenT< cms::cuda::Product< cms::cudatest::Thing > >	dstToken_

TestCUDAProducerGPUKernel	gpuAlgo_

cms::cuda::host::noncached::unique_ptr< float >	hostData_

const std::string	label_

const edm::EDGetTokenT< cms::cuda::Product< cms::cudatest::Thing > >	srcToken_

Additional Inherited Members
Public Types inherited from edm::stream::EDProducer< edm::ExternalWork >
typedef CacheContexts< T... >	CacheTypes

typedef CacheTypes::GlobalCache	GlobalCache

typedef AbilityChecker< T... >	HasAbility

typedef CacheTypes::LuminosityBlockCache	LuminosityBlockCache

typedef LuminosityBlockContextT< LuminosityBlockCache, RunCache, GlobalCache >	LuminosityBlockContext

typedef CacheTypes::LuminosityBlockSummaryCache	LuminosityBlockSummaryCache

typedef CacheTypes::RunCache	RunCache

typedef RunContextT< RunCache, GlobalCache >	RunContext

typedef CacheTypes::RunSummaryCache	RunSummaryCache

Detailed Description

Definition at line 24 of file TestCUDAProducerGPUEWTask.cc.

Constructor & Destructor Documentation

◆ TestCUDAProducerGPUEWTask()

TestCUDAProducerGPUEWTask::TestCUDAProducerGPUEWTask ( edm::ParameterSet const & iConfig )

explicit

Definition at line 48 of file TestCUDAProducerGPUEWTask.cc.

     : label_{iConfig.getParameter<std::string>("@module_label")},
       srcToken_{consumes<cms::cuda::Product<cms::cudatest::Thing>>(iConfig.getParameter<edm::InputTag>("src"))},
       dstToken_{produces<cms::cuda::Product<cms::cudatest::Thing>>()} {
   edm::Service<CUDAService> cs;
   if (cs->enabled()) {
     hostData_ = cms::cuda::make_host_noncached_unique<float>();
   }
 }

References edm::ParameterSet::getParameter(), and AlCaHLTBitMon_QueryRunRegistry::string.

◆ ~TestCUDAProducerGPUEWTask()

TestCUDAProducerGPUEWTask::~TestCUDAProducerGPUEWTask ( )

overridedefault

Member Function Documentation

◆ acquire()

void TestCUDAProducerGPUEWTask::acquire	(	edm::Event const &	iEvent,
		edm::EventSetup const &	iSetup,
		edm::WaitingTaskWithArenaHolder	waitingTaskHolder
	)

override

Definition at line 71 of file TestCUDAProducerGPUEWTask.cc.

                                                                                          {
   edm::LogVerbatim("TestCUDAProducerGPUEWTask") << label_ << " TestCUDAProducerGPUEWTask::acquire begin event "
                                                 << iEvent.id().event() << " stream " << iEvent.streamID();
  
   auto const& in = iEvent.get(srcToken_);
   cms::cuda::ScopedContextAcquire ctx{in, waitingTaskHolder, ctxState_};
  
   cms::cudatest::Thing const& input = ctx.get(in);
  
   devicePtr_ = gpuAlgo_.runAlgo(label_, input.get(), ctx.stream());
   // Mimick the need to transfer some of the GPU data back to CPU to
   // be used for something within this module, or to be put in the
   // event.
   cudaCheck(
       cudaMemcpyAsync(hostData_.get(), devicePtr_.get() + 10, sizeof(float), cudaMemcpyDeviceToHost, ctx.stream()));
   // Push a task to run addSimpleWork() after the asynchronous work
   // (and acquire()) has finished instead of produce()
   ctx.pushNextTask([iev = iEvent.id().event(), istr = iEvent.streamID(), this](cms::cuda::ScopedContextTask ctx) {
     addSimpleWork(iev, istr, ctx);
   });
  
   edm::LogVerbatim("TestCUDAProducerGPUEWTask") << label_ << " TestCUDAProducerGPUEWTask::acquire end event "
                                                 << iEvent.id().event() << " stream " << iEvent.streamID();
 }

References addSimpleWork(), ctxState_, cudaCheck, devicePtr_, dqmMemoryStats::float, gpuAlgo_, hostData_, iEvent, recoMuon::in, input, label_, TestCUDAProducerGPUKernel::runAlgo(), and srcToken_.

◆ addSimpleWork()

void TestCUDAProducerGPUEWTask::addSimpleWork	(	edm::EventNumber_t	eventID,
		edm::StreamID	streamID,
		cms::cuda::ScopedContextTask &	ctx
	)

private

Definition at line 98 of file TestCUDAProducerGPUEWTask.cc.

                                                                              {
   if (*hostData_ < 13) {
     edm::LogVerbatim("TestCUDAProducerGPUEWTask")
         << label_ << " TestCUDAProducerGPUEWTask::addSimpleWork begin event " << eventID << " stream " << streamID
         << " 10th element " << *hostData_ << " not satisfied, queueing more work";
     cudaCheck(
         cudaMemcpyAsync(hostData_.get(), devicePtr_.get() + 10, sizeof(float), cudaMemcpyDeviceToHost, ctx.stream()));
  
     ctx.pushNextTask(
         [eventID, streamID, this](cms::cuda::ScopedContextTask ctx) { addSimpleWork(eventID, streamID, ctx); });
     gpuAlgo_.runSimpleAlgo(devicePtr_.get(), ctx.stream());
     edm::LogVerbatim("TestCUDAProducerGPUEWTask")
         << label_ << " TestCUDAProducerGPUEWTask::addSimpleWork end event " << eventID << " stream " << streamID;
   } else {
     edm::LogVerbatim("TestCUDAProducerGPUEWTask")
         << label_ << " TestCUDAProducerGPUEWTask::addSimpleWork event " << eventID << " stream " << streamID
         << " 10th element " << *hostData_ << " not queueing more work";
   }
 }

References cudaCheck, devicePtr_, dqmMemoryStats::float, gpuAlgo_, hostData_, label_, cms::cuda::ScopedContextTask::pushNextTask(), TestCUDAProducerGPUKernel::runSimpleAlgo(), and cms::cuda::impl::ScopedContextBase::stream().

Referenced by acquire().

◆ fillDescriptions()

void TestCUDAProducerGPUEWTask::fillDescriptions ( edm::ConfigurationDescriptions & descriptions )

static

Definition at line 58 of file TestCUDAProducerGPUEWTask.cc.

                                                                                            {
   edm::ParameterSetDescription desc;
   desc.add<edm::InputTag>("src", edm::InputTag());
   descriptions.addWithDefaultLabel(desc);
   descriptions.setComment(
       "This EDProducer is part of the TestCUDAProducer* family. It models a GPU algorithm this is not the first "
       "algorithm in the chain of the GPU EDProducers, and that transfers some data from GPU to CPU multiple times "
       "alternating the transfers and kernel executions (e.g. to decide which kernel to run next based on a value from "
       "GPU). A synchronization between GPU and CPU is needed after each transfer. The synchronizations are implemented "
       "with the ExternalWork extension and explicit TBB tasks within the module. Produces "
       "cms::cuda::Product<cms::cudatest::Thing>.");
 }

References edm::ParameterSetDescription::add(), edm::ConfigurationDescriptions::addWithDefaultLabel(), HLT_2018_cff::InputTag, and edm::ConfigurationDescriptions::setComment().

◆ produce()

void TestCUDAProducerGPUEWTask::produce	(	edm::Event &	iEvent,
		const edm::EventSetup &	iSetup
	)

override

Definition at line 120 of file TestCUDAProducerGPUEWTask.cc.

                                                                                      {
   edm::LogVerbatim("TestCUDAProducerGPUEWTask")
       << label_ << " TestCUDAProducerGPUEWTask::produce begin event " << iEvent.id().event() << " stream "
       << iEvent.streamID() << " 10th element " << *hostData_;
   if (*hostData_ != 13) {
     throw cms::Exception("Assert") << "Expecting 10th element to be 13, got " << *hostData_;
   }
  
   cms::cuda::ScopedContextProduce ctx{ctxState_};
  
   ctx.emplace(iEvent, dstToken_, std::move(devicePtr_));
  
   edm::LogVerbatim("TestCUDAProducerGPUEWTask") << label_ << " TestCUDAProducerGPUEWTask::produce end event "
                                                 << iEvent.id().event() << " stream " << iEvent.streamID();
 }