|
struct | Mma |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, complex< double >, LayoutA, complex< double >, LayoutB, complex< double >, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, complex< double >, LayoutA, double, LayoutB, complex< double >, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, complex< float >, LayoutA, complex< float >, LayoutB, complex< float >, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, complex< float >, LayoutA, float, LayoutB, complex< float >, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, double, LayoutA, complex< double >, LayoutB, complex< double >, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, double, LayoutA, double, LayoutB, double, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, ElementA, LayoutA, ElementB, LayoutB, ElementC, LayoutC, Operator > |
| Matrix multiply-add operation - specialized for 1x1x1x1 matrix multiply operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, float, LayoutA, complex< float >, LayoutB, complex< float >, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, float, LayoutA, float, LayoutB, float, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, half_t, LayoutA, half_t, LayoutB, float, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 1 >, 1, int, LayoutA, int, LayoutB, int, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 2 >, 1, int16_t, layout::RowMajor, int16_t, layout::ColumnMajor, int, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 1, 4 >, 1, int8_t, LayoutA, int8_t, LayoutB, int, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 1, 2, 1 >, 1, half_t, LayoutA, half_t, LayoutB, half_t, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 16, 16, 4 >, 32, half_t, LayoutA, half_t, LayoutB, ElementC, LayoutC, Operator > |
| Matrix multiply-add operation specialized for the entire warp. More...
|
|
struct | Mma< gemm::GemmShape< 16, 8, 8 >, 32, half_t, layout::RowMajor, half_t, layout::ColumnMajor, float, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: F32 = F16 * F16 + F32. More...
|
|
struct | Mma< gemm::GemmShape< 16, 8, 8 >, 32, half_t, layout::RowMajor, half_t, layout::ColumnMajor, half_t, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation - F16 = F16 * F16 + F16. More...
|
|
struct | Mma< gemm::GemmShape< 2, 1, 1 >, 1, half_t, LayoutA, half_t, LayoutB, half_t, LayoutC, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 2, 2, 1 >, 1, half_t, layout::ColumnMajor, half_t, layout::RowMajor, half_t, layout::ColumnMajor, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 2, 2, 1 >, 1, half_t, layout::ColumnMajor, half_t, layout::RowMajor, half_t, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 128 >, 32, uint1b_t, layout::RowMajor, uint1b_t, layout::ColumnMajor, int, layout::RowMajor, OpXorPopc > |
| Matrix multiply-add operation. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 16 >, 32, int8_t, layout::RowMajor, int8_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: S32 = S8 * S8 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 16 >, 32, int8_t, layout::RowMajor, int8_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAddSaturate > |
| Matrix multiply-add operation: S32 = S8 * S8 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 16 >, 32, int8_t, layout::RowMajor, uint8_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: S32 = S8 * U8 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 16 >, 32, int8_t, layout::RowMajor, uint8_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAddSaturate > |
| Matrix multiply-add operation: S32 = S8 * U8 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 16 >, 32, uint8_t, layout::RowMajor, int8_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: S32 = U8 * S8 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 16 >, 32, uint8_t, layout::RowMajor, int8_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAddSaturate > |
| Matrix multiply-add operation: S32 = U8 * S8 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 16 >, 32, uint8_t, layout::RowMajor, uint8_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: S32 = S8 * U8 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 16 >, 32, uint8_t, layout::RowMajor, uint8_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAddSaturate > |
| Matrix multiply-add operation: S32 = S8 * U8 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 32 >, 32, int4b_t, layout::RowMajor, int4b_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: S32 = S4 * S4 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 32 >, 32, int4b_t, layout::RowMajor, int4b_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAddSaturate > |
| Matrix multiply-add operation: S32 = S4 * S4 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 32 >, 32, int4b_t, layout::RowMajor, uint4b_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: S32 = S4 * U4 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 32 >, 32, int4b_t, layout::RowMajor, uint4b_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAddSaturate > |
| Matrix multiply-add operation: S32 = S4 * U4 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 32 >, 32, uint4b_t, layout::RowMajor, int4b_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: S32 = U4 * S4 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 32 >, 32, uint4b_t, layout::RowMajor, int4b_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAddSaturate > |
| Matrix multiply-add operation: S32 = U4 * S4 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 32 >, 32, uint4b_t, layout::RowMajor, uint4b_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: S32 = U4 * U4 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 32 >, 32, uint4b_t, layout::RowMajor, uint4b_t, layout::ColumnMajor, int, layout::RowMajor, OpMultiplyAddSaturate > |
| Matrix multiply-add operation: S32 = U4 * U4 + S32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 4 >, 8, half_t, layout::ColumnMajor, half_t, layout::ColumnMajor, float, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: F32 = F16 * F16 + F32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 4 >, 8, half_t, layout::ColumnMajor, half_t, layout::ColumnMajor, half_t, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: F16 = F16 * F16 + F16. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 4 >, 8, half_t, layout::ColumnMajor, half_t, layout::RowMajor, float, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: F32 = F16 * F16 + F32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 4 >, 8, half_t, layout::ColumnMajor, half_t, layout::RowMajor, half_t, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: F16 = F16 * F16 + F16. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 4 >, 8, half_t, layout::RowMajor, half_t, layout::ColumnMajor, float, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: F32 = F16 * F16 + F32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 4 >, 8, half_t, layout::RowMajor, half_t, layout::ColumnMajor, half_t, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: F16 = F16 * F16 + F16. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 4 >, 8, half_t, layout::RowMajor, half_t, layout::RowMajor, float, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: F32 = F16 * F16 + F32. More...
|
|
struct | Mma< gemm::GemmShape< 8, 8, 4 >, 8, half_t, layout::RowMajor, half_t, layout::RowMajor, half_t, layout::RowMajor, OpMultiplyAdd > |
| Matrix multiply-add operation: F16 = F16 * F16 + F16. More...
|
|
struct | PtxWmma |
| WMMA Matrix multiply-add operation. More...
|
|
struct | PtxWmmaLoadA |
| WMMA PTX string load for A, B, and C matrices. More...
|
|
struct | PtxWmmaLoadB |
|
struct | PtxWmmaLoadC |
|
struct | PtxWmmaStoreD |
| WMMA store for matrix D. More...
|
|
struct | Sm50 |
|
struct | Sm60 |
|
struct | Sm61 |
|
struct | Sm70 |
|
struct | Sm72 |
|
struct | Sm75 |
|
struct | Wmma< Shape_, cutlass::half_t, LayoutA_, cutlass::half_t, LayoutB_, ElementC_, LayoutC_, cutlass::arch::OpMultiplyAdd > |
|
struct | Wmma< Shape_, cutlass::int4b_t, LayoutA_, cutlass::int4b_t, LayoutB_, int32_t, LayoutC_, cutlass::arch::OpMultiplyAdd > |
|
struct | Wmma< Shape_, cutlass::uint1b_t, LayoutA_, cutlass::uint1b_t, LayoutB_, int32_t, LayoutC_, cutlass::arch::OpXorPopc > |
|
struct | Wmma< Shape_, int8_t, LayoutA_, int8_t, LayoutB_, int32_t, LayoutC_, cutlass::arch::OpMultiplyAdd > |
|
struct | Wmma< Shape_, uint8_t, LayoutA_, uint8_t, LayoutB_, int32_t, LayoutC_, cutlass::arch::OpMultiplyAdd > |
|
|
template<typename Layout , int MatrixCount> |
__device__ void | ldsm (Array< unsigned, MatrixCount > &D, void const *ptr) |
|
template<> |
__device__ void | ldsm< layout::RowMajor, 1 > (Array< unsigned, 1 > &D, void const *ptr) |
|
template<> |
__device__ void | ldsm< layout::RowMajor, 2 > (Array< unsigned, 2 > &D, void const *ptr) |
|
template<> |
__device__ void | ldsm< layout::RowMajor, 4 > (Array< unsigned, 4 > &D, void const *ptr) |
|
template<> |
__device__ void | ldsm< layout::ColumnMajor, 1 > (Array< unsigned, 1 > &D, void const *ptr) |
|
template<> |
__device__ void | ldsm< layout::ColumnMajor, 2 > (Array< unsigned, 2 > &D, void const *ptr) |
|
template<> |
__device__ void | ldsm< layout::ColumnMajor, 4 > (Array< unsigned, 4 > &D, void const *ptr) |
|
template<typename T , int N> |
CUTLASS_HOST_DEVICE Array< T, N > | operator* (Array< T, N > const &a, Array< T, N > const &b) |
|
template<typename T , int N> |
CUTLASS_HOST_DEVICE Array< T, N > | operator+ (Array< T, N > const &a, Array< T, N > const &b) |
|
template<typename T , int N> |
CUTLASS_HOST_DEVICE Array< T, N > | operator- (Array< T, N > const &a, Array< T, N > const &b) |
|
template<typename T , int N> |
CUTLASS_HOST_DEVICE Array< T, N > | mac (Array< T, N > const &a, Array< T, N > const &b, Array< T, N > const &c) |
|
template<typename Element , typename Accumulator , int N> |
CUTLASS_HOST_DEVICE Accumulator | dot (Array< T, N > const &a, Array< T, N > const &b, Accumulator accum) |
|
template<> |
CUTLASS_HOST_DEVICE Array< half_t, 2 > | operator* (Array< half_t, 2 > const &a, Array< half_t, 2 > const &b) |
|
template<> |
CUTLASS_HOST_DEVICE Array< half_t, 2 > | operator+ (AArray< half_t, 2 > const &a, Array< half_t, 2 > const &b) |
|
template<> |
CUTLASS_HOST_DEVICE Array< half_t, 2 > | operator- (Array< half_t, 2 > const &a, Array< half_t, 2 > const &b) |
|
template<> |
CUTLASS_HOST_DEVICE Array< half_t, 2 > | mac (Array< half_t, 2 > const &a, Array< half_t, 2 > const &b, Array< half_t, 2 > const &c) |
| Multiply-accumulate operators - specialized for half_t x 2. More...
|
|
template<> |
CUTLASS_HOST_DEVICE half_t | dot (Array< half_t, 2 > const &a, Array< half_t, 2 > const &b, half_t accum) |
| Dot product operator - specialized for half_t <- (half_t * half_t) x 2 + half_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE float | dot (Array< half_t, 2 > const &a, Array< half_t, 2 > const &b, float accum) |
| Dot product operator - specialized for float <- (half_t * half_t) x 2 + float. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< int8_t, 4 > const &a, Array< int8_t, 4 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (int8_t * int8_t) x 4 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< uint8_t, 4 > const &a, Array< int8_t, 4 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (uint8_t * int8_t) x 4 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< int8_t, 4 > const &a, Array< uint8_t, 4 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (int8_t * uint8_t) x 4 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< uint8_t, 4 > const &a, Array< uint8_t, 4 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (uint8_t * uint8_t) x 4 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< int16_t, 2 > const &a, Array< int8_t, 2 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (int16_t * int8_t) x 2 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< uint16_t, 2 > const &a, Array< int8_t, 2 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (uint16_t * int8_t) x 2 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< int16_t, 2 > const &a, Array< uint8_t, 2 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (int16_t * int8_t) x 2 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< uint16_t, 2 > const &a, Array< uint8_t, 2 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (uint16_t * int8_t) x 2 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< int16_t, 2 > const &a, Array< int16_t, 2 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (int16_t * int16_t) x 2 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< uint16_t, 2 > const &a, Array< int16_t, 2 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (uint16_t * int16_t) x 2 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< int16_t, 2 > const &a, Array< uint16_t, 2 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (int16_t * int16_t) x 2 + int32_t. More...
|
|
template<> |
CUTLASS_HOST_DEVICE int32_t | dot (Array< uint16_t, 2 > const &a, Array< uint16_t, 2 > const &b, int32_t accum) |
| Dot product operator - specialized for int32_t <- (uint16_t * int16_t) x 2 + int32_t. More...
|
|