Skip to content

Updated SoA View accessors from raw pointers to span #48377

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
cmsbuild merged 1 commit into from
Oct 1, 2025
Merged

Updated SoA View accessors from raw pointers to span #48377

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
6 changes: 3 additions & 3 deletions CUDADataFormats/TrackingRecHit/interface/TrackingRecHitSoADevice.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -49,9 +49,9 @@ class TrackingRecHitSoADevice : public cms::cuda::PortableDeviceCollection<Track
auto ret = cms::cuda::make_host_unique<float[]>(4 * nHits(), stream);
size_t rowSize = sizeof(float) * nHits();

size_t srcPitch = ptrdiff_t(view().yLocal()) - ptrdiff_t(view().xLocal());
cudaCheck(
cudaMemcpy2DAsync(ret.get(), rowSize, view().xLocal(), srcPitch, rowSize, 4, cudaMemcpyDeviceToHost, stream));
size_t srcPitch = ptrdiff_t(view().yLocal().data()) - ptrdiff_t(view().xLocal().data());
cudaCheck(cudaMemcpy2DAsync(
ret.get(), rowSize, view().xLocal().data(), srcPitch, rowSize, 4, cudaMemcpyDeviceToHost, stream));

return ret;
} //move to utilities
Expand Down
4 changes: 2 additions & 2 deletions CUDADataFormats/TrackingRecHit/test/TrackingRecHitSoA_test.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -45,11 +45,11 @@ namespace testTrackingRecHitSoA {
cudaCheck(cudaDeviceSynchronize());
cms::cuda::fillManyFromVector(&(hits.view().phiBinner()),
10,
hits.view().iphi(),
hits.view().iphi().data(),
hits.view().hitsLayerStart().data(),
2000,
256,
hits.view().phiBinnerStorage(),
hits.view().phiBinnerStorage().data(),
stream);
cudaCheck(cudaDeviceSynchronize());
show<TrackerTraits><<<10, 1000, 0, stream>>>(hits.view());
Expand Down
4 changes: 2 additions & 2 deletions DQM/SiPixelHeterogeneous/plugins/SiPixelCompareTrackSoA.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -157,8 +157,8 @@ void SiPixelCompareTrackSoA<T>::analyze(const edm::Event& iEvent, const edm::Eve
auto const& tsoaGPU = *tsoaHandleGPU;
auto maxTracksCPU = tsoaCPU.view().metadata().size(); //this should be same for both?
auto maxTracksGPU = tsoaGPU.view().metadata().size(); //this should be same for both?
auto const* qualityCPU = tsoaCPU.view().quality();
auto const* qualityGPU = tsoaGPU.view().quality();
auto const qualityCPU = tsoaCPU.view().quality();
auto const qualityGPU = tsoaGPU.view().quality();
int32_t nTracksCPU = 0;
int32_t nTracksGPU = 0;
int32_t nLooseAndAboveTracksCPU = 0;
Expand Down
4 changes: 2 additions & 2 deletions DQM/SiPixelHeterogeneous/plugins/SiPixelCompareTracks.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -166,8 +166,8 @@ void SiPixelCompareTracks::analyzeSeparate(U tokenRef, V tokenTar, const edm::Ev
auto maxTracksRef = tsoaRef.view().metadata().size(); //this should be same for both?
auto maxTracksTar = tsoaTar.view().metadata().size(); //this should be same for both?

auto const* qualityRef = tsoaRef.view().quality();
auto const* qualityTar = tsoaTar.view().quality();
auto const qualityRef = tsoaRef.view().quality();
auto const qualityTar = tsoaTar.view().quality();

int32_t nTracksRef = 0;
int32_t nTracksTar = 0;
Expand Down
2 changes: 1 addition & 1 deletion DQM/SiPixelHeterogeneous/plugins/SiPixelMonitorTrackSoA.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -86,7 +86,7 @@ void SiPixelMonitorTrackSoA<T>::analyze(const edm::Event& iEvent, const edm::Eve
using helper = TracksUtilities<T>;
auto const& tsoa = *tsoaHandle.product();
auto maxTracks = tsoa.view().metadata().size();
auto const* quality = tsoa.view().quality();
auto const quality = tsoa.view().quality();
int32_t nTracks = 0;
int32_t nLooseAndAboveTracks = 0;

Expand Down
2 changes: 1 addition & 1 deletion DQM/SiPixelHeterogeneous/plugins/SiPixelMonitorTrackSoAAlpaka.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -84,7 +84,7 @@ void SiPixelMonitorTrackSoAAlpaka::analyze(const edm::Event& iEvent, const edm::

auto const& tsoa = *tsoaHandle.product();
auto maxTracks = tsoa.view().metadata().size();
auto const* quality = tsoa.view().quality();
auto const quality = tsoa.view().quality();
int32_t nTracks = 0;
int32_t nLooseAndAboveTracks = 0;

Expand Down
16 changes: 10 additions & 6 deletions DataFormats/SiPixelDigi/interface/SiPixelDigisSoA.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -3,6 +3,7 @@

#include <cstddef>
#include <cstdint>
#include <span>
#include <vector>

namespace legacy {
Expand All @@ -17,12 +18,15 @@ namespace legacy {
class SiPixelDigisSoA {
public:
SiPixelDigisSoA() = default;
explicit SiPixelDigisSoA(
size_t nDigis, const uint32_t* pdigi, const uint32_t* rawIdArr, const uint16_t* adc, const int32_t* clus)
: pdigi_(pdigi, pdigi + nDigis),
rawIdArr_(rawIdArr, rawIdArr + nDigis),
adc_(adc, adc + nDigis),
clus_(clus, clus + nDigis) {}
explicit SiPixelDigisSoA(size_t nDigis,
std::span<const uint32_t> pdigi,
std::span<const uint32_t> rawIdArr,
std::span<const uint16_t> adc,
std::span<const int32_t> clus)
: pdigi_(pdigi.data(), pdigi.data() + nDigis),
rawIdArr_(rawIdArr.data(), rawIdArr.data() + nDigis),
adc_(adc.data(), adc.data() + nDigis),
clus_(clus.data(), clus.data() + nDigis) {}

~SiPixelDigisSoA() = default;

Expand Down
1 change: 0 additions & 1 deletion DataFormats/SiPixelDigiSoA/interface/SiPixelDigiErrorsDevice.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -26,7 +26,6 @@ class SiPixelDigiErrorsDevice : public PortableDeviceCollection<SiPixelDigiError
explicit SiPixelDigiErrorsDevice(size_t maxFedWords, TDev const& device)
: PortableDeviceCollection<SiPixelDigiErrorsSoA, TDev>(maxFedWords, device) {}

auto& error_data() const { return (*this->view().pixelErrors()); }
auto maxFedWords() const { return maxFedWords_; }

private:
Expand Down
3 changes: 0 additions & 3 deletions DataFormats/SiPixelDigiSoA/interface/SiPixelDigiErrorsHost.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,9 +22,6 @@ class SiPixelDigiErrorsHost : public PortableHostCollection<SiPixelDigiErrorsSoA

int maxFedWords() const { return maxFedWords_; }

auto& error_data() { return (*view().pixelErrors()); }
auto const& error_data() const { return (*view().pixelErrors()); }

Comment on lines -25 to -27
Copy link
Contributor

@fwyzard fwyzard Sep 29, 2025

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

For consistency, can you remove error_data() also from DataFormats/SiPixelDigiSoA/interface/SiPixelDigiErrorsDevice.h ?

private:
int maxFedWords_ = 0;
};
Expand Down
46 changes: 30 additions & 16 deletions DataFormats/SoATemplate/interface/SoACommon.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -677,30 +677,36 @@ namespace cms::soa {
// Column
template <typename T, byte_size_type alignment, bool restrictQualify>
struct SoAColumnAccessorsImpl<T, SoAColumnType::column, SoAAccessType::mutableAccess, alignment, restrictQualify> {
SOA_HOST_DEVICE SOA_INLINE SoAColumnAccessorsImpl(const SoAParametersImpl<SoAColumnType::column, T>& params)
: params_(params) {}
SOA_HOST_DEVICE SOA_INLINE T* operator()() { return params_.addr_; }
SOA_HOST_DEVICE SOA_INLINE SoAColumnAccessorsImpl(const SoAParametersImpl<SoAColumnType::column, T>& params,
size_type size)
: params_(params), size_(size) {}
SOA_HOST_DEVICE SOA_INLINE std::span<T> operator()() { return std::span<T>(params_.addr_, size_); }

using NoParamReturnType = T*;
using NoParamReturnType = std::span<T>;
using ParamReturnType = T&;
SOA_HOST_DEVICE SOA_INLINE T& operator()(size_type index) { return params_.addr_[index]; }

private:
SoAParametersImpl<SoAColumnType::column, T> params_;
size_type size_ = 0;
};

// Const column
template <typename T, byte_size_type alignment, bool restrictQualify>
struct SoAColumnAccessorsImpl<T, SoAColumnType::column, SoAAccessType::constAccess, alignment, restrictQualify> {
SOA_HOST_DEVICE SOA_INLINE SoAColumnAccessorsImpl(const SoAConstParametersImpl<SoAColumnType::column, T>& params)
: params_(params) {}
SOA_HOST_DEVICE SOA_INLINE const T* operator()() const { return params_.addr_; }
using NoParamReturnType = const T*;
SOA_HOST_DEVICE SOA_INLINE SoAColumnAccessorsImpl(const SoAConstParametersImpl<SoAColumnType::column, T>& params,
size_type size)
: params_(params), size_(size) {}
SOA_HOST_DEVICE SOA_INLINE std::span<const T> operator()() const {
return std::span<const T>(params_.addr_, size_);
}
using NoParamReturnType = std::span<const T>;
using ParamReturnType = const T&;
SOA_HOST_DEVICE SOA_INLINE T const& operator()(size_type index) const { return params_.addr_[index]; }

private:
SoAConstParametersImpl<SoAColumnType::column, T> params_;
const size_type size_ = 0;
};

// Scalar
Expand Down Expand Up @@ -738,33 +744,41 @@ namespace cms::soa {
// Eigen-type
template <typename T, byte_size_type alignment, bool restrictQualify>
struct SoAColumnAccessorsImpl<T, SoAColumnType::eigen, SoAAccessType::mutableAccess, alignment, restrictQualify> {
SOA_HOST_DEVICE SOA_INLINE SoAColumnAccessorsImpl(const SoAParametersImpl<SoAColumnType::eigen, T>& params)
: params_(params) {}
SOA_HOST_DEVICE SOA_INLINE typename T::Scalar* operator()() { return params_.addr_; }
using NoParamReturnType = typename T::Scalar*;
SOA_HOST_DEVICE SOA_INLINE SoAColumnAccessorsImpl(const SoAParametersImpl<SoAColumnType::eigen, T>& params,
size_type size)
: params_(params), size_(size) {}
SOA_HOST_DEVICE SOA_INLINE std::span<typename T::Scalar> operator()() {
return std::span<typename T::Scalar>(params_.addr_, size_);
}
using NoParamReturnType = std::span<typename T::Scalar>;
using ParamReturnType = typename SoAValue<SoAColumnType::eigen, T, alignment, restrictQualify>::MapType;
SOA_HOST_DEVICE SOA_INLINE ParamReturnType operator()(size_type index) {
return SoAValue<SoAColumnType::eigen, T, alignment, restrictQualify>(index, params_)();
}

private:
SoAParametersImpl<SoAColumnType::eigen, T> params_;
size_type size_ = 0;
};

// Const Eigen-type
template <typename T, byte_size_type alignment, bool restrictQualify>
struct SoAColumnAccessorsImpl<T, SoAColumnType::eigen, SoAAccessType::constAccess, alignment, restrictQualify> {
SOA_HOST_DEVICE SOA_INLINE SoAColumnAccessorsImpl(const SoAConstParametersImpl<SoAColumnType::eigen, T>& params)
: params_(params) {}
SOA_HOST_DEVICE SOA_INLINE typename T::Scalar const* operator()() const { return params_.addr_; }
using NoParamReturnType = typename T::Scalar const*;
SOA_HOST_DEVICE SOA_INLINE SoAColumnAccessorsImpl(const SoAConstParametersImpl<SoAColumnType::eigen, T>& params,
size_type size)
: params_(params), size_(size) {}
SOA_HOST_DEVICE SOA_INLINE std::span<typename T::Scalar const> operator()() const {
return std::span<typename T::Scalar const>(params_.addr_, size_);
}
using NoParamReturnType = std::span<typename T::Scalar const>;
using ParamReturnType = typename SoAValue<SoAColumnType::eigen, T, alignment, restrictQualify>::CMapType;
SOA_HOST_DEVICE SOA_INLINE ParamReturnType operator()(size_type index) const {
return SoAConstValue<SoAColumnType::eigen, T, alignment, restrictQualify>(index, params_)();
}

private:
SoAConstParametersImpl<SoAColumnType::eigen, T> params_;
const size_type size_ = 0;
};

/* A helper template stager to avoid commas inside macros */
Expand Down
Loading