rmm::device_scalar< T > Class Template Reference

Container for a single object of type T in device memory. More...

#include <device_scalar.hpp>

Public Types
using	value_type = typename device_uvector<T>::value_type
	T, the type of the scalar element.
using	size_type = typename device_uvector<T>::size_type
	The type used for the size.
using	reference = typename device_uvector<T>::reference
	value_type&
using	const_reference = typename device_uvector<T>::const_reference
	const value_type&
using	pointer
	The type of the pointer returned by data().
using	const_pointer = typename device_uvector<T>::const_pointer

Public Member Functions
	device_scalar (device_scalar &&) noexcept=default
	Default move constructor.
device_scalar &	operator= (device_scalar &&) noexcept=default
	Default move assignment operator.
	device_scalar (device_scalar const &)=delete
	Copy ctor is deleted as it doesn't allow a stream argument.
device_scalar &	operator= (device_scalar const &)=delete
	Copy assignment is deleted as it doesn't allow a stream argument.
	device_scalar ()=delete
	Default constructor is deleted as it doesn't allow a stream argument.
	device_scalar (cuda::stream_ref stream, cuda::mr::any_resource< cuda::mr::device_accessible > mr=mr::get_current_device_resource_ref())
	Construct a new uninitialized device_scalar.
	device_scalar (value_type const &initial_value, cuda::stream_ref stream, cuda::mr::any_resource< cuda::mr::device_accessible > mr=mr::get_current_device_resource_ref())
	Construct a new device_scalar with an initial value.
	device_scalar (value_type const &&, cuda::stream_ref stream, cuda::mr::any_resource< cuda::mr::device_accessible > mr=mr::get_current_device_resource_ref())=delete
	device_scalar (device_scalar const &other, cuda::stream_ref stream, cuda::mr::any_resource< cuda::mr::device_accessible > mr=mr::get_current_device_resource_ref())
	Construct a new device_scalar by deep copying the contents of another device_scalar, using the specified stream and memory resource.
value_type	value (cuda::stream_ref stream) const
	Copies the value from device to host, synchronizes, and returns the value.
void	set_value_async (value_type const &value, cuda::stream_ref stream)
	Sets the value of the device_scalar to the value of v.
void	set_value_async (value_type const &&, cuda::stream_ref)=delete
void	set_value_to_zero_async (cuda::stream_ref stream)
	Sets the value of the device_scalar to zero on the specified stream.
pointer	data () noexcept
	Returns pointer to object in device memory.
const_pointer	data () const noexcept
	Returns const pointer to object in device memory.
constexpr size_type	size () const noexcept
	The size of the scalar: always 1.
cuda::stream_ref	stream () const noexcept
	Stream associated with the device memory allocation.
void	set_stream (cuda::stream_ref stream) noexcept
	Sets the stream to be used for deallocation.

Detailed Description

template<typename T>
class rmm::device_scalar< T >

Container for a single object of type T in device memory.

T must be trivially copyable.

Template Parameters

T	The object's type

Member Typedef Documentation

const_pointer

template<typename T>

using rmm::device_scalar< T >::const_pointer = typename device_uvector<T>::const_pointer

The type of the iterator returned by data() const

pointer

template<typename T>

using rmm::device_scalar< T >::pointer

Initial value:

 
typename device_uvector<T>::pointer

The type of the pointer returned by data().

Constructor & Destructor Documentation

device_scalar() [1/3]

template<typename T>

rmm::device_scalar< T >::device_scalar ( cuda::stream_ref stream, cuda::mr::any_resource< cuda::mr::device_accessible > mr = mr::get_current_device_resource_ref() )

inlineexplicit

Construct a new uninitialized device_scalar.

Does not synchronize the stream.

Note

This device_scalar is only safe to access in kernels and copies on the specified CUDA stream, or on another stream only if a dependency is enforced (e.g. using cudaStreamWaitEvent()).

Exceptions

rmm::bad_alloc

if allocating the device memory fails.

Parameters

stream	Stream on which to perform asynchronous allocation.
mr	Optional, resource with which to allocate.

device_scalar() [2/3]

template<typename T>

rmm::device_scalar< T >::device_scalar ( value_type const & initial_value, cuda::stream_ref stream, cuda::mr::any_resource< cuda::mr::device_accessible > mr = mr::get_current_device_resource_ref() )

inlineexplicit

Construct a new device_scalar with an initial value.

Does not synchronize the stream.

Note

This device_scalar is only safe to access in kernels and copies on the specified CUDA stream, or on another stream only if a dependency is enforced (e.g. using cudaStreamWaitEvent()).

Exceptions

rmm::bad_alloc	if allocating the device memory for initial_value fails.
rmm::bad_alloc	If the provided memory resource cannot allocate with alignment to satisfy the alignment requirements of the value type.
rmm::cuda_error	if copying initial_value to device memory fails.

Parameters

initial_value	The initial value of the object in device memory.
stream	Optional, stream on which to perform allocation and copy.
mr	Optional, resource with which to allocate.

