Cutlass CUDA Templates for Linear Algebra Subroutines and Solvers

Classes | Public Types | Static Public Member Functions | Static Public Attributes | List of all members
cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ > Struct Template Reference

`#include <gemm_traits.h>`

## Classes
struct | MainLoopSharedStorage
struct | Params
| Parameters object constructable on the host. More...
union | SharedStorage
| The storage in shared memory. More...

## Public Types
typedef GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ > | This_
| This traits. More...
typedef cutlass::gemm::Gemm< This_ > | KernelClass
| The struct that consumes this Traits. More...
typedef GemmConfig_ | GemmConfig
| The configuration. More...
typedef GemmConfig::OutputTile | OutputTile
| The output tile. More...
typedef GlobalLoadStreamA_ | GlobalLoadStreamA
| The stream to load A from global memory to shared memory. More...
typedef GlobalLoadStreamA_::Scalar | ScalarA
| The scalar for A. More...
typedef GlobalLoadStreamB_ | GlobalLoadStreamB
| The stream to load B from global memory to shared memory. More...
typedef GlobalLoadStreamB_::Scalar | ScalarB
| The scalar for B. More...
typedef SharedLoadStreamA_ | SharedLoadStreamA
| The iterator for A to load from shared memory. More...
typedef SharedLoadStreamB_ | SharedLoadStreamB
| The iterator for B to load from shared memory. More...
typedef GemmConfig::MultiplyAdd | MultiplyAdd
| The multiply-add functor. More...
typedef Epilogue_ | Epilogue
| The epilogue. More...
typedef Epilogue::ScalarC | ScalarC
| The scalars in the epilogue. More...
typedef Epilogue::ScalarD | ScalarD
typedef BlockSwizzle_ | BlockSwizzle
| The block swizzle to reorganize the grid. More...
typedef Index_ | Index
| The index. More...
typedef ClearAccumulators_ | ClearAccumulators
| Clear the accumulators. More...
typedef GlobalLoadStreamPair< GlobalLoadStreamA, GlobalLoadStreamB, GemmConfig::kResidueInProlog > | GlobalLoadStream
| Assemble the global load streams for A/B. More...
typedef GlobalLoadStream::ThreadblockTileStorage | ThreadblockTileStorage
| Memory needed to store the threadblock-scoped GEMM tile. More...
typedef SharedStreamPair< SharedLoadStreamA, SharedLoadStreamB > | SharedStream
| Assemble the shared load streams for A/B. More...

## Static Public Member Functions
static CUTLASS_DEVICE void | shared_load_fence (bool in_loop)
| The memory fence for shared loads. More...
static CUTLASS_DEVICE void | shared_store_fence (bool in_loop)
| The memory fence for shared stores. More...

## Static Public Attributes
static MatrixLayout::Kind const | kLayoutA = GlobalLoadStreamA::kLayout
| The layout of A. More...
static MatrixLayout::Kind const | kLayoutB = GlobalLoadStreamB::kLayout
| The layout of B. More...

## Member Typedef Documentation

## ◆ BlockSwizzle

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef BlockSwizzle_ cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::BlockSwizzle

## ◆ ClearAccumulators

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef ClearAccumulators_ cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::ClearAccumulators

## ◆ Epilogue

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef Epilogue_ cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::Epilogue

## ◆ GemmConfig

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GemmConfig_ cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::GemmConfig

## ◆ GlobalLoadStream

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GlobalLoadStreamPair<GlobalLoadStreamA, GlobalLoadStreamB, GemmConfig::kResidueInProlog> cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::GlobalLoadStream

## ◆ GlobalLoadStreamA

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GlobalLoadStreamA_ cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::GlobalLoadStreamA

## ◆ GlobalLoadStreamB

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GlobalLoadStreamB_ cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::GlobalLoadStreamB

## ◆ Index

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef Index_ cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::Index

## ◆ KernelClass

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef cutlass::gemm::Gemm<This_> cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::KernelClass

## ◆ MultiplyAdd

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GemmConfig::MultiplyAdd cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::MultiplyAdd

## ◆ OutputTile

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GemmConfig::OutputTile cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::OutputTile

## ◆ ScalarA

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GlobalLoadStreamA_::Scalar cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::ScalarA

## ◆ ScalarB

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GlobalLoadStreamB_::Scalar cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::ScalarB

## ◆ ScalarC

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef Epilogue::ScalarC cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::ScalarC

## ◆ ScalarD

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef Epilogue::ScalarD cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::ScalarD

## ◆ SharedLoadStreamA

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef SharedLoadStreamA_ cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::SharedLoadStreamA

## ◆ SharedLoadStreamB

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef SharedLoadStreamB_ cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::SharedLoadStreamB

## ◆ SharedStream

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef SharedStreamPair<SharedLoadStreamA, SharedLoadStreamB> cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::SharedStream

## ◆ This_

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GemmTraits<GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_> cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::This_

## ◆ ThreadblockTileStorage

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

typedef GlobalLoadStream::ThreadblockTileStorage cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::ThreadblockTileStorage

## Member Function Documentation

## ◆ shared_load_fence()

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

static CUTLASS_DEVICE void cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::shared_load_fence | ( | bool | in_loop | ) |
| inline static

## ◆ shared_store_fence()

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

static CUTLASS_DEVICE void cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::shared_store_fence | ( | bool | in_loop | ) |
| inline static

## Member Data Documentation

## ◆ kLayoutA

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

MatrixLayout::Kind const cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::kLayoutA = GlobalLoadStreamA::kLayout
| static

## ◆ kLayoutB

template<typename GemmConfig_, typename GlobalLoadStreamA_, typename GlobalLoadStreamB_, typename SharedLoadStreamA_, typename SharedLoadStreamB_, typename Epilogue_, typename BlockSwizzle_ = IdentityBlockSwizzle, typename Index_ = int, typename ClearAccumulators_ = ClearAccumulators<typename GemmConfig_::Accumulators::Element>>

MatrixLayout::Kind const cutlass::gemm::GemmTraits< GemmConfig_, GlobalLoadStreamA_, GlobalLoadStreamB_, SharedLoadStreamA_, SharedLoadStreamB_, Epilogue_, BlockSwizzle_, Index_, ClearAccumulators_ >::kLayoutB = GlobalLoadStreamB::kLayout
| static

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

- gemm_traits.h