NVGPUDialect.h File Reference

#include "mlir/Bytecode/BytecodeOpInterface.h"
#include "mlir/IR/BuiltinTypes.h"
#include "mlir/IR/Dialect.h"
#include "mlir/IR/OpDefinition.h"
#include "mlir/Interfaces/InferTypeOpInterface.h"
#include "mlir/Interfaces/SideEffectInterfaces.h"
#include "mlir/Dialect/NVGPU/IR/NVGPUEnums.h.inc"
#include "mlir/Dialect/NVGPU/IR/NVGPUAttrDefs.h.inc"
#include "mlir/Dialect/NVGPU/IR/NVGPUTypes.h.inc"
#include "mlir/Dialect/NVGPU/IR/NVGPUDialect.h.inc"
#include "mlir/Dialect/NVGPU/IR/NVGPU.h.inc"

Go to the source code of this file.

Macros
#define	GET_ATTRDEF_CLASSES
#define	GET_TYPEDEF_CLASSES
#define	GET_OP_CLASSES

Variables
constexpr int	kWarpSize = 32
constexpr int	kWgmmaSizeM = 64
	M size of wgmma.mma_async instruction. More...
constexpr unsigned	kMaxTMATensorDimension = 5
	Maximum TMA tile dimension (tensorRank) must be non-zero and less than or equal to the maximum supported dimensionality of 5. More...
constexpr unsigned	kMaxTMADimension = 256
	Maximum TMA tile size (boxDim), which specifies number of elements to be traversed along each of the kMaxTMATensorDimension (tensorRank) dimensions, must be non-zero and less than or equal to 256. More...
constexpr unsigned	kMaxTMALastdimByte = 128
	Last dimension of 2D+ TMA must be 128 bytes. More...

Macro Definition Documentation

GET_ATTRDEF_CLASSES

#define GET_ATTRDEF_CLASSES

Definition at line 40 of file NVGPUDialect.h.

GET_OP_CLASSES

#define GET_OP_CLASSES

Definition at line 48 of file NVGPUDialect.h.

GET_TYPEDEF_CLASSES

#define GET_TYPEDEF_CLASSES

Definition at line 43 of file NVGPUDialect.h.

Variable Documentation

kMaxTMADimension

constexpr unsigned kMaxTMADimension = 256

constexpr

Maximum TMA tile size (boxDim), which specifies number of elements to be traversed along each of the kMaxTMATensorDimension (tensorRank) dimensions, must be non-zero and less than or equal to 256.

Definition at line 36 of file NVGPUDialect.h.

Referenced by verifyTmaDescriptorWithMemref().

kMaxTMALastdimByte

constexpr unsigned kMaxTMALastdimByte = 128

constexpr

Last dimension of 2D+ TMA must be 128 bytes.

Definition at line 38 of file NVGPUDialect.h.

Referenced by verifyTmaDescriptorWithMemref().

kMaxTMATensorDimension

constexpr unsigned kMaxTMATensorDimension = 5

constexpr

Maximum TMA tile dimension (tensorRank) must be non-zero and less than or equal to the maximum supported dimensionality of 5.

Definition at line 32 of file NVGPUDialect.h.

kWarpSize

constexpr int kWarpSize = 32

constexpr

Definition at line 25 of file NVGPUDialect.h.

kWgmmaSizeM

constexpr int kWgmmaSizeM = 64

constexpr

M size of wgmma.mma_async instruction.

Definition at line 28 of file NVGPUDialect.h.

Referenced by isAllowedSizeM().