CUTLASS
CUDA Templates for Linear Algebra Subroutines and Solvers
Classes | Public Types | Public Member Functions | Static Public Member Functions | Static Public Attributes | List of all members
cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero > Class Template Reference

Partial specialization for column-major output exchanges problem size and operand.

#include <gemm.h>

Classes

struct  Arguments
 Argument structure. More...
 

Public Types

using ElementA = ElementA_
 
using LayoutA = LayoutA_
 
using TensorRefA = TensorRef< ElementA const, LayoutA >
 
using ElementB = ElementB_
 
using LayoutB = LayoutB_
 
using TensorRefB = TensorRef< ElementB const, LayoutB >
 
using ElementC = ElementC_
 
using LayoutC = layout::ColumnMajor
 
using TensorRefC = TensorRef< ElementC const, LayoutC >
 
using TensorRefD = TensorRef< ElementC, LayoutC >
 
using ElementAccumulator = ElementAccumulator_
 
using OperatorClass = OperatorClass_
 
using ArchTag = ArchTag_
 
using ThreadblockShape = ThreadblockShape_
 
using WarpShape = WarpShape_
 
using InstructionShape = InstructionShape_
 
using EpilogueOutputOp = EpilogueOutputOp_
 
using ThreadblockSwizzle = ThreadblockSwizzle_
 
using Operator = Operator_
 
using UnderlyingOperator = Gemm< ElementB, typename layout::LayoutTranspose< LayoutB >::type, ElementA, typename layout::LayoutTranspose< LayoutA >::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero >
 
using UnderlyingArguments = typename UnderlyingOperator::Arguments
 
using GemmKernel = typename UnderlyingOperator::GemmKernel
 

Public Member Functions

 Gemm ()
 Constructs the GEMM. More...
 
Status initialize (Arguments const &args, void *workspace=nullptr, cudaStream_t stream=nullptr)
 Initializes GEMM state from arguments. More...
 
Status update (Arguments const &args, void *workspace=nullptr)
 Lightweight update given a subset of arguments. More...
 
Status run (cudaStream_t stream=nullptr)
 Runs the kernel using initialized state. More...
 
Status operator() (cudaStream_t stream=nullptr)
 Runs the kernel using initialized state. More...
 
Status operator() (Arguments const &args, void *workspace=nullptr, cudaStream_t stream=nullptr)
 Runs the kernel using initialized state. More...
 

Static Public Member Functions

static UnderlyingArguments to_underlying_arguments (Arguments const &args)
 Helper to construct a transposed equivalent for the underying GEMM operator. More...
 
static Status can_implement (Arguments const &args)
 Determines whether the GEMM can execute the given problem. More...
 
static size_t get_workspace_size (Arguments const &args)
 Gets the workspace size. More...
 

Static Public Attributes

static int const kStages = Stages
 
static int const kAlignmentA = AlignmentA
 
static int const kAlignmentB = AlignmentB
 
static bool const kSplitKSerial = SplitKSerial
 
static bool const kIsBetaZero = IsBetaZero
 
static int const kAlignmentC = UnderlyingOperator::kAlignmentC
 

Member Typedef Documentation

template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::ArchTag = ArchTag_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::ElementA = ElementA_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::ElementAccumulator = ElementAccumulator_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::ElementB = ElementB_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::ElementC = ElementC_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::EpilogueOutputOp = EpilogueOutputOp_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::GemmKernel = typename UnderlyingOperator::GemmKernel
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::InstructionShape = InstructionShape_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::LayoutA = LayoutA_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::LayoutB = LayoutB_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::LayoutC = layout::ColumnMajor
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::Operator = Operator_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::OperatorClass = OperatorClass_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::TensorRefA = TensorRef<ElementA const, LayoutA>
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::TensorRefB = TensorRef<ElementB const, LayoutB>
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::TensorRefC = TensorRef<ElementC const, LayoutC>
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::TensorRefD = TensorRef<ElementC, LayoutC>
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::ThreadblockShape = ThreadblockShape_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::ThreadblockSwizzle = ThreadblockSwizzle_
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::UnderlyingArguments = typename UnderlyingOperator::Arguments
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::UnderlyingOperator = Gemm< ElementB, typename layout::LayoutTranspose<LayoutB>::type, ElementA, typename layout::LayoutTranspose<LayoutA>::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero >
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
using cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::WarpShape = WarpShape_

Constructor & Destructor Documentation

template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::Gemm ( )
inline

Member Function Documentation

template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
static Status cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::can_implement ( Arguments const &  args)
inlinestatic
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
static size_t cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::get_workspace_size ( Arguments const &  args)
inlinestatic
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
Status cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::initialize ( Arguments const &  args,
void *  workspace = nullptr,
cudaStream_t  stream = nullptr 
)
inline
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
Status cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::operator() ( cudaStream_t  stream = nullptr)
inline
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
Status cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::operator() ( Arguments const &  args,
void *  workspace = nullptr,
cudaStream_t  stream = nullptr 
)
inline
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
Status cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::run ( cudaStream_t  stream = nullptr)
inline
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
static UnderlyingArguments cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::to_underlying_arguments ( Arguments const &  args)
inlinestatic
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
Status cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::update ( Arguments const &  args,
void *  workspace = nullptr 
)
inline

Member Data Documentation

template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
int const cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::kAlignmentA = AlignmentA
static
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
int const cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::kAlignmentB = AlignmentB
static
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
int const cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::kAlignmentC = UnderlyingOperator::kAlignmentC
static
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
bool const cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::kIsBetaZero = IsBetaZero
static
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
bool const cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::kSplitKSerial = SplitKSerial
static
template<typename ElementA_ , typename LayoutA_ , typename ElementB_ , typename LayoutB_ , typename ElementC_ , typename ElementAccumulator_ , typename OperatorClass_ , typename ArchTag_ , typename ThreadblockShape_ , typename WarpShape_ , typename InstructionShape_ , typename EpilogueOutputOp_ , typename ThreadblockSwizzle_ , int Stages, int AlignmentA, int AlignmentB, bool SplitKSerial, typename Operator_ , bool IsBetaZero>
int const cutlass::gemm::device::Gemm< ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero >::kStages = Stages
static
The documentation for this class was generated from the following file:
Generated by   doxygen 1.8.11