data/detail/Filler.device.t.hh

0001 //---------------------------------*-CUDA-*----------------------------------//
0002 // Copyright 2021-2024 UT-Battelle, LLC, and other Celeritas developers.
0003 // See the top-level COPYRIGHT file for details.
0004 // SPDX-License-Identifier: (Apache-2.0 OR MIT)
0005 //---------------------------------------------------------------------------//
0006 //! \file corecel/data/detail/Filler.device.t.hh
0007 //---------------------------------------------------------------------------//
0008 #pragma once
0009
0010 #include "Filler.hh"
0011
0012 #include <thrust/device_malloc.h>
0013 #include <thrust/device_ptr.h>
0014 #include <thrust/execution_policy.h>
0015 #include <thrust/fill.h>
0016
0017 #include "corecel/DeviceRuntimeApi.hh"
0018
0019 #include "corecel/sys/Thrust.device.hh"
0020
0021 namespace celeritas
0022 {
0023 namespace detail
0024 {
0025 //---------------------------------------------------------------------------//
0026 template<class T>
0027 void Filler<T, MemSpace::device>::operator()(Span<T> data) const
0028 {
0029     if (stream_)
0030     {
0031         thrust::fill_n(thrust_execute_on(stream_),
0032                        thrust::device_pointer_cast<T>(data.data()),
0033                        data.size(),
0034                        value_);
0035     }
0036     else
0037     {
0038         thrust::fill_n(thrust_execution_policy<ThrustExecMode::Sync>(),
0039                        thrust::device_pointer_cast<T>(data.data()),
0040                        data.size(),
0041                        value_);
0042     }
0043     CELER_DEVICE_CHECK_ERROR();
0044 }
0045
0046 //---------------------------------------------------------------------------//
0047 }  // namespace detail
0048 }  // namespace celeritas