diff --git a/Event/DAQEvent/include/Event/ODIN.h b/Event/DAQEvent/include/Event/ODIN.h
index d5bbd3f4f0ddbaeacb761b8444d036125378114a..e95104d3dfe0c81a97297a139eabb6acd510a2ae 100644
--- a/Event/DAQEvent/include/Event/ODIN.h
+++ b/Event/DAQEvent/include/Event/ODIN.h
@@ -18,6 +18,13 @@ namespace LHCb {
#include <cstdint>
#include <cstring>
+// define macros to target host and device build for accelerators
+#if defined( __CUDACC__ ) || defined( __HIP__ )
+# define ACCEL_TARGET_SPEC __host__ __device__
+#else
+# define ACCEL_TARGET_SPEC
+#endif
+
namespace LHCb {
#ifndef ODIN_WITHOUT_GAUDI
@@ -33,7 +40,7 @@ namespace LHCb {
namespace ODINImplementation::details {
/// Helper to extract COUNT bits starting from OFFSET in a buffer.
template <unsigned int COUNT, unsigned int OFFSET>
- auto get_bits( LHCb::span<const std::uint32_t> data ) {
+ ACCEL_TARGET_SPEC auto get_bits( LHCb::span<const std::uint32_t> data ) {
static_assert( COUNT != 0 && COUNT <= 64, "invalid COUNT parameter" );
if constexpr ( COUNT == 1 ) {
return ( data[OFFSET / 32] & ( 1 << OFFSET % 32 ) ) ? true : false;
@@ -60,7 +67,7 @@ namespace LHCb {
}
/// Helper to set COUNT bits starting from OFFSET in a buffer using the passed value.
template <unsigned int COUNT, unsigned int OFFSET, typename VALUE>
- void set_bits( LHCb::span<std::uint32_t> data, VALUE value ) {
+ ACCEL_TARGET_SPEC void set_bits( LHCb::span<std::uint32_t> data, VALUE value ) {
static_assert( COUNT != 0 && COUNT <= 64, "invalid COUNT parameter" );
if constexpr ( COUNT == 1 ) {
if ( value ) {
@@ -144,8 +151,9 @@ namespace LHCb {
std::array<std::uint32_t, BANK_SIZE / sizeof( std::uint32_t )> data{0};
- ODIN() = default;
- ODIN( LHCb::span<const std::uint32_t> buffer ) {
+ // Note that this cannot be `= default` to support accelerator builds
+ ACCEL_TARGET_SPEC ODIN() {}
+ ACCEL_TARGET_SPEC ODIN( LHCb::span<const std::uint32_t> buffer ) {
assert( buffer.size() == data.size() );
std::memcpy( data.data(), buffer.data(), sizeof( data ) );
}
@@ -199,84 +207,98 @@ namespace LHCb {
EventNumberOffset = 8 * 32
};
- auto runNumber() const { return details::get_bits<RunNumberSize, RunNumberOffset>( data ); }
+ ACCEL_TARGET_SPEC auto runNumber() const { return details::get_bits<RunNumberSize, RunNumberOffset>( data ); }
void setRunNumber( std::uint32_t value ) { details::set_bits<RunNumberSize, RunNumberOffset>( data, value ); }
- auto eventType() const { return details::get_bits<EventTypeSize, EventTypeOffset>( data ); }
+ ACCEL_TARGET_SPEC auto eventType() const { return details::get_bits<EventTypeSize, EventTypeOffset>( data ); }
void setEventType( std::uint16_t value ) { details::set_bits<EventTypeSize, EventTypeOffset>( data, value ); }
void setEventType( EventTypes value ) { setEventType( static_cast<std::uint16_t>( value ) ); }
- auto calibrationStep() const { return details::get_bits<CalibrationStepSize, CalibrationStepOffset>( data ); }
+ ACCEL_TARGET_SPEC auto calibrationStep() const {
+ return details::get_bits<CalibrationStepSize, CalibrationStepOffset>( data );
+ }
void setCalibrationStep( std::uint16_t value ) {
details::set_bits<CalibrationStepSize, CalibrationStepOffset>( data, value );
}
- auto gpsTime() const { return details::get_bits<GpsTimeSize, GpsTimeOffset>( data ); }
+ ACCEL_TARGET_SPEC auto gpsTime() const { return details::get_bits<GpsTimeSize, GpsTimeOffset>( data ); }
void setGpsTime( std::uint64_t value ) { details::set_bits<GpsTimeSize, GpsTimeOffset>( data, value ); }
- auto triggerConfigurationKey() const {
+ ACCEL_TARGET_SPEC auto triggerConfigurationKey() const {
return details::get_bits<TriggerConfigurationKeySize, TriggerConfigurationKeyOffset>( data );
}
void setTriggerConfigurationKey( std::uint32_t value ) {
details::set_bits<TriggerConfigurationKeySize, TriggerConfigurationKeyOffset>( data, value );
}
- auto partitionID() const { return details::get_bits<PartitionIDSize, PartitionIDOffset>( data ); }
+ ACCEL_TARGET_SPEC auto partitionID() const {
+ return details::get_bits<PartitionIDSize, PartitionIDOffset>( data );
+ }
void setPartitionID( std::uint32_t value ) {
details::set_bits<PartitionIDSize, PartitionIDOffset>( data, value );
}
- auto bunchId() const { return details::get_bits<BunchIdSize, BunchIdOffset>( data ); }
- void setBunchId( std::uint16_t value ) { details::set_bits<BunchIdSize, BunchIdOffset>( data, value ); }
- BXTypes bunchCrossingType() const {
+ ACCEL_TARGET_SPEC auto bunchId() const { return details::get_bits<BunchIdSize, BunchIdOffset>( data ); }
+ void setBunchId( std::uint16_t value ) { details::set_bits<BunchIdSize, BunchIdOffset>( data, value ); }
+ ACCEL_TARGET_SPEC BXTypes bunchCrossingType() const {
return static_cast<BXTypes>( details::get_bits<BunchCrossingTypeSize, BunchCrossingTypeOffset>( data ) );
}
void setBunchCrossingType( BXTypes value ) {
details::set_bits<BunchCrossingTypeSize, BunchCrossingTypeOffset>( data, static_cast<std::uint8_t>( value ) );
}
- auto nonZeroSuppressionMode() const {
+ ACCEL_TARGET_SPEC auto nonZeroSuppressionMode() const {
return details::get_bits<NonZeroSuppressionModeSize, NonZeroSuppressionModeOffset>( data );
}
void setNonZeroSuppressionMode( bool value ) {
details::set_bits<NonZeroSuppressionModeSize, NonZeroSuppressionModeOffset>( data, value );
}
- auto timeAlignmentEventCentral() const {
+ ACCEL_TARGET_SPEC auto timeAlignmentEventCentral() const {
return details::get_bits<TimeAlignmentEventCentralSize, TimeAlignmentEventCentralOffset>( data );
}
void setTimeAlignmentEventCentral( bool value ) {
details::set_bits<TimeAlignmentEventCentralSize, TimeAlignmentEventCentralOffset>( data, value );
}
- auto timeAlignmentEventWindow() const {
+ ACCEL_TARGET_SPEC auto timeAlignmentEventWindow() const {
return details::get_bits<TimeAlignmentEventWindowSize, TimeAlignmentEventWindowOffset>( data );
}
void setTimeAlignmentEventWindow( std::uint8_t value ) {
details::set_bits<TimeAlignmentEventWindowSize, TimeAlignmentEventWindowOffset>( data, value );
}
- auto stepRunEnable() const { return details::get_bits<StepRunEnableSize, StepRunEnableOffset>( data ); }
+ ACCEL_TARGET_SPEC auto stepRunEnable() const {
+ return details::get_bits<StepRunEnableSize, StepRunEnableOffset>( data );
+ }
void setStepRunEnable( bool value ) { details::set_bits<StepRunEnableSize, StepRunEnableOffset>( data, value ); }
- auto triggerType() const { return details::get_bits<TriggerTypeSize, TriggerTypeOffset>( data ); }
+ ACCEL_TARGET_SPEC auto triggerType() const {
+ return details::get_bits<TriggerTypeSize, TriggerTypeOffset>( data );
+ }
void setTriggerType( std::uint8_t value ) {
details::set_bits<TriggerTypeSize, TriggerTypeOffset>( data, value );
}
void setTriggerType( TriggerTypes value ) { setTriggerType( static_cast<std::uint8_t>( value ) ); }
- auto timeAlignmentEventFirst() const {
+ ACCEL_TARGET_SPEC auto timeAlignmentEventFirst() const {
return details::get_bits<TimeAlignmentEventFirstSize, TimeAlignmentEventFirstOffset>( data );
}
void setTimeAlignmentEventFirst( bool value ) {
details::set_bits<TimeAlignmentEventFirstSize, TimeAlignmentEventFirstOffset>( data, value );
}
- auto calibrationType() const { return details::get_bits<CalibrationTypeSize, CalibrationTypeOffset>( data ); }
+ ACCEL_TARGET_SPEC auto calibrationType() const {
+ return details::get_bits<CalibrationTypeSize, CalibrationTypeOffset>( data );
+ }
void setCalibrationType( std::uint8_t value ) {
details::set_bits<CalibrationTypeSize, CalibrationTypeOffset>( data, value );
}
void setCalibrationType( CalibrationTypes value ) { setCalibrationType( static_cast<std::uint8_t>( value ) ); }
- auto orbitNumber() const { return details::get_bits<OrbitNumberSize, OrbitNumberOffset>( data ); }
+ ACCEL_TARGET_SPEC auto orbitNumber() const {
+ return details::get_bits<OrbitNumberSize, OrbitNumberOffset>( data );
+ }
void setOrbitNumber( std::uint32_t value ) {
details::set_bits<OrbitNumberSize, OrbitNumberOffset>( data, value );
}
- auto eventNumber() const { return details::get_bits<EventNumberSize, EventNumberOffset>( data ); }
+ ACCEL_TARGET_SPEC auto eventNumber() const {
+ return details::get_bits<EventNumberSize, EventNumberOffset>( data );
+ }
void setEventNumber( std::uint64_t value ) {
details::set_bits<EventNumberSize, EventNumberOffset>( data, value );
}