cub::CacheModifiedInputIterator

Defined in cub/iterator/cache_modified_input_iterator.cuh

template<CacheLoadModifier MODIFIER, typename ValueType, typename OffsetT = ptrdiff_t> class CacheModifiedInputIterator

A random-access input wrapper for dereferencing array values using a PTX cache load modifier.

Overview

CacheModifiedInputIterator is a random-access input iterator that wraps a native device pointer of type ValueType*. ValueType references are made by reading ValueType values through loads modified by MODIFIER.
Can be used to load any data type from memory using PTX cache load modifiers (e.g., “LOAD_LDG”, “LOAD_CG”, “LOAD_CA”, “LOAD_CS”, “LOAD_CV”, etc.).
Can be constructed, manipulated, and exchanged within and between host and device functions, but can only be dereferenced within device functions.
Compatible with Thrust API v1.7 or newer.

Snippet

The code snippet below illustrates the use of CacheModifiedInputIterator to dereference a device array of double using the “ldg” PTX load modifier (i.e., load values through texture cache).

#include <cub/cub.cuh>   // or equivalently <cub/iterator/cache_modified_input_iterator.cuh>

// Declare, allocate, and initialize a device array
double *d_in;            // e.g., [8.0, 6.0, 7.0, 5.0, 3.0, 0.0, 9.0]

// Create an iterator wrapper
cub::CacheModifiedInputIterator<cub::LOAD_LDG, double> itr(d_in);

// Within device code:
printf("%f\n", itr[0]);  // 8.0
printf("%f\n", itr[1]);  // 6.0
printf("%f\n", itr[6]);  // 9.0

Template Parameters

CacheLoadModifier – The cub::CacheLoadModifier to use when accessing data
ValueType – The value type of this iterator
OffsetT – The difference type of this iterator (Default: ptrdiff_t)

Public Types

using self_type = CacheModifiedInputIterator : My own type.

using difference_type = OffsetT : Type to express the result of subtracting one iterator from another.

using value_type = ValueType : The type of the element the iterator can point to.

using pointer = ValueType*: The type of a pointer to an element the iterator can point to.

using reference = ValueType : The type of a reference to an element the iterator can point to.

using iterator_category = typename THRUST_NS_QUALIFIER::detail::iterator_facade_category<THRUST_NS_QUALIFIER::device_system_tag, THRUST_NS_QUALIFIER::random_access_traversal_tag, value_type, reference>::type

Public Functions

template<typename QualifiedValueType> inline CacheModifiedInputIterator(QualifiedValueType *ptr)

Constructor.

Parameters: ptr – Native pointer to wrap

inline self_type operator++(int): Postfix increment.

inline self_type operator++(): Prefix increment.

inline reference operator*() const: Indirection.

template<typename Distance> inline self_type operator+(Distance n) const: Addition.

template<typename Distance> inline self_type &operator+=(Distance n): Addition assignment.

template<typename Distance> inline self_type operator-(Distance n) const: Subtraction.

template<typename Distance> inline self_type &operator-=(Distance n): Subtraction assignment.

inline difference_type operator-(self_type other) const: Distance.

template<typename Distance> inline reference operator[](Distance n) const: Array subscript.

inline pointer operator->(): Structure dereference.

inline bool operator==(const self_type &rhs) const: Equal to.

inline bool operator!=(const self_type &rhs) const: Not equal to.

Public Members

ValueType *ptr: Wrapped native pointer.

Friends

inline friend ::std::ostream &operator<<(::std::ostream &os, const self_type&): ostream operator