MLIR  19.0.0git
Classes | Namespaces | Macros | Functions
Passes.h File Reference
#include "Utils.h"
#include "mlir/Dialect/GPU/IR/GPUDialect.h"
#include "mlir/IR/PatternMatch.h"
#include "mlir/Pass/Pass.h"
#include <optional>
#include "mlir/Dialect/GPU/Transforms/Passes.h.inc"

Go to the source code of this file.

Classes

class  mlir::gpu::SerializeToBlobPass
 Base pass class to serialize kernel functions through LLVM into user-specified IR and add the resulting blob as module attribute. More...
 

Namespaces

 llvm
 Include the generated interface declarations.
 
 mlir
 Include the generated interface declarations.
 
 mlir::func
 
 mlir::gpu
 

Macros

#define GEN_PASS_DECL
 
#define GEN_PASS_REGISTRATION
 Generate the code for registering passes. More...
 

Functions

std::unique_ptr< Passmlir::createGpuLauchSinkIndexComputationsPass ()
 Pass that moves ops which are likely an index computation into gpu.launch body. More...
 
std::unique_ptr< OperationPass< ModuleOp > > mlir::createGpuKernelOutliningPass (StringRef dataLayoutStr=StringRef())
 Replaces gpu.launch with gpu.launch_func by moving the region into a separate kernel function. More...
 
std::unique_ptr< OperationPass< func::FuncOp > > mlir::createGpuAsyncRegionPass ()
 Rewrites a function region so that GPU ops execute asynchronously. More...
 
std::unique_ptr< OperationPass< func::FuncOp > > mlir::createGpuMapParallelLoopsPass ()
 Maps the parallel loops found in the given function to workgroups. More...
 
void mlir::populateGpuGlobalIdPatterns (RewritePatternSet &patterns)
 Collect a set of patterns to rewrite GlobalIdOp op within the GPU dialect. More...
 
void mlir::populateGpuShufflePatterns (RewritePatternSet &patterns)
 Collect a set of patterns to rewrite shuffle ops within the GPU dialect. More...
 
void mlir::populateGpuAllReducePatterns (RewritePatternSet &patterns)
 Collect a set of patterns to rewrite all-reduce ops within the GPU dialect. More...
 
void mlir::populateGpuBreakDownSubgrupReducePatterns (RewritePatternSet &patterns, unsigned maxShuffleBitwidth=32, PatternBenefit benefit=1)
 Collect a set of patterns to break down subgroup_reduce ops into smaller ones supported by the target of size <= maxShuffleBitwidth, where size is the subgroup_reduce value bitwidth. More...
 
void mlir::populateGpuLowerSubgroupReduceToShufflePattenrs (RewritePatternSet &patterns, unsigned subgroupSize, unsigned shuffleBitwidth=32, PatternBenefit benefit=1)
 Collect a set of patterns to lower gpu.subgroup_reduce into gpu.shuffle ops over shuffleBitwidth scalar types. More...
 
void mlir::populateGpuRewritePatterns (RewritePatternSet &patterns)
 Collect all patterns to rewrite ops within the GPU dialect. More...
 
LogicalResult mlir::gpu::transformGpuModulesToBinaries (Operation *op, OffloadingLLVMTranslationAttrInterface handler=nullptr, const gpu::TargetOptions &options={})
 Searches for all GPU modules in op and transforms them into GPU binary operations. More...
 
void mlir::registerGpuSerializeToHsacoPass ()
 Register pass to serialize GPU kernel functions to a HSAco binary annotation. More...
 
std::unique_ptr< Pass > mlir::createGpuSerializeToHsacoPass (StringRef triple, StringRef arch, StringRef features, int optLevel)
 Create an instance of the GPU kernel function to HSAco binary serialization pass. More...
 
void mlir::populateGpuDecomposeMemrefsPatterns (RewritePatternSet &patterns)
 Collect a set of patterns to decompose memrefs ops. More...
 
std::unique_ptr< Passmlir::createGpuDecomposeMemrefsPass ()
 Pass decomposes memref ops inside gpu.launch body. More...
 
void mlir::populateGpuEliminateBarriersPatterns (RewritePatternSet &patterns)
 Erase barriers that do not enforce conflicting memory side effects. More...
 

Macro Definition Documentation

◆ GEN_PASS_DECL

#define GEN_PASS_DECL

Definition at line 35 of file Passes.h.

◆ GEN_PASS_REGISTRATION

#define GEN_PASS_REGISTRATION

Generate the code for registering passes.

Definition at line 173 of file Passes.h.