Public Types | Public Member Functions | Static Public Member Functions | List of all members
cudf::mutable_column_device_view Class Reference

A mutable, non-owning view of device data as a column of elements that is trivially copyable and usable in CUDA device code. More...

Inheritance diagram for cudf::mutable_column_device_view:
cudf::detail::column_device_view_base

Public Types

using count_it = thrust::counting_iterator< size_type >
 Iterator for navigating this column.
 
template<typename T >
using iterator = thrust::transform_iterator< detail::mutable_value_accessor< T >, count_it >
 

Public Member Functions

 mutable_column_device_view (mutable_column_device_view const &)=default
 
 mutable_column_device_view (mutable_column_device_view &&)=default
 
mutable_column_device_viewoperator= (mutable_column_device_view const &)=default
 
mutable_column_device_viewoperator= (mutable_column_device_view &&)=default
 
 mutable_column_device_view (mutable_column_view column, void *h_ptr, void *d_ptr)
 Creates an instance of this class using the specified host memory pointer (h_ptr) to store child objects and the device memory pointer (d_ptr) as a base for any child object pointers. More...
 
template<typename T = void, CUDF_ENABLE_IF(std::is_same_v< T, void > or is_rep_layout_compatible< T >()) >
T * head () const noexcept
 Returns pointer to the base device memory allocation casted to the specified type. More...
 
template<typename T , CUDF_ENABLE_IF(is_rep_layout_compatible< T >()) >
T * data () const noexcept
 Returns the underlying data casted to the specified type, plus the offset. More...
 
template<typename T , CUDF_ENABLE_IF(is_rep_layout_compatible< T >()) >
T & element (size_type element_index) const noexcept
 Returns reference to element at the specified index. More...
 
bitmask_type * null_mask () const noexcept
 Returns raw pointer to the underlying bitmask allocation. More...
 
template<typename T , CUDF_ENABLE_IF(mutable_column_device_view::has_element_accessor< T >()) >
iterator< T > begin ()
 Return first element (accounting for offset) after underlying data is casted to the specified type. More...
 
template<typename T , CUDF_ENABLE_IF(mutable_column_device_view::has_element_accessor< T >()) >
iterator< T > end ()
 Return one past the last element after underlying data is casted to the specified type. More...
 
mutable_column_device_view child (size_type child_index) const noexcept
 Returns the specified child. More...
 
void set_mask_word (size_type word_index, bitmask_type new_word) const noexcept
 Updates the specified bitmask word in the null_mask() with a new word. More...
 
void destroy ()
 Destroy the mutable_column_device_view object. More...
 
- Public Member Functions inherited from cudf::detail::column_device_view_base
 column_device_view_base (column_device_view_base const &)=default
 
 column_device_view_base (column_device_view_base &&)=default
 
column_device_view_baseoperator= (column_device_view_base const &)=default
 
column_device_view_baseoperator= (column_device_view_base &&)=default
 
template<typename T = void, CUDF_ENABLE_IF(std::is_same_v< T, void > or is_rep_layout_compatible< T >()) >
T const * head () const noexcept
 Returns pointer to the base device memory allocation casted to the specified type. More...
 
template<typename T , CUDF_ENABLE_IF(is_rep_layout_compatible< T >()) >
T const * data () const noexcept
 Returns the underlying data casted to the specified type, plus the offset. More...
 
size_type size () const noexcept
 Returns the number of elements in the column.
 
data_type type () const noexcept
 Returns the element type.
 
bool nullable () const noexcept
 Indicates whether the column can contain null elements, i.e., if it has an allocated bitmask. More...
 
bitmask_type const * null_mask () const noexcept
 Returns raw pointer to the underlying bitmask allocation. More...
 
size_type offset () const noexcept
 Returns the index of the first element relative to the base memory allocation, i.e., what is returned from head<T>().
 
bool is_valid (size_type element_index) const noexcept
 Returns whether the specified element holds a valid value (i.e., not null). More...
 
bool is_valid_nocheck (size_type element_index) const noexcept
 Returns whether the specified element holds a valid value (i.e., not null) More...
 
bool is_null (size_type element_index) const noexcept
 Returns whether the specified element is null. More...
 
bool is_null_nocheck (size_type element_index) const noexcept
 Returns whether the specified element is null. More...
 
bitmask_type get_mask_word (size_type word_index) const noexcept
 Returns the the specified bitmask word from the null_mask(). More...
 

Static Public Member Functions

static std::unique_ptr< mutable_column_device_view, std::function< void(mutable_column_device_view *)> > create (mutable_column_view source_view, rmm::cuda_stream_view stream=rmm::cuda_stream_default)
 Factory to construct a column view that is usable in device memory. More...
 
template<typename T >
static constexpr bool has_element_accessor ()
 For a given T, indicates if mutable_column_device_view::element<T>() has a valid overload.
 
static std::size_t extent (mutable_column_view source_view)
 Return the size in bytes of the amount of memory needed to hold a device view of the specified column and it's children. More...
 

Additional Inherited Members

- Protected Member Functions inherited from cudf::detail::column_device_view_base
 column_device_view_base (data_type type, size_type size, void const *data, bitmask_type const *null_mask, size_type offset)
 
- Protected Attributes inherited from cudf::detail::column_device_view_base
data_type _type {type_id::EMPTY}
 Element type.
 
cudf::size_type _size {}
 Number of elements.
 
void const * _data {}
 Pointer to device memory containing elements.
 
bitmask_type const * _null_mask {}
 
size_type _offset {}
 

Detailed Description

A mutable, non-owning view of device data as a column of elements that is trivially copyable and usable in CUDA device code.

Definition at line 794 of file column_device_view.cuh.

Constructor & Destructor Documentation

◆ mutable_column_device_view()

cudf::mutable_column_device_view::mutable_column_device_view ( mutable_column_view  column,
void *  h_ptr,
void *  d_ptr 
)

Creates an instance of this class using the specified host memory pointer (h_ptr) to store child objects and the device memory pointer (d_ptr) as a base for any child object pointers.

Parameters
columnColumn view from which to create this instance.
h_ptrHost memory pointer on which to place any child data.
d_ptrDevice memory pointer on which to base any child pointers.

Member Function Documentation

◆ begin()

template<typename T , CUDF_ENABLE_IF(mutable_column_device_view::has_element_accessor< T >()) >
iterator<T> cudf::mutable_column_device_view::begin ( )
inline

Return first element (accounting for offset) after underlying data is casted to the specified type.

This function does not participate in overload resolution if mutable_column_device_view::has_element_accessor<T>() is false.

Template Parameters
TThe desired type
Returns
T* Pointer to the first element after casting

Definition at line 937 of file column_device_view.cuh.

◆ child()

mutable_column_device_view cudf::mutable_column_device_view::child ( size_type  child_index) const
inlinenoexcept

Returns the specified child.

Parameters
child_indexThe index of the desired child
Returns
column_view The requested child column_view

Definition at line 964 of file column_device_view.cuh.

◆ create()

static std::unique_ptr<mutable_column_device_view, std::function<void(mutable_column_device_view*)> > cudf::mutable_column_device_view::create ( mutable_column_view  source_view,
rmm::cuda_stream_view  stream = rmm::cuda_stream_default 
)
static

Factory to construct a column view that is usable in device memory.

Allocates and copies views of source_view's children to device memory to make them accessible in device code.

If source_view.num_children() == 0, then no device memory is allocated.

Returns a std::unique_ptr<mutable_column_device_view> with a custom deleter to free the device memory allocated for the children.

A mutable_column_device_view should be passed by value into GPU kernels.

Parameters
source_viewThe column_view to make usable in device code
streamCUDA stream used for device memory operations for children columns.
Returns
A unique_ptr to a mutable_column_device_view that makes the data from source_view available in device memory.

◆ data()

template<typename T , CUDF_ENABLE_IF(is_rep_layout_compatible< T >()) >
T* cudf::mutable_column_device_view::data ( ) const
inlinenoexcept

Returns the underlying data casted to the specified type, plus the offset.

This function does not participate in overload resolution if is_rep_layout_compatible<T> is false.

Note
If offset() == 0, then head<T>() == data<T>()
Template Parameters
TThe type to cast to
Returns
T* Typed pointer to underlying data, including the offset

Definition at line 872 of file column_device_view.cuh.

◆ destroy()

void cudf::mutable_column_device_view::destroy ( )

Destroy the mutable_column_device_view object.

Note
Does not free the column data, simply frees the device memory allocated to hold the child views.

◆ element()

template<typename T , CUDF_ENABLE_IF(is_rep_layout_compatible< T >()) >
T& cudf::mutable_column_device_view::element ( size_type  element_index) const
inlinenoexcept

Returns reference to element at the specified index.

This function accounts for the offset.

This function does not participate in overload resolution if is_rep_layout_compatible<T> is false. Specializations of this function may exist for types T where is_rep_layout_compatible<T> is false.

Template Parameters
TThe element type
Parameters
element_indexPosition of the desired element

Definition at line 891 of file column_device_view.cuh.

◆ end()

template<typename T , CUDF_ENABLE_IF(mutable_column_device_view::has_element_accessor< T >()) >
iterator<T> cudf::mutable_column_device_view::end ( )
inline

Return one past the last element after underlying data is casted to the specified type.

This function does not participate in overload resolution if mutable_column_device_view::has_element_accessor<T>() is false.

Template Parameters
TThe desired type
Returns
T const* Pointer to one past the last element after casting

Definition at line 953 of file column_device_view.cuh.

◆ extent()

static std::size_t cudf::mutable_column_device_view::extent ( mutable_column_view  source_view)
static

Return the size in bytes of the amount of memory needed to hold a device view of the specified column and it's children.

Parameters
source_viewThe column_view to use for this calculation.

◆ head()

template<typename T = void, CUDF_ENABLE_IF(std::is_same_v< T, void > or is_rep_layout_compatible< T >()) >
T* cudf::mutable_column_device_view::head ( ) const
inlinenoexcept

Returns pointer to the base device memory allocation casted to the specified type.

This function will only participate in overload resolution if is_rep_layout_compatible<T>() or std::is_same_v<T,void> are true.

Note
If offset() == 0, then head<T>() == data<T>()
It should be rare to need to access the head<T>() allocation of a column, and instead, accessing the elements should be done via data<T>().
Template Parameters
Thetype to cast to
Returns
T* Typed pointer to underlying data

Definition at line 854 of file column_device_view.cuh.

◆ null_mask()

bitmask_type* cudf::mutable_column_device_view::null_mask ( ) const
inlinenoexcept

Returns raw pointer to the underlying bitmask allocation.

Note
This function does not account for the offset().
If null_count() == 0, this may return nullptr.

Definition at line 914 of file column_device_view.cuh.

◆ set_mask_word()

void cudf::mutable_column_device_view::set_mask_word ( size_type  word_index,
bitmask_type  new_word 
) const
inlinenoexcept

Updates the specified bitmask word in the null_mask() with a new word.

Note
It is undefined behavior to call this function if nullable() == false.
Parameters
word_indexThe index of the word to update
new_wordThe new bitmask word

Definition at line 1019 of file column_device_view.cuh.


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