device_scalar() [3/3]

template<typename T>

rmm::device_scalar< T >::device_scalar ( device_scalar< T > const & other, cuda::stream_ref stream, cuda::mr::any_resource< cuda::mr::device_accessible > mr = mr::get_current_device_resource_ref() )

inline

Construct a new device_scalar by deep copying the contents of another device_scalar, using the specified stream and memory resource.

Exceptions

rmm::bad_alloc	If creating the new allocation fails.
rmm::cuda_error	if copying from other fails.

Parameters

other	The device_scalar whose contents will be copied
stream	The stream to use for the allocation and copy
mr	The resource to use for allocating the new device_scalar

Member Function Documentation

data() [1/2]

template<typename T>

const_pointer rmm::device_scalar< T >::data ( ) const

inlinenodiscardnoexcept

Returns const pointer to object in device memory.

Note

If the returned device pointer is used on a CUDA stream different from the stream specified to the constructor, then appropriate dependencies must be inserted between the streams (e.g. using cudaStreamWaitEvent() or cudaStreamSynchronize()), otherwise there may be a race condition.

Returns

Const pointer to underlying device memory

data() [2/2]

template<typename T>

pointer rmm::device_scalar< T >::data ( )

inlinenodiscardnoexcept

Returns pointer to object in device memory.

Note

If the returned device pointer is used on a CUDA stream different from the stream specified to the constructor, then appropriate dependencies must be inserted between the streams (e.g. using cudaStreamWaitEvent() or cudaStreamSynchronize()), otherwise there may be a race condition.

Returns

Pointer to underlying device memory

operator=()

template<typename T>

device_scalar & rmm::device_scalar< T >::operator= ( device_scalar< T > && )

defaultnoexcept

Default move assignment operator.

Returns

device_scalar& A reference to the assigned-to object

set_stream()

template<typename T>

void rmm::device_scalar< T >::set_stream ( cuda::stream_ref stream )

inlinenoexcept

Sets the stream to be used for deallocation.

Parameters

stream

Stream to be used for deallocation

set_value_async()

template<typename T>

void rmm::device_scalar< T >::set_value_async ( value_type const & value, cuda::stream_ref stream )

inline

Sets the value of the device_scalar to the value of v.

Note

If the stream specified to this function is different from the stream specified to the constructor, then appropriate dependencies must be inserted between the streams (e.g. using cudaStreamWaitEvent() or cudaStreamSynchronize()) before and after calling this function, otherwise there may be a race condition.

This function does not synchronize stream before returning. Therefore, the object referenced by v should not be destroyed or modified until stream has been synchronized. Otherwise, behavior is undefined.

Note

This function incurs a host to device memcpy and should be used carefully.

Example:

rmm::device_scalar<int32_t> s;
 
int v{42};
 
// Copies 42 to device storage on `stream`. Does _not_ synchronize
vec.set_value_async(v, stream);
...
cudaStreamSynchronize(stream);
// Synchronization is required before `v` can be modified
v = 13;

Exceptions

rmm::cuda_error

if copying value to device memory fails.

Parameters

value	The host value which will be copied to device
stream	CUDA stream on which to perform the copy

set_value_to_zero_async()

template<typename T>

void rmm::device_scalar< T >::set_value_to_zero_async ( cuda::stream_ref stream )

inline

Sets the value of the device_scalar to zero on the specified stream.

Note

If the stream specified to this function is different from the stream specified to the constructor, then appropriate dependencies must be inserted between the streams (e.g. using cudaStreamWaitEvent() or cudaStreamSynchronize()) before and after calling this function, otherwise there may be a race condition.

This function does not synchronize stream before returning.

Note

This function incurs a device memset and should be used carefully.

Parameters

stream

CUDA stream on which to perform the copy

size()

template<typename T>

size_type rmm::device_scalar< T >::size ( ) const

inlinenodiscardconstexprnoexcept

The size of the scalar: always 1.

Returns

The size of the scalar: always 1

stream()

template<typename T>

cuda::stream_ref rmm::device_scalar< T >::stream ( ) const

inlinenodiscardnoexcept

Stream associated with the device memory allocation.

Returns

Stream associated with the device memory allocation

value()

template<typename T>

value_type rmm::device_scalar< T >::value ( cuda::stream_ref stream ) const

inlinenodiscard

Copies the value from device to host, synchronizes, and returns the value.

Synchronizes stream after copying the data from device to host.

Note

If the stream specified to this function is different from the stream specified to the constructor, then an appropriate dependency must be inserted between the streams (e.g. using cudaStreamWaitEvent() or cudaStreamSynchronize()) before calling this function, otherwise there may be a race condition.

Exceptions

rmm::cuda_error	If the copy fails.
rmm::cuda_error	If synchronizing stream fails.

Returns

T The value of the scalar.

Parameters

stream

CUDA stream on which to perform the copy and synchronize.

---

The documentation for this class was generated from the following file:

• device_scalar.hpp