Cutlass
CUDA Templates for Linear Algebra Subroutines and Solvers
Classes | Public Types | Public Member Functions | Static Public Member Functions | Public Attributes | Static Public Attributes | List of all members
cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ > Struct Template Reference

#include <gemm_global_stream.h>

Classes

struct  Params
 The params. More...
 
struct  SharedStorage
 

Public Types

typedef LoadIterator_ LoadIterator
 The load iterator. More...
 
typedef Transformer_ Transformer
 The transformer. More...
 
typedef StoreIterator_ StoreIterator
 The store iterator to write to shared memory. More...
 
typedef LoadIterator::Fragment FetchedFragment
 The fragment that is copied from shared memory. More...
 
typedef Transformer::OutputFragment TransformedFragment
 The fragment that is obtained after the transformation by the transformer. More...
 
typedef TransformedFragment Fragment
 Make sure the fragments match. More...
 
typedef LoadIterator::Scalar Scalar
 The scalar type of the iterator. More...
 
typedef LoadIterator::Pointer Pointer
 The pointer. More...
 
typedef LoadIterator::Index Index
 The index. More...
 
typedef LoadIterator::LongIndex LongIndex
 The index. More...
 
typedef LoadIterator::Tile Tile
 The tile. More...
 
typedef TileAllocation< typename StoreIterator::Scalar, typename StoreIterator::Tile > ThreadblockTileStorage
 Shared memory allocation for the tile. More...
 
typedef ThreadblockTileStorage::TensorRef ThreadblockTileRef
 Tensor reference to threadblock tile. More...
 

Public Member Functions

CUTLASS_DEVICE GlobalLoadStream (Params const &_params, SharedStorage &shared_storage, ThreadblockTileRef const &threadblock_tile_ref, Coord< 3 > const bounds, Coord< 3 > const &_threadblock_offset)
 Ctor. More...
 
CUTLASS_DEVICE void copy ()
 Load the data from shared memory to the fetch fragment. More...
 
CUTLASS_DEVICE void commit ()
 Commit the data. More...
 
CUTLASS_DEVICE void residue (Index k, bool skip_clear=false)
 Execute the residue code. More...
 
CUTLASS_DEVICE void move_to_residue (Index k, Index kTileK)
 Move to the residue portion. More...
 
CUTLASS_DEVICE void rollback (void)
 Rollback to the beginning of the first tile. More...
 
CUTLASS_DEVICE GlobalLoadStreamoperator+= (Coord< 3 > const &offset)
 Adds a Coord<3> to the underlying global load iterator. More...
 
CUTLASS_DEVICE GlobalLoadStreamadd_batch_offset (int batch_id)
 Adds an offset based on batch stride. More...
 

Static Public Member Functions

static CUTLASS_HOST_DEVICE Coord< 3 > project_coordinate (Coord< 3 > const &coord, Index d_offset=0)
 Maps a coordinate in the GEMM's (K, N, M) coordinate system to global memory. More...
 

Public Attributes

Params params
 Parameters. More...
 
Coord< 3 > threadblock_offset
 Threadblock offset. More...
 
Coord< 3 > multiplicand_bounds
 Multiplicand bounds. More...
 
LoadIterator load_iterator
 The iterator. More...
 
FetchedFragment fetched_fragment
 The fragment to fetch from shared memory. More...
 
Transformer transformer
 The transformer. More...
 
TransformedFragment transformed_fragment
 The fragment to convert the data after it has been fetched from shared memory. More...
 
StoreIterator store_iterator
 The store iterator. More...
 

Static Public Attributes

static GemmOperand::Kind const kOperand = Operand
 Indicates the type of GEMM operand. More...
 
static MatrixLayout::Kind const kLayout = LoadIterator::kLayout
 Make sure the transformed fragment is the same as the store fragment. More...
 

Member Typedef Documentation

◆ FetchedFragment

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef LoadIterator::Fragment cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::FetchedFragment

◆ Fragment

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef TransformedFragment cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::Fragment

The output fragment.

◆ Index

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef LoadIterator::Index cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::Index

◆ LoadIterator

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef LoadIterator_ cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::LoadIterator

◆ LongIndex

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef LoadIterator::LongIndex cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::LongIndex

◆ Pointer

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef LoadIterator::Pointer cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::Pointer

◆ Scalar

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef LoadIterator::Scalar cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::Scalar

◆ StoreIterator

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef StoreIterator_ cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::StoreIterator

◆ ThreadblockTileRef

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef ThreadblockTileStorage::TensorRef cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::ThreadblockTileRef

◆ ThreadblockTileStorage

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef TileAllocation<typename StoreIterator::Scalar, typename StoreIterator::Tile> cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::ThreadblockTileStorage

◆ Tile

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef LoadIterator::Tile cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::Tile

◆ TransformedFragment

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef Transformer::OutputFragment cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::TransformedFragment

◆ Transformer

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
typedef Transformer_ cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::Transformer

Constructor & Destructor Documentation

◆ GlobalLoadStream()

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
CUTLASS_DEVICE cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::GlobalLoadStream ( Params const &  _params,
SharedStorage shared_storage,
ThreadblockTileRef const &  threadblock_tile_ref,
Coord< 3 > const  bounds,
Coord< 3 > const &  _threadblock_offset 
)
inline

Member Function Documentation

◆ add_batch_offset()

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
CUTLASS_DEVICE GlobalLoadStream& cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::add_batch_offset ( int  batch_id)
inline

◆ commit()

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
CUTLASS_DEVICE void cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::commit ( )
inline

◆ copy()

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
CUTLASS_DEVICE void cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::copy ( )
inline

◆ move_to_residue()

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
CUTLASS_DEVICE void cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::move_to_residue ( Index  k,
Index  kTileK 
)
inline

◆ operator+=()

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
CUTLASS_DEVICE GlobalLoadStream& cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::operator+= ( Coord< 3 > const &  offset)
inline

◆ project_coordinate()

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
static CUTLASS_HOST_DEVICE Coord<3> cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::project_coordinate ( Coord< 3 > const &  coord,
Index  d_offset = 0 
)
inlinestatic

◆ residue()

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
CUTLASS_DEVICE void cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::residue ( Index  k,
bool  skip_clear = false 
)
inline

◆ rollback()

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
CUTLASS_DEVICE void cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::rollback ( void  )
inline

Member Data Documentation

◆ fetched_fragment

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
FetchedFragment cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::fetched_fragment

◆ kLayout

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
MatrixLayout::Kind const cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::kLayout = LoadIterator::kLayout
static

The layout.

◆ kOperand

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
GemmOperand::Kind const cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::kOperand = Operand
static

◆ load_iterator

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
LoadIterator cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::load_iterator

◆ multiplicand_bounds

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
Coord<3> cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::multiplicand_bounds

◆ params

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
Params cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::params

◆ store_iterator

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
StoreIterator cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::store_iterator

◆ threadblock_offset

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
Coord<3> cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::threadblock_offset

◆ transformed_fragment

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
TransformedFragment cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::transformed_fragment

◆ transformer

template<GemmOperand::Kind Operand, typename LoadIterator_ , typename StoreIterator_ , typename Transformer_ >
Transformer cutlass::gemm::GlobalLoadStream< Operand, LoadIterator_, StoreIterator_, Transformer_ >::transformer

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