Cutlass
CUDA Templates for Linear Algebra Subroutines and Solvers
Classes | Public Types | Public Member Functions | Public Attributes | Static Public Attributes | List of all members
cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ > Struct Template Reference

An iterator implementing Tile Store Iterator Concept for storing a tile to memory.

#include <tile_iterator.h>

Inheritance diagram for cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >:
cutlass::TileIteratorBase< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >

Classes

struct  Params
 Parameters. More...
 

Public Types

typedef TileIteratorBase< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ > Base
 Base class. More...
 
typedef Base::Traits Traits
 concept TileTraits More...
 
typedef Base::Scalar Scalar
 Scalar element. More...
 
typedef Base::FragmentElement FragmentElement
 Fragment element. More...
 
typedef Base::Index Index
 Index type. More...
 
typedef Base::LongIndex LongIndex
 Long index type. More...
 
typedef Base::Skew Skew
 Skew quantity. More...
 
typedef Base::Tile Tile
 Tile shape. More...
 
typedef Base::Delta Delta
 Delta. More...
 
typedef Base::Iterations Iterations
 Iterations. More...
 
typedef Base::ThreadOffset ThreadOffset
 ThreadOffset functor. More...
 
typedef Base::FragmentShape FragmentShape
 Fragment type. More...
 
typedef Base::AccessType AccessType
 Memory access type. More...
 
typedef Base::Fragment Fragment
 Fragment definition. More...
 
typedef Base::FragmentIterator FragmentIterator
 Fragment iterator definition. More...
 
typedef Base::FragmentConstIterator FragmentConstIterator
 Fragment const iterator definition. More...
 
typedef Base::PredicateVector PredicateVector
 Default predicate mask type. More...
 
typedef Base::Storage SharedStorage
 Storage object which may be stored to. More...
 
typedef Base::Params BaseParams
 IteratorBase parameters. More...
 
typedef ScalarPointer
 Pointer to underlying type. More...
 
typedef TensorRef< Scalar, 4 > TensorRef
 Tensor reference for the store iterator. More...
 
- Public Types inherited from cutlass::TileIteratorBase< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >
typedef Traits_ Traits
 concept TileTraits More...
 
typedef Scalar_ Scalar
 Scalar element. More...
 
typedef FragmentElement_ FragmentElement
 Fragment element. More...
 
typedef Index_ Index
 Index type. More...
 
typedef long long LongIndex
 Long index. More...
 
typedef Skew_ Skew
 Skew quantity. More...
 
typedef Traits::Tile Tile
 Tile shape. More...
 
typedef Traits::Delta Delta
 Distance along each dimension. More...
 
typedef Traits::ImmediateOffsetStrides ImmediateOffsetStrides
 The strides in each dimension between different loads/stores. More...
 
typedef Traits::Iterations Iterations
 Iterations. More...
 
typedef Traits::ThreadOffset ThreadOffset
 Thread offset. More...
 
typedef Vectorize< FragmentElement, kAccessSize >::Type AccessType
 The elements loaded/store by one instruction. More...
 
typedef Fragment< Scalar, ShapeCount< Tile >::kCount, kFragmentSizeStorage
 The storage. More...
 
typedef Fragment< FragmentElement, ShapeCount< Iterations >::kCount *kAccessSizeFragment
 The fragment. More...
 
typedef FragmentIterator< Fragment, Iterations, AccessTypeFragmentIterator
 The fragment iterator. More...
 
typedef FragmentConstIterator< Fragment, Iterations, AccessTypeFragmentConstIterator
 The fragment const iterator. More...
 
typedef FragmentIterator::FragmentShape FragmentShape
 The shape of the fragment. More...
 
typedef PredicateVector< ShapeCount< Iterations >::kCount > PredicateVector
 Default predicate mask type. More...
 

Public Member Functions

template<typename PredicateIterator >
CUTLASS_HOST_DEVICE void initialize_predicates (PredicateIterator predicate_it, Coord< 3 > const &bounds, Coord< 3 > const &block_offset=make_Coord(0, 0, 0))
 Initializes a predicate vector using a RegularTilePredicateFunctor. More...
 
template<typename PredicateIterator , typename PredicateFunctor >
CUTLASS_HOST_DEVICE void initialize_predicates (PredicateIterator predicate_it, PredicateFunctor const &functor, Coord< 3 > const &block_offset)
 Initializes a predicate vector using an arbitrary predicate functor. More...
 
CUTLASS_HOST_DEVICE TileStoreIterator ()
 Default constructor. More...
 
CUTLASS_HOST_DEVICE TileStoreIterator (Params const &_params, Coord< 3 > const &block_offset=make_Coord(0, 0, 0), ThreadOffset thread_offset_func=ThreadOffset())
 Constructs a tile store iterator. More...
 
CUTLASS_HOST_DEVICE TileStoreIterator (Params const &, Scalar *ptr, ThreadOffset thread_offset_func=ThreadOffset())
 Constructs a tile store iterator. More...
 
CUTLASS_HOST_DEVICE void inc_d ()
 Increment in the D dimension. More...
 
