MLIR  20.0.0git
Classes | Namespaces | Typedefs | Functions
Utils.h File Reference
#include "mlir/Dialect/GPU/IR/GPUDialect.h"
#include "mlir/Dialect/SCF/IR/DeviceMappingInterface.h"
#include "mlir/Dialect/Transform/Interfaces/TransformInterfaces.h"
#include "mlir/IR/OpImplementation.h"
#include "mlir/IR/PatternMatch.h"

Go to the source code of this file.

Classes

struct  mlir::transform::gpu::IdBuilderResult
 Helper type for functions that generate ids for the mapping of a scf.forall. More...
 
struct  mlir::transform::gpu::GpuIdBuilder
 Helper struct for configuring the rewrite of mapped scf.forall ops to various gpu id configurations. More...
 
struct  mlir::transform::gpu::GpuBlockIdBuilder
 Builder for gpu::BlockIdOps used to map scf.forall to blocks. More...
 
struct  mlir::transform::gpu::GpuWarpgroupIdBuilder
 Builder for warpgroup ids used to map scf.forall to reindexed warpgroups. More...
 
struct  mlir::transform::gpu::GpuWarpIdBuilder
 Builder for warp ids used to map scf.forall to reindexed warps. More...
 
struct  mlir::transform::gpu::GpuThreadIdBuilder
 Builder for warp ids used to map scf.forall to reindexed threads. More...
 

Namespaces

 mlir
 Include the generated interface declarations.
 
 mlir::gpu
 
 mlir::scf
 
 mlir::transform
 
 mlir::transform::gpu
 

Typedefs

using mlir::transform::gpu::GpuIdBuilderFnType = std::function< IdBuilderResult(RewriterBase &, Location, ArrayRef< int64_t >, ArrayRef< int64_t >)>
 Common gpu id builder type, allows the configuration of lowering for various mapping schemes. More...
 

Functions

DiagnosedSilenceableFailure mlir::transform::gpu::checkGpuLimits (TransformOpInterface transformOp, std::optional< int64_t > gridDimX, std::optional< int64_t > gridDimY, std::optional< int64_t > gridDimZ, std::optional< int64_t > blockDimX, std::optional< int64_t > blockDimY, std::optional< int64_t > blockDimZ)
 Determine if the size of the kernel configuration is supported by the GPU architecture being used. More...
 
DiagnosedSilenceableFailure mlir::transform::gpu::createGpuLaunch (RewriterBase &rewriter, Location loc, TransformOpInterface transformOp, mlir::gpu::LaunchOp &launchOp, std::optional< int64_t > gridDimX=std::nullopt, std::optional< int64_t > gridDimY=std::nullopt, std::optional< int64_t > gridDimZ=std::nullopt, std::optional< int64_t > blockDimX=std::nullopt, std::optional< int64_t > blockDimY=std::nullopt, std::optional< int64_t > blockDimZ=std::nullopt)
 Create an empty-body gpu::LaunchOp using the provided kernel settings and put a terminator within. More...
 
DiagnosedSilenceableFailure mlir::transform::gpu::alterGpuLaunch (RewriterBase &rewriter, mlir::gpu::LaunchOp gpuLaunch, TransformOpInterface transformOp, std::optional< int64_t > gridDimX=std::nullopt, std::optional< int64_t > gridDimY=std::nullopt, std::optional< int64_t > gridDimZ=std::nullopt, std::optional< int64_t > blockDimX=std::nullopt, std::optional< int64_t > blockDimY=std::nullopt, std::optional< int64_t > blockDimZ=std::nullopt)
 Alter kernel configuration of the given kernel. More...
 
DiagnosedSilenceableFailure mlir::transform::gpu::findTopLevelForallOp (Operation *target, scf::ForallOp &topLevelForallOp, TransformOpInterface transformOp)
 Find the unique top level scf::ForallOp within a given target op. More...