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/NVGPUTypeDefs.h.inc"
#include "mlir/Dialect/NVGPU/IR/NVGPUDialect.h.inc"
#include "mlir/Dialect/NVGPU/IR/NVGPUOps.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	kMaxTotalBlockdim = 1024
constexpr int	kMaxBlockdimx = 1024
constexpr int	kMaxBlockdimy = 1024
constexpr int	kMaxBlockdimz = 64
constexpr int	kMaxTotalGriddim = 2147483647
constexpr int	kMaxGriddimx = 2147483647
constexpr int	kMaxGriddimy = 65535
constexpr int	kMaxGriddimz = 65535
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 52 of file NVGPUDialect.h.

GET_OP_CLASSES

#define GET_OP_CLASSES

Definition at line 60 of file NVGPUDialect.h.

GET_TYPEDEF_CLASSES

#define GET_TYPEDEF_CLASSES

Definition at line 55 of file NVGPUDialect.h.

Variable Documentation

kMaxBlockdimx

constexpr int kMaxBlockdimx = 1024

constexpr

Definition at line 31 of file NVGPUDialect.h.

Referenced by mlir::transform::gpu::checkGpuLimits().

kMaxBlockdimy

constexpr int kMaxBlockdimy = 1024

constexpr

Definition at line 32 of file NVGPUDialect.h.

Referenced by mlir::transform::gpu::checkGpuLimits().

kMaxBlockdimz

constexpr int kMaxBlockdimz = 64

constexpr

Definition at line 33 of file NVGPUDialect.h.

Referenced by mlir::transform::gpu::checkGpuLimits().

kMaxGriddimx

constexpr int kMaxGriddimx = 2147483647

constexpr

Definition at line 35 of file NVGPUDialect.h.

Referenced by mlir::transform::gpu::checkGpuLimits().

kMaxGriddimy

constexpr int kMaxGriddimy = 65535

constexpr

Definition at line 36 of file NVGPUDialect.h.

Referenced by mlir::transform::gpu::checkGpuLimits().

kMaxGriddimz

constexpr int kMaxGriddimz = 65535

constexpr

Definition at line 37 of file NVGPUDialect.h.

Referenced by mlir::transform::gpu::checkGpuLimits().

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 48 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 50 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 44 of file NVGPUDialect.h.

kMaxTotalBlockdim

constexpr int kMaxTotalBlockdim = 1024

constexpr

Definition at line 30 of file NVGPUDialect.h.

Referenced by mlir::transform::gpu::checkGpuLimits().

kMaxTotalGriddim

constexpr int kMaxTotalGriddim = 2147483647

constexpr

Definition at line 34 of file NVGPUDialect.h.

Referenced by mlir::transform::gpu::checkGpuLimits().

kWarpSize

constexpr int kWarpSize = 32

constexpr

Definition at line 26 of file NVGPUDialect.h.

kWgmmaSizeM

constexpr int kWgmmaSizeM = 64

constexpr

M size of wgmma.mma_async instruction.

Definition at line 40 of file NVGPUDialect.h.

Referenced by isAllowedSizeM().