CUTLASS: cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable > Class Template Reference

| | CUTLASS

CUDA Templates for Linear Algebra Subroutines and Solvers |

• Main Page

• Modules

• Namespaces

• Classes

• Files

• Class List

• Class Index

• Class Hierarchy

• Class Members

• cutlass

• gemm

• threadblock

• MmaBase

Classes | Public Types | Public Member Functions | Static Public Attributes | Protected Attributes | List of all members

cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable > Class Template Reference

#include <mma_base.h>

|

Classes

| | class | SharedStorage | | | Shared storage object needed by threadblock-scoped GEMM. More...
| | |

|

Public Types

| | using | Shape = Shape_ | | | Policy describing tuning details. More...
| | | | using | Policy = Policy_ | | | | using | Operator = typename Policy::Operator | | | Warp-level Mma. More...
| | | | using | WarpGemm = typename Policy::Operator::Shape | | | | using | WarpCount = GemmShape< Shape::kM/WarpGemm::kM, Shape::kN/WarpGemm::kN, Shape::kK/WarpGemm::kK > | | | Shape describing the number of warps filling the CTA. More...
| | | | using | TensorRefA = TensorRef< typename Operator::ElementA, typename Operator::LayoutA > | | | Tensor reference to the A operand. More...
| | | | using | TensorRefB = TensorRef< typename Operator::ElementB, typename Operator::LayoutB > | | | Tensor reference to the B operand. More...
| | |

|

Public Member Functions

| | CUTLASS_DEVICE | MmaBase (SharedStorage &shared_storage, int thread_idx, int warp_idx, int lane_idx) | | | Construct from tensor references. More...
| | |

|

Static Public Attributes

| | static int const | kWarpGemmIterations | | | Number of warp-level GEMM oeprations. More...
| | | | static int const | kStages = Stages | | | Number of stages. More...
| | |

|

Protected Attributes

| | Operator::IteratorA | warp_tile_iterator_A_ | | | Iterator to load a warp-scoped tile of A operand from shared memory. More...
| | | | Operator::IteratorB | warp_tile_iterator_B_ | | | Iterator to load a warp-scoped tile of B operand from shared memory. More...
| | |

Detailed Description

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool> class cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >

Structure to compute the matrix product targeting CUDA cores and SIMT math instructions.

Member Typedef Documentation

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

| using cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::Operator = typename Policy::Operator |

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

| using cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::Policy = Policy_ |

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

| using cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::Shape = Shape_ |

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

| using cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::TensorRefA = TensorRef<typename Operator::ElementA, typename Operator::LayoutA> |

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

| using cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::TensorRefB = TensorRef<typename Operator::ElementB, typename Operator::LayoutB> |

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

| using cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::WarpCount = GemmShape<Shape::kM / WarpGemm::kM, Shape::kN / WarpGemm::kN, Shape::kK / WarpGemm::kK> |

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

| using cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::WarpGemm = typename Policy::Operator::Shape |

Shape describing the overall GEMM computed from shared memory by each warp.

Constructor & Destructor Documentation

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

|

| CUTLASS_DEVICE cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::MmaBase | ( | SharedStorage & | shared_storage, | | | | int | thread_idx, | | | | int | warp_idx, | | | | int | lane_idx | | | ) | | |

| inline |

Parameters

| shared_storage | Shared storage needed for internal use by threadblock-scoped GEMM ID within the threadblock | | thread_idx | ID of warp | | warp_idx | ID of each thread within a warp |

Member Data Documentation

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

|

| int const cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::kStages = Stages |

| static |

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

|

| int const cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::kWarpGemmIterations |

| static |

Initial value:

=

(WarpGemm::kK / Operator::Policy::MmaShape::kK)

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

|

| Operator::IteratorA cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::warp_tile_iterator_A_ |

| protected |

template<typename Shape_, typename Policy_, int Stages, typename Enable = bool>

|

| Operator::IteratorB cutlass::gemm::threadblock::MmaBase< Shape_, Policy_, Stages, Enable >::warp_tile_iterator_B_ |

| protected |

---

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

• mma_base.h

---

Generated by 1.8.11