All Classes Namespaces Files Functions Variables Typedefs Enumerations Enumerator Friends Groups
Public Types | Public Methods | Public Members | Friends | List of all members
cub::CacheModifiedInputIterator< MODIFIER, ValueType, OffsetT > Class Template Reference

Detailed description

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

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

  • CacheModifiedInputIteratorTis 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.
The code snippet below illustrates the use of CacheModifiedInputIteratorTto 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
// 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
CacheLoadModifierThe cub::CacheLoadModifier to use when accessing data
ValueTypeThe value type of this iterator
OffsetTThe difference type of this iterator (Default: ptrdiff_t)

Public Types

typedef CacheModifiedInputIterator self_type
 My own type.
typedef OffsetT difference_type
 Type to express the result of subtracting one iterator from another.
typedef ValueType value_type
 The type of the element the iterator can point to.
typedef ValueType * pointer
 The type of a pointer to an element the iterator can point to.
typedef ValueType reference
 The type of a reference to an element the iterator can point to.
 The iterator category.

Public Methods

template<typename QualifiedValueType >
__host__ __device__ __forceinline__ CacheModifiedInputIterator (QualifiedValueType *ptr)
 Constructor. More...
__host__ __device__
__forceinline__ self_type 
operator++ (int)
 Postfix increment.
__host__ __device__
__forceinline__ self_type 
operator++ ()
 Prefix increment.
__device__ __forceinline__
operator* () const
template<typename Distance >
__host__ __device__
__forceinline__ self_type 
operator+ (Distance n) const
template<typename Distance >
__host__ __device__
__forceinline__ self_type
operator+= (Distance n)
 Addition assignment.
template<typename Distance >
__host__ __device__
__forceinline__ self_type 
operator- (Distance n) const
template<typename Distance >
__host__ __device__
__forceinline__ self_type
operator-= (Distance n)
 Subtraction assignment.
__host__ __device__
operator- (self_type other) const
template<typename Distance >
__device__ __forceinline__
operator[] (Distance n) const
 Array subscript.
__device__ __forceinline__ pointer operator-> ()
 Structure dereference.
__host__ __device__
__forceinline__ bool 
operator== (const self_type &rhs)
 Equal to.
__host__ __device__
__forceinline__ bool 
operator!= (const self_type &rhs)
 Not equal to.

Public Members

ValueType * ptr
 Wrapped native pointer.


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

Constructor & Destructor Documentation

template<CacheLoadModifier MODIFIER, typename ValueType, typename OffsetT = ptrdiff_t>
template<typename QualifiedValueType >
__host__ __device__ __forceinline__ cub::CacheModifiedInputIterator< MODIFIER, ValueType, OffsetT >::CacheModifiedInputIterator ( QualifiedValueType *  ptr)


ptrNative pointer to wrap

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