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...