MLIR
17.0.0git
|
Namespaces | |
amd | |
Classes | |
struct | KernelDim3 |
Utility class for the GPU dialect to represent triples of Value s accessible through .x , .y , and .z similarly to CUDA notation. More... | |
class | AsyncTokenType |
struct | MMAMatrixStorageType |
MMAMatrixType storage and uniquing. More... | |
class | MMAMatrixType |
MMAMatrix represents a matrix held by a subgroup for matrix-matrix multiply accumulate operations. More... | |
class | SparseHandleType |
class | SerializeToBlobPass |
Base pass class to serialize kernel functions through LLVM into user-specified IR and add the resulting blob as module attribute. More... | |
Typedefs | |
using | SparseEnvHandleType = SparseHandleType< SparseHandleKind::Env > |
using | SparseDnVecHandleType = SparseHandleType< SparseHandleKind::DnVec > |
using | SparseDnMatHandleType = SparseHandleType< SparseHandleKind::DnMat > |
using | SparseSpMatHandleType = SparseHandleType< SparseHandleKind::SpMat > |
Enumerations | |
enum class | SparseHandleKind { Env , DnVec , DnMat , SpMat } |
Functions | |
void | addAsyncDependency (Operation *op, Value token) |
void | registerTransformDialectExtension (DialectRegistry ®istry) |
StringRef | getMappingAttrName () |
Name of the mapping attribute produced by loop mappers. More... | |
LogicalResult | setMappingAttr (scf::ParallelOp ploopOp, ArrayRef< ParallelLoopDimMappingAttr > mapping) |
Sets the mapping attribute of a scf.parallel operation. More... | |
std::string | getDefaultGpuBinaryAnnotation () |
Returns the default annotation name for GPU binary blobs. More... | |
static MappingLevel & | operator++ (MappingLevel &mappingLevel) |
Bounded increment on MappingLevel. More... | |
static Processor | getHardwareIdForMapping (MappingLevel level, int dimension) |
Computed the hardware id to use for a given mapping level. More... | |
static void | mapParallelOp (ParallelOp parallelOp, MappingLevel mappingLevel=MapGrid) |
Add mapping information to the given parallel loop. More... | |
Variables | |
static constexpr int | kNumHardwareIds = 3 |
using mlir::gpu::SparseDnMatHandleType = typedef SparseHandleType<SparseHandleKind::DnMat> |
Definition at line 181 of file GPUDialect.h.
using mlir::gpu::SparseDnVecHandleType = typedef SparseHandleType<SparseHandleKind::DnVec> |
Definition at line 180 of file GPUDialect.h.
using mlir::gpu::SparseEnvHandleType = typedef SparseHandleType<SparseHandleKind::Env> |
Definition at line 179 of file GPUDialect.h.
using mlir::gpu::SparseSpMatHandleType = typedef SparseHandleType<SparseHandleKind::SpMat> |
Definition at line 182 of file GPUDialect.h.
|
strong |
Enumerator | |
---|---|
Env | |
DnVec | |
DnMat | |
SpMat |
Definition at line 168 of file GPUDialect.h.
Definition at line 538 of file GPUDialect.cpp.
std::string mlir::gpu::getDefaultGpuBinaryAnnotation | ( | ) |
Returns the default annotation name for GPU binary blobs.
Definition at line 33 of file SerializeToBlob.cpp.
|
static |
Computed the hardware id to use for a given mapping level.
Will assign x,y and z hardware ids for the first 3 dimensions and use sequential after. TODO: Make this use x for the inner-most loop that is distributed to map to x, the next innermost to y and the next innermost to z.
Definition at line 73 of file ParallelLoopMapper.cpp.
References kNumHardwareIds.
Referenced by mapParallelOp().
StringRef mlir::gpu::getMappingAttrName | ( | ) |
Name of the mapping attribute produced by loop mappers.
Definition at line 31 of file ParallelLoopMapper.cpp.
Referenced by mlir::configureParallelLoopToGPULegality(), mapParallelOp(), and processParallelLoop().
|
static |
Add mapping information to the given parallel loop.
Do not add mapping information if the loop already has it. Also, don't start a mapping at a nested loop.
Definition at line 109 of file ParallelLoopMapper.cpp.
References mlir::Builder::getAttr(), mlir::Builder::getDimIdentityMap(), getHardwareIdForMapping(), getMappingAttrName(), and setMappingAttr().
|
static |
Bounded increment on MappingLevel.
Increments to the next level unless Sequential was already reached.
Definition at line 60 of file ParallelLoopMapper.cpp.
void mlir::gpu::registerTransformDialectExtension | ( | DialectRegistry & | registry | ) |
Definition at line 920 of file GPUTransformOps.cpp.
References mlir::DialectRegistry::addExtensions().
Referenced by mlir::registerAllDialects().
LogicalResult mlir::gpu::setMappingAttr | ( | scf::ParallelOp | ploopOp, |
ArrayRef< ParallelLoopDimMappingAttr > | mapping | ||
) |
Sets the mapping attribute of a scf.parallel operation.
Verifies that the mapping passed is valid.
ploopOp
.Referenced by mapParallelOp().
|
staticconstexpr |
Definition at line 56 of file ParallelLoopMapper.cpp.
Referenced by getHardwareIdForMapping().