diff --git a/DataFormats/SoATemplate/interface/SoACommon.h b/DataFormats/SoATemplate/interface/SoACommon.h index 8ab80fc88c697..49f5049377b44 100644 --- a/DataFormats/SoATemplate/interface/SoACommon.h +++ b/DataFormats/SoATemplate/interface/SoACommon.h @@ -52,6 +52,8 @@ #define _VALUE_TYPE_EIGEN_COLUMN 2 /* The size type need to be "hardcoded" in the template parameters for classes serialized by ROOT */ +/* In practice, using a typedef as a template parameter to the Layout or its ViewTemplateFreeParams member + * declaration fails ROOT dictionary generation. */ #define CMS_SOA_BYTE_SIZE_TYPE std::size_t namespace cms::soa { @@ -132,6 +134,8 @@ namespace cms::soa { return reinterpret_cast(addr) % alignment; } + TupleOrPointerType tupleOrPointer() { return addr_; } + public: // scalar or column ValueType const* addr_ = nullptr; @@ -166,6 +170,8 @@ namespace cms::soa { return reinterpret_cast(addr) % alignment; } + TupleOrPointerType tupleOrPointer() { return {addr_, stride_}; } + public: // address and stride ScalarType const* addr_ = nullptr; @@ -201,6 +207,8 @@ namespace cms::soa { return reinterpret_cast(addr) % alignment; } + TupleOrPointerType tupleOrPointer() { return addr_; } + public: // scalar or column ValueType* addr_ = nullptr; @@ -234,6 +242,8 @@ namespace cms::soa { return reinterpret_cast(addr) % alignment; } + TupleOrPointerType tupleOrPointer() { return {addr_, stride_}; } + public: // address and stride ScalarType* addr_ = nullptr; diff --git a/DataFormats/SoATemplate/interface/SoAView.h b/DataFormats/SoATemplate/interface/SoAView.h index 8cf0307c71ad0..f219bd137a0cc 100644 --- a/DataFormats/SoATemplate/interface/SoAView.h +++ b/DataFormats/SoATemplate/interface/SoAView.h @@ -231,6 +231,15 @@ namespace cms::soa { #define _DECLARE_VIEW_MEMBER_LIST(R, DATA, LAYOUT_MEMBER_NAME) \ BOOST_PP_EXPAND(_DECLARE_VIEW_MEMBER_LIST_IMPL LAYOUT_MEMBER_NAME) +/** + * Generator of view member list. + */ +#define _DECLARE_VIEW_OTHER_MEMBER_LIST_IMPL(LAYOUT, MEMBER, NAME) \ + (const_cast_SoAParametersImpl(other.BOOST_PP_CAT(NAME, Parameters_)).tupleOrPointer()) + +#define _DECLARE_VIEW_OTHER_MEMBER_LIST(R, DATA, LAYOUT_MEMBER_NAME) \ + BOOST_PP_EXPAND(_DECLARE_VIEW_OTHER_MEMBER_LIST_IMPL LAYOUT_MEMBER_NAME) + /** * Generator of member initializer for copy constructor. */ @@ -390,7 +399,7 @@ namespace cms::soa { template RestrictQualifier::ParamReturnType \ LOCAL_NAME(size_type _soa_impl_index) { \ if constexpr (rangeChecking == cms::soa::RangeChecking::enabled) { \ - if (_soa_impl_index >= base_type::elements_) \ + if (_soa_impl_index >= base_type::elements_ or _soa_impl_index < 0) \ SOA_THROW_OUT_OF_RANGE("Out of range index in mutable " #LOCAL_NAME "(size_type index)") \ } \ return typename cms::soa::SoAAccessors:: \ @@ -428,7 +437,7 @@ namespace cms::soa { template RestrictQualifier::ParamReturnType \ LOCAL_NAME(size_type _soa_impl_index) const { \ if constexpr (rangeChecking == cms::soa::RangeChecking::enabled) { \ - if (_soa_impl_index >= elements_) \ + if (_soa_impl_index >= elements_ or _soa_impl_index < 0) \ SOA_THROW_OUT_OF_RANGE("Out of range index in const " #LOCAL_NAME "(size_type index)") \ } \ return typename cms::soa::SoAAccessors:: \ @@ -535,6 +544,9 @@ namespace cms::soa { template \ using SoAConstValueWithConf = cms::soa::SoAConstValue; \ \ + template \ + friend struct VIEW; \ + \ /** \ * Helper/friend class allowing SoA introspection. \ */ \ @@ -582,6 +594,23 @@ namespace cms::soa { VIEW(VIEW const&) = default; \ VIEW& operator=(VIEW const&) = default; \ \ + /* Copy constructor for other parameters */ \ + template \ + VIEW(VIEW const& other): base_type{other.elements_, \ + _ITERATE_ON_ALL_COMMA(_DECLARE_VIEW_OTHER_MEMBER_LIST, BOOST_PP_EMPTY(), VALUE_LIST) \ + } {} \ + /* Copy operator for other parameters */ \ + template \ + VIEW& operator=(VIEW const& other) { static_cast(*this) = static_cast(other); } \ + \ /* Movable */ \ VIEW(VIEW &&) = default; \ VIEW& operator=(VIEW &&) = default; \ @@ -620,7 +649,7 @@ namespace cms::soa { SOA_HOST_DEVICE SOA_INLINE \ element operator[](size_type _soa_impl_index) { \ if constexpr (rangeChecking == cms::soa::RangeChecking::enabled) { \ - if (_soa_impl_index >= base_type::elements_) \ + if (_soa_impl_index >= base_type::elements_ or _soa_impl_index < 0) \ SOA_THROW_OUT_OF_RANGE("Out of range index in " #VIEW "::operator[]") \ } \ return element{_soa_impl_index, _ITERATE_ON_ALL_COMMA(_DECLARE_VIEW_ELEMENT_CONSTR_CALL, ~, VALUE_LIST)}; \ @@ -673,6 +702,9 @@ namespace cms::soa { template \ friend struct VIEW; \ \ + template \ + friend struct CONST_VIEW; \ + \ /* For CUDA applications, we align to the 128 bytes of the cache lines. \ * See https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#global-memory-3-0 this is still valid \ * up to compute capability 8.X. \ @@ -739,6 +771,23 @@ namespace cms::soa { CONST_VIEW(CONST_VIEW const&) = default; \ CONST_VIEW& operator=(CONST_VIEW const&) = default; \ \ + /* Copy constructor for other parameters */ \ + template \ + CONST_VIEW(CONST_VIEW const& other): CONST_VIEW{other.elements_, \ + _ITERATE_ON_ALL_COMMA(_DECLARE_VIEW_OTHER_MEMBER_LIST, BOOST_PP_EMPTY(), VALUE_LIST) \ + } {} \ + /* Copy operator for other parameters */ \ + template \ + CONST_VIEW& operator=(CONST_VIEW const& other) { *this = other; } \ + \ /* Movable */ \ CONST_VIEW(CONST_VIEW &&) = default; \ CONST_VIEW& operator=(CONST_VIEW &&) = default; \ @@ -761,7 +810,7 @@ namespace cms::soa { SOA_HOST_DEVICE SOA_INLINE \ const_element operator[](size_type _soa_impl_index) const { \ if constexpr (rangeChecking == cms::soa::RangeChecking::enabled) { \ - if (_soa_impl_index >= elements_) \ + if (_soa_impl_index >= elements_ or _soa_impl_index < 0) \ SOA_THROW_OUT_OF_RANGE("Out of range index in " #CONST_VIEW "::operator[]") \ } \ return const_element{ \ diff --git a/DataFormats/SoATemplate/test/SoALayoutAndView_t.cu b/DataFormats/SoATemplate/test/SoALayoutAndView_t.cu index d8a92a56bf37a..0e30f8301f099 100644 --- a/DataFormats/SoATemplate/test/SoALayoutAndView_t.cu +++ b/DataFormats/SoATemplate/test/SoALayoutAndView_t.cu @@ -35,6 +35,8 @@ GENERATE_SOA_LAYOUT(SoAHostDeviceLayoutTemplate, using SoAHostDeviceLayout = SoAHostDeviceLayoutTemplate<>; using SoAHostDeviceView = SoAHostDeviceLayout::View; +using SoAHostDeviceRangeCheckingView = + SoAHostDeviceLayout::ViewTemplate; using SoAHostDeviceConstView = SoAHostDeviceLayout::ConstView; GENERATE_SOA_LAYOUT(SoADeviceOnlyLayoutTemplate, @@ -126,6 +128,12 @@ int main(void) { cudaCheck(cudaMallocHost(&h_buf, hostDeviceSize)); SoAHostDeviceLayout h_soahdLayout(h_buf, numElements); SoAHostDeviceView h_soahd(h_soahdLayout); + + // Validation of range checking variants initialization + SoAHostDeviceRangeCheckingView h_soahdrc(h_soahdLayout); + [[maybe_unused]] SoAHostDeviceRangeCheckingView h_soahdrc2 = h_soahdLayout; + SoAHostDeviceRangeCheckingView h_soahdrc3{h_soahd}; + [[maybe_unused]] SoAHostDeviceRangeCheckingView h_soahdrc4 = h_soahd; SoAHostDeviceConstView h_soahd_c(h_soahdLayout); // Alocate buffer, stores and views on the device (single, shared buffer). @@ -248,29 +256,50 @@ int main(void) { } } - // Validation of range checking - try { - // Get a view like the default, except for range checking - SoAHostDeviceLayout::ViewTemplate - soa1viewRangeChecking(h_soahdLayout); - // This should throw an exception - [[maybe_unused]] auto si = soa1viewRangeChecking[soa1viewRangeChecking.metadata().size()]; - std::cout << "Fail: expected range-check exception (operator[]) not caught on the host." << std::endl; - assert(false); - } catch (const std::out_of_range&) { - std::cout << "Pass: expected range-check exception (operator[]) successfully caught on the host." << std::endl; + { + // Get a view like the default, except for range checking (direct initialization from layout) + SoAHostDeviceRangeCheckingView soa1viewRangeChecking(h_soahdLayout); + try { + [[maybe_unused]] auto si = soa1viewRangeChecking[soa1viewRangeChecking.metadata().size()]; + std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host (overflow)." + << std::endl; + assert(false); + } catch (const std::out_of_range&) { + } + try { + [[maybe_unused]] auto si = soa1viewRangeChecking[-1]; + std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host (underflow)." + << std::endl; + assert(false); + } catch (const std::out_of_range&) { + } + [[maybe_unused]] auto si = soa1viewRangeChecking[soa1viewRangeChecking.metadata().size() - 1]; + [[maybe_unused]] auto si2 = soa1viewRangeChecking[0]; + std::cout << "Pass: expected range-check exceptions (view-level index access) successfully caught on the host " + "(layout initialization)." + << std::endl; } - try { - // Get a view like the default, except for range checking - SoAHostDeviceLayout::ViewTemplate - soa1viewRangeChecking(h_soahdLayout); - // This should throw an exception - [[maybe_unused]] auto si = soa1viewRangeChecking[soa1viewRangeChecking.metadata().size()]; - std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host." << std::endl; - assert(false); - } catch (const std::out_of_range&) { - std::cout << "Pass: expected range-check exception (view-level index access) successfully caught on the host." + { + // Validation of view initialized range checking view initialization + try { + [[maybe_unused]] auto si = h_soahdrc3[h_soahdrc3.metadata().size()]; + std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host (overflow)." + << std::endl; + assert(false); + } catch (const std::out_of_range&) { + } + try { + [[maybe_unused]] auto si = h_soahdrc3[-1]; + std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host (underflow)." + << std::endl; + assert(false); + } catch (const std::out_of_range&) { + } + [[maybe_unused]] auto si = h_soahdrc3[h_soahdrc3.metadata().size() - 1]; + [[maybe_unused]] auto si2 = h_soahdrc3[0]; + std::cout << "Pass: expected range-check exceptions (view-level index access) successfully caught on the host " + "(view initialization)." << std::endl; } diff --git a/DataFormats/SoATemplate/test/SoALayoutAndView_t.hip.cc b/DataFormats/SoATemplate/test/SoALayoutAndView_t.hip.cc index 1965d0757d358..6a041248f3995 100644 --- a/DataFormats/SoATemplate/test/SoALayoutAndView_t.hip.cc +++ b/DataFormats/SoATemplate/test/SoALayoutAndView_t.hip.cc @@ -37,6 +37,8 @@ GENERATE_SOA_LAYOUT(SoAHostDeviceLayoutTemplate, using SoAHostDeviceLayout = SoAHostDeviceLayoutTemplate<>; using SoAHostDeviceView = SoAHostDeviceLayout::View; +using SoAHostDeviceRangeCheckingView = + SoAHostDeviceLayout::ViewTemplate; using SoAHostDeviceConstView = SoAHostDeviceLayout::ConstView; GENERATE_SOA_LAYOUT(SoADeviceOnlyLayoutTemplate, @@ -128,6 +130,12 @@ int main(void) { hipCheck(hipHostMalloc((void**)&h_buf, hostDeviceSize)); SoAHostDeviceLayout h_soahdLayout(h_buf, numElements); SoAHostDeviceView h_soahd(h_soahdLayout); + + // Validation of range checking variants initialization + SoAHostDeviceRangeCheckingView h_soahdrc(h_soahdLayout); + [[maybe_unused]] SoAHostDeviceRangeCheckingView h_soahdrc2 = h_soahdLayout; + [[maybe_unused]] SoAHostDeviceRangeCheckingView h_soahdrc3{h_soahd}; + [[maybe_unused]] SoAHostDeviceRangeCheckingView h_soahdrc4 = h_soahd; SoAHostDeviceConstView h_soahd_c(h_soahdLayout); // Alocate buffer, stores and views on the device (single, shared buffer). @@ -250,29 +258,50 @@ int main(void) { } } - // Validation of range checking - try { - // Get a view like the default, except for range checking - SoAHostDeviceLayout::ViewTemplate - soa1viewRangeChecking(h_soahdLayout); - // This should throw an exception - [[maybe_unused]] auto si = soa1viewRangeChecking[soa1viewRangeChecking.metadata().size()]; - std::cout << "Fail: expected range-check exception (operator[]) not caught on the host." << std::endl; - assert(false); - } catch (const std::out_of_range&) { - std::cout << "Pass: expected range-check exception (operator[]) successfully caught on the host." << std::endl; + { + // Get a view like the default, except for range checking (direct initialization from layout) + SoAHostDeviceRangeCheckingView soa1viewRangeChecking(h_soahdLayout); + try { + [[maybe_unused]] auto si = soa1viewRangeChecking[soa1viewRangeChecking.metadata().size()]; + std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host (overflow)." + << std::endl; + assert(false); + } catch (const std::out_of_range&) { + } + try { + [[maybe_unused]] auto si = soa1viewRangeChecking[-1]; + std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host (underflow)." + << std::endl; + assert(false); + } catch (const std::out_of_range&) { + } + [[maybe_unused]] auto si = soa1viewRangeChecking[soa1viewRangeChecking.metadata().size() - 1]; + [[maybe_unused]] auto si2 = soa1viewRangeChecking[0]; + std::cout << "Pass: expected range-check exceptions (view-level index access) successfully caught on the host " + "(layout initialization)." + << std::endl; } - try { - // Get a view like the default, except for range checking - SoAHostDeviceLayout::ViewTemplate - soa1viewRangeChecking(h_soahdLayout); - // This should throw an exception - [[maybe_unused]] auto si = soa1viewRangeChecking[soa1viewRangeChecking.metadata().size()]; - std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host." << std::endl; - assert(false); - } catch (const std::out_of_range&) { - std::cout << "Pass: expected range-check exception (view-level index access) successfully caught on the host." + { + // Validation of view initialized range checking view initialization + try { + [[maybe_unused]] auto si = h_soahdrc3[h_soahdrc3.metadata().size()]; + std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host (overflow)." + << std::endl; + assert(false); + } catch (const std::out_of_range&) { + } + try { + [[maybe_unused]] auto si = h_soahdrc3[-1]; + std::cout << "Fail: expected range-check exception (view-level index access) not caught on the host (underflow)." + << std::endl; + assert(false); + } catch (const std::out_of_range&) { + } + [[maybe_unused]] auto si = h_soahdrc3[h_soahdrc3.metadata().size() - 1]; + [[maybe_unused]] auto si2 = h_soahdrc3[0]; + std::cout << "Pass: expected range-check exceptions (view-level index access) successfully caught on the host " + "(view initialization)." << std::endl; }