Back to Cutlass

CUTLASS: cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ > Struct Template Reference

docs/structcutlass_1_1gemm_1_1kernel_1_1DefaultGemv.html

4.4.221.0 KB
Original Source

| | CUTLASS

CUDA Templates for Linear Algebra Subroutines and Solvers |

Public Types | List of all members

cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ > Struct Template Reference

#include <default_gemv.h>

|

Public Types

| | using | ThreadBlockShape = ThreadBlockShape_ | | | Shape of Threadblock-level matrix operation (concept: GemmShape) More...
| | | | using | ThreadShape = ThreadShape_ | | | Shape of warp-level matrix operation (concept: GemmShape) More...
| | | | using | ElementA = ElementA_ | | | Data type of multiplicand A. More...
| | | | using | LayoutA = LayoutA_ | | | Layout of multiplicand A. More...
| | | | using | ElementB = ElementB_ | | | Data type of multiplicand B. More...
| | | | using | LayoutB = LayoutB_ | | | Layout of multiplicand B. More...
| | | | using | ElementAccumulator = ElementAccumulator_ | | | Data type of accumulators. More...
| | | | using | LayoutAccumulator = LayoutCD_ | | | Data type of accumulators (same as C/D) More...
| | | | using | ElementCD = ElementCD_ | | | Data type of input/output matrix C/D. More...
| | | | using | LayoutCD = LayoutCD_ | | | Layout of input/output matrix C/D. More...
| | | | using | Core = typename cutlass::gemm::threadblock::DefaultGemvCore< ThreadBlockShape, ThreadShape, ElementA, LayoutA, ElementB, LayoutB, ElementAccumulator, LayoutAccumulator > | | | | using | ThreadBlockGemv = cutlass::gemm::threadblock::Gemv< Core > | | | | using | IteratorA = typename ThreadBlockGemv::IteratorA | | | | using | IteratorB = typename ThreadBlockGemv::IteratorB | | | | using | IteratorPolicyCD = typename platform::conditional< platform::is_same< LayoutCD, layout::RowMajor >::value, cutlass::transform::PitchLinearTilePolicyStripminedThreadContiguous< layout::PitchLinearShape< ThreadBlockShape::kN, ThreadBlockShape::kM >, Core::kThreadsPerN, ThreadShape::kN >, cutlass::transform::PitchLinearTilePolicyStripminedThreadStrided< layout::PitchLinearShape< ThreadBlockShape::kM, ThreadBlockShape::kN >, Core::kThreadsPerN, ThreadShape::kM >>::type | | | Policy for the iterator that reads/writes C/D. More...
| | | | using | IteratorCD = cutlass::transform::threadblock::PredicatedTileIterator< cutlass::MatrixShape< ThreadBlockShape::kM, ThreadBlockShape::kN >, ElementCD, LayoutCD, 0, IteratorPolicyCD > | | | Iterator that reads/writes C/D. More...
| | | | using | FragmentCD = typename IteratorCD::Fragment | | | Fragment storage for C/D. More...
| | | | using | ThreadBlockSwizzle = cutlass::gemm::threadblock::GemvBatchedStridedThreadblockDefaultSwizzle | | |

Member Typedef Documentation

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::Core = typename cutlass::gemm::threadblock::DefaultGemvCore< ThreadBlockShape, ThreadShape, ElementA, LayoutA, ElementB, LayoutB, ElementAccumulator, LayoutAccumulator> |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::ElementA = ElementA_ |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::ElementAccumulator = ElementAccumulator_ |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::ElementB = ElementB_ |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::ElementCD = ElementCD_ |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::FragmentCD = typename IteratorCD::Fragment |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::IteratorA = typename ThreadBlockGemv::IteratorA |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::IteratorB = typename ThreadBlockGemv::IteratorB |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::IteratorCD = cutlass::transform::threadblock::PredicatedTileIterator< cutlass::MatrixShape<ThreadBlockShape::kM, ThreadBlockShape::kN>, ElementCD, LayoutCD, 0, IteratorPolicyCD> |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::IteratorPolicyCD = typename platform::conditional< platform::is_same<LayoutCD, layout::RowMajor>::value, cutlass::transform::PitchLinearTilePolicyStripminedThreadContiguous< layout::PitchLinearShape<ThreadBlockShape::kN, ThreadBlockShape::kM>, Core::kThreadsPerN, ThreadShape::kN>, cutlass::transform::PitchLinearTilePolicyStripminedThreadStrided< layout::PitchLinearShape<ThreadBlockShape::kM, ThreadBlockShape::kN>, Core::kThreadsPerN, ThreadShape::kM>>::type |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::LayoutA = LayoutA_ |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::LayoutAccumulator = LayoutCD_ |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::LayoutB = LayoutB_ |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::LayoutCD = LayoutCD_ |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::ThreadBlockGemv = cutlass::gemm::threadblock::Gemv<Core> |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::ThreadBlockShape = ThreadBlockShape_ |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::ThreadBlockSwizzle = cutlass::gemm::threadblock::GemvBatchedStridedThreadblockDefaultSwizzle |

template<typename ThreadBlockShape_ , typename ThreadShape_ , typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementCD_ , typename LayoutCD_ , typename ElementAccumulator_ = ElementCD_>

| using cutlass::gemm::kernel::DefaultGemv< ThreadBlockShape_, ThreadShape_, ElementA_, LayoutA_, ElementB_, LayoutB_, ElementCD_, LayoutCD_, ElementAccumulator_ >::ThreadShape = ThreadShape_ |


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


Generated by 1.8.11