CUTLASS_HOST_DEVICE void inc_h ()
 Increment in the H dimension. More...
 
CUTLASS_HOST_DEVICE void inc_w ()
 Increment in the W dimension. More...
 
CUTLASS_HOST_DEVICE void inc_advance ()
 Increment in the next dimension. More...
 
CUTLASS_HOST_DEVICE void inc_stage ()
 Increment the stage. More...
 
CUTLASS_HOST_DEVICE TileStoreIteratoroperator+= (Coord< 3 > const &offset)
 Adds a vector offset to the iterator. More...
 
CUTLASS_HOST_DEVICE void add_pointer_offset (LongIndex offset)
 Adds a raw offset to the pointer. More...
 
CUTLASS_HOST_DEVICE void store_element (AccessType const &value, int d, int h, int w, int c)
 Stores a single fragment element into memory. More...
 
template<typename Fragment , typename PredicateIterator >
CUTLASS_HOST_DEVICE void store_post_increment (Fragment const &fragment, PredicateIterator pred_it)
 Stores a fragment and advances to the next tile. More...
 
template<typename Fragment >
CUTLASS_HOST_DEVICE void store_post_increment (Fragment const &fragment)
 Stores a fragment and advances to the next tile. More...
 
template<typename Fragment , typename PredicateIterator >
CUTLASS_HOST_DEVICE void store (Fragment const &fragment, PredicateIterator pred_it) const
 Stores a fragment without advancing the iterator. More...
 
template<typename Fragment >
CUTLASS_HOST_DEVICE void store (Fragment const &fragment) const
 Stores a fragment without advancing the iterator. More...
 
CUTLASS_HOST_DEVICE void load_element (AccessType &value, int d, int h, int w, int c) const
 Loads a single fragment element from memory. More...
 
template<typename Fragment , typename PredicateIterator >
CUTLASS_HOST_DEVICE void load_post_increment (Fragment &fragment, PredicateIterator pred_it)
 Loads a fragment and advances the iterator to the next tile. More...
 
template<typename Fragment >
CUTLASS_HOST_DEVICE void load_post_increment (Fragment &fragment)
 Loads a fragment and advances the iterator to the next tile. More...
 
template<typename Fragment , typename PredicateIterator >
CUTLASS_HOST_DEVICE void load (Fragment &fragment, PredicateIterator pred_it) const
 Loads a fragment without advancing the iterator.. More...
 
template<typename Fragment >
CUTLASS_HOST_DEVICE void load (Fragment &fragment) const
 Loads a fragment without advancing the iterator.. More...
 
template<typename Fragment >
CUTLASS_HOST_DEVICE void load (Fragment &fragment, int d)
 Loads a fragment without advancing the iterator.. More...
 
- Public Member Functions inherited from cutlass::TileIteratorBase< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >
CUTLASS_HOST_DEVICE bool valid (int d, int h, int w, int c) const
 Is the iterator valid? More...
 

Public Attributes

Params params
 Parameters structure. More...
 
Coord< 4 > thread_offset
 Offset of an individual lane from the start of the tile. More...
 
int stage
 The stage. More...
 

Static Public Attributes

static IteratorAdvance::Kind const kAdvance = Base::kAdvance
 Specifies in which dimension post-increment accesses advance. More...
 
static FragmentElementType::Kind const kFragmentElementType = Base::kFragmentElementType
 Specifies type of iterator fragment storage (Salar or WmmaMatrix) More...
 
static MemorySpace::Kind const kMemorySpace = Base::kMemorySpace
 Source or destination memory space. More...
 
static int const kAccessSize = Base::kAccessSize
 The number of scalars accessed per load/store. More...
 
- Static Public Attributes inherited from cutlass::TileIteratorBase< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >
static IteratorAdvance::Kind const kAdvance = Advance_
 Specifies dimension in which post-increment accesses advance. More...
 
static FragmentElementType::Kind const kFragmentElementType = FragmentElementType_
 Specifies iterator storage fragment type (Scalar or WmmaMatrix) More...
 
static MemorySpace::Kind const kMemorySpace = MemorySpace
 Source or destination memory space. More...
 
static int const kAccessSize = Traits::kAccessSize
 The number of scalars accessed per load/store. More...
 
static int const kFragmentSize
 The size of storage needed per fragment. More...
 

Additional Inherited Members

- Static Public Member Functions inherited from cutlass::TileIteratorBase< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >
template<typename PredicateIterator , typename PredicateFunctor >
static CUTLASS_HOST_DEVICE void initialize_predicates (PredicateIterator predicate_it, PredicateFunctor const &predicate_func, Coord< 3 > const &offset)
 Initializes a predicate vector. More...
 

Member Typedef Documentation

◆ AccessType

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::AccessType cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::AccessType

◆ Base

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef TileIteratorBase<Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_> cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Base

◆ BaseParams

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Params cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::BaseParams

◆ Delta

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Delta cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Delta

◆ Fragment

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Fragment cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Fragment

◆ FragmentConstIterator

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::FragmentConstIterator cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::FragmentConstIterator

◆ FragmentElement

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::FragmentElement cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::FragmentElement

◆ FragmentIterator

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::FragmentIterator cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::FragmentIterator

◆ FragmentShape

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::FragmentShape cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::FragmentShape

◆ Index

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Index cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Index

◆ Iterations

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Iterations cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Iterations

◆ LongIndex

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::LongIndex cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::LongIndex

◆ Pointer

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Scalar* cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Pointer

◆ PredicateVector

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::PredicateVector cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::PredicateVector

◆ Scalar

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Scalar cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Scalar

◆ SharedStorage

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Storage cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::SharedStorage

◆ Skew

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Skew cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Skew

◆ TensorRef

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef TensorRef<Scalar, 4> cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::TensorRef

◆ ThreadOffset

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::ThreadOffset cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::ThreadOffset

◆ Tile

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Tile cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Tile

◆ Traits

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
typedef Base::Traits cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::Traits

Constructor & Destructor Documentation

◆ TileStoreIterator() [1/3]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::TileStoreIterator ( )
inline

◆ TileStoreIterator() [2/3]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::TileStoreIterator ( Params const &  _params,
Coord< 3 > const &  block_offset = make_Coord(0, 0, 0),
ThreadOffset  thread_offset_func = ThreadOffset() 
)
inline

◆ TileStoreIterator() [3/3]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::TileStoreIterator ( Params const &  ,
Scalar ptr,
ThreadOffset  thread_offset_func = ThreadOffset() 
)
inline

Member Function Documentation

◆ add_pointer_offset()

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::add_pointer_offset ( LongIndex  offset)
inline

◆ inc_advance()

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::inc_advance ( )
inline

◆ inc_d()

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::inc_d ( )
inline

◆ inc_h()

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::inc_h ( )
inline

◆ inc_stage()

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::inc_stage ( )
inline

◆ inc_w()

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::inc_w ( )
inline

◆ initialize_predicates() [1/2]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename PredicateIterator >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::initialize_predicates ( PredicateIterator  predicate_it,
Coord< 3 > const &  bounds,
Coord< 3 > const &  block_offset = make_Coord(0,                                                                                           0,                                                                                           0) 
)
inline

◆ initialize_predicates() [2/2]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename PredicateIterator , typename PredicateFunctor >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::initialize_predicates ( PredicateIterator  predicate_it,
PredicateFunctor const &  functor,
Coord< 3 > const &  block_offset 
)
inline

◆ load() [1/3]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename Fragment , typename PredicateIterator >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::load ( Fragment fragment,
PredicateIterator  pred_it 
) const
inline

◆ load() [2/3]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename Fragment >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::load ( Fragment fragment) const
inline

◆ load() [3/3]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename Fragment >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::load ( Fragment fragment,
int  d 
)
inline

◆ load_element()

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::load_element ( AccessType value,
int  d,
int  h,
int  w,
int  c 
) const
inline

◆ load_post_increment() [1/2]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename Fragment , typename PredicateIterator >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::load_post_increment ( Fragment fragment,
PredicateIterator  pred_it 
)
inline

◆ load_post_increment() [2/2]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename Fragment >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::load_post_increment ( Fragment fragment)
inline

◆ operator+=()

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE TileStoreIterator& cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::operator+= ( Coord< 3 > const &  offset)
inline

◆ store() [1/2]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename Fragment , typename PredicateIterator >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::store ( Fragment const &  fragment,
PredicateIterator  pred_it 
) const
inline

◆ store() [2/2]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename Fragment >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::store ( Fragment const &  fragment) const
inline

◆ store_element()

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::store_element ( AccessType const &  value,
int  d,
int  h,
int  w,
int  c 
)
inline

◆ store_post_increment() [1/2]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename Fragment , typename PredicateIterator >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::store_post_increment ( Fragment const &  fragment,
PredicateIterator  pred_it 
)
inline

◆ store_post_increment() [2/2]

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
template<typename Fragment >
CUTLASS_HOST_DEVICE void cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::store_post_increment ( Fragment const &  fragment)
inline

Member Data Documentation

◆ kAccessSize

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
int const cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::kAccessSize = Base::kAccessSize
static

◆ kAdvance

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
IteratorAdvance::Kind const cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::kAdvance = Base::kAdvance
static

◆ kFragmentElementType

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
FragmentElementType::Kind const cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::kFragmentElementType = Base::kFragmentElementType
static

◆ kMemorySpace

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
MemorySpace::Kind const cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::kMemorySpace = Base::kMemorySpace
static

◆ params

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
Params cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::params

◆ stage

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
int cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::stage

◆ thread_offset

template<typename Traits_ , typename Scalar_ , IteratorAdvance::Kind Advance_ = IteratorAdvance::kH, MemorySpace::Kind MemorySpace = MemorySpace::kGeneric, typename Index_ = int, typename FragmentElement_ = Scalar_, FragmentElementType::Kind FragmentElementType_ = FragmentElementType::kScalar, typename Skew_ = Shape<0, 0, 0, 0>>
Coord<4> cutlass::TileStoreIterator< Traits_, Scalar_, Advance_, MemorySpace, Index_, FragmentElement_, FragmentElementType_, Skew_ >::thread_offset

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