MLIR 24.0.0git
Passes.h File Reference
#include "mlir/Dialect/GPU/IR/GPUDialect.h"
#include "mlir/Dialect/GPU/Utils/GPUUtils.h"
#include "mlir/IR/PatternMatch.h"
#include "mlir/Pass/Pass.h"
#include "mlir/Dialect/GPU/Transforms/Passes.h.inc"

Go to the source code of this file.

Classes

struct  mlir::GpuKernelOutliningPassOptions
struct  mlir::GpuMapParallelLoopsPassOptions
struct  mlir::GpuModuleToBinaryPassOptions

Namespaces

namespace  llvm
 The OpAsmOpInterface, see OpAsmInterface.td for more details.
namespace  mlir
 Include the generated interface declarations.
namespace  mlir::func
namespace  mlir::gpu

Macros

#define GEN_PASS_DECL
#define GEN_PASS_DECL_GPUASYNCREGIONPASS
#define GEN_PASS_DECL_GPUDECOMPOSEMEMREFSPASS
#define GEN_PASS_DECL_GPUELIMINATEBARRIERS
#define GEN_PASS_DECL_GPUKERNELOUTLININGPASS
#define GEN_PASS_DECL_GPULAUNCHSINKINDEXCOMPUTATIONSPASS
#define GEN_PASS_DECL_GPUMAPPARALLELLOOPSPASS
#define GEN_PASS_DECL_GPUMODULETOBINARYPASS
#define GEN_PASS_REGISTRATION
 Generate the code for registering passes.
#define GEN_PASS_REGISTRATION_GPUASYNCREGIONPASS
#define GEN_PASS_REGISTRATION_GPUDECOMPOSEMEMREFSPASS
#define GEN_PASS_REGISTRATION_GPUELIMINATEBARRIERS
#define GEN_PASS_REGISTRATION_GPUKERNELOUTLININGPASS
#define GEN_PASS_REGISTRATION_GPULAUNCHSINKINDEXCOMPUTATIONSPASS
#define GEN_PASS_REGISTRATION_GPUMAPPARALLELLOOPSPASS
#define GEN_PASS_REGISTRATION_GPUMODULETOBINARYPASS

Functions

std::unique_ptr<::mlir::Pass > mlir::createGpuAsyncRegionPass ()
std::unique_ptr<::mlir::Pass > mlir::createGpuDecomposeMemrefsPass ()
std::unique_ptr<::mlir::Pass > mlir::createGpuEliminateBarriers ()
std::unique_ptr<::mlir::Pass > mlir::createGpuKernelOutliningPass ()
std::unique_ptr<::mlir::Pass > mlir::createGpuKernelOutliningPass (GpuKernelOutliningPassOptions options)
std::unique_ptr<::mlir::Pass > mlir::createGpuLaunchSinkIndexComputationsPass ()
std::unique_ptr<::mlir::Pass > mlir::createGpuMapParallelLoopsPass ()
std::unique_ptr<::mlir::Pass > mlir::createGpuMapParallelLoopsPass (GpuMapParallelLoopsPassOptions options)
std::unique_ptr<::mlir::Pass > mlir::createGpuModuleToBinaryPass ()
std::unique_ptr<::mlir::Pass > mlir::createGpuModuleToBinaryPass (GpuModuleToBinaryPassOptions options)
void mlir::populateGpuGlobalIdPatterns (RewritePatternSet &patterns)
 Collect a set of patterns to rewrite GlobalIdOp op within the GPU dialect.
void mlir::populateGpuSubgroupIdPatterns (RewritePatternSet &patterns)
 Collect a set of patterns to rewrite SubgroupIdOp op within the GPU dialect.
void mlir::populateGpuShufflePatterns (RewritePatternSet &patterns)
 Collect a set of patterns to rewrite shuffle ops within the GPU dialect.
void mlir::populateGpuAllReducePatterns (RewritePatternSet &patterns)
 Collect a set of patterns to rewrite all-reduce ops within the GPU dialect.
void mlir::populateGpuBreakDownSubgroupReducePatterns (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.
void mlir::populateGpuLowerSubgroupReduceToShufflePatterns (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.
void mlir::populateGpuLowerClusteredSubgroupReduceToShufflePatterns (RewritePatternSet &patterns, unsigned subgroupSize, unsigned shuffleBitwidth=32, PatternBenefit benefit=1)
 Disjoint counterpart of populateGpuLowerSubgroupReduceToShufflePatterns that only matches gpu.subgroup_reduce ops with a cluster_size.
void mlir::populateGpuRewritePatterns (RewritePatternSet &patterns)
 Collect all patterns to rewrite ops within the GPU dialect.
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.
void mlir::populateGpuDecomposeMemrefsPatterns (RewritePatternSet &patterns)
 Collect a set of patterns to decompose memrefs ops.
void mlir::populateGpuEliminateBarriersPatterns (RewritePatternSet &patterns)
 Erase barriers that do not enforce conflicting memory side effects.
void mlir::registerGpuAsyncRegionPass ()
void mlir::registerGpuAsyncRegionPassPass ()
void mlir::registerGpuDecomposeMemrefsPass ()
void mlir::registerGpuDecomposeMemrefsPassPass ()
void mlir::registerGpuEliminateBarriers ()
void mlir::registerGpuEliminateBarriersPass ()
void mlir::registerGpuKernelOutliningPass ()
void mlir::registerGpuKernelOutliningPassPass ()
void mlir::registerGpuLaunchSinkIndexComputationsPass ()
void mlir::registerGpuLaunchSinkIndexComputationsPassPass ()
void mlir::registerGpuMapParallelLoopsPass ()
void mlir::registerGpuMapParallelLoopsPassPass ()
void mlir::registerGpuModuleToBinaryPass ()
void mlir::registerGpuModuleToBinaryPassPass ()
void mlir::registerGPUPasses ()

Macro Definition Documentation

◆ GEN_PASS_DECL

#define GEN_PASS_DECL

Definition at line 34 of file Passes.h.

◆ GEN_PASS_DECL_GPUASYNCREGIONPASS

#define GEN_PASS_DECL_GPUASYNCREGIONPASS

Definition at line 6 of file Passes.h.

◆ GEN_PASS_DECL_GPUDECOMPOSEMEMREFSPASS

#define GEN_PASS_DECL_GPUDECOMPOSEMEMREFSPASS

Definition at line 7 of file Passes.h.

◆ GEN_PASS_DECL_GPUELIMINATEBARRIERS

#define GEN_PASS_DECL_GPUELIMINATEBARRIERS

Definition at line 8 of file Passes.h.

◆ GEN_PASS_DECL_GPUKERNELOUTLININGPASS

#define GEN_PASS_DECL_GPUKERNELOUTLININGPASS

Definition at line 9 of file Passes.h.

◆ GEN_PASS_DECL_GPULAUNCHSINKINDEXCOMPUTATIONSPASS

#define GEN_PASS_DECL_GPULAUNCHSINKINDEXCOMPUTATIONSPASS

Definition at line 10 of file Passes.h.

◆ GEN_PASS_DECL_GPUMAPPARALLELLOOPSPASS

#define GEN_PASS_DECL_GPUMAPPARALLELLOOPSPASS

Definition at line 11 of file Passes.h.

◆ GEN_PASS_DECL_GPUMODULETOBINARYPASS

#define GEN_PASS_DECL_GPUMODULETOBINARYPASS

Definition at line 12 of file Passes.h.

◆ GEN_PASS_REGISTRATION

#define GEN_PASS_REGISTRATION

Generate the code for registering passes.

Definition at line 101 of file Passes.h.

◆ GEN_PASS_REGISTRATION_GPUASYNCREGIONPASS

#define GEN_PASS_REGISTRATION_GPUASYNCREGIONPASS

Definition at line 617 of file Passes.h.

◆ GEN_PASS_REGISTRATION_GPUDECOMPOSEMEMREFSPASS

#define GEN_PASS_REGISTRATION_GPUDECOMPOSEMEMREFSPASS

Definition at line 618 of file Passes.h.

◆ GEN_PASS_REGISTRATION_GPUELIMINATEBARRIERS

#define GEN_PASS_REGISTRATION_GPUELIMINATEBARRIERS

Definition at line 619 of file Passes.h.

◆ GEN_PASS_REGISTRATION_GPUKERNELOUTLININGPASS

#define GEN_PASS_REGISTRATION_GPUKERNELOUTLININGPASS

Definition at line 620 of file Passes.h.

◆ GEN_PASS_REGISTRATION_GPULAUNCHSINKINDEXCOMPUTATIONSPASS

#define GEN_PASS_REGISTRATION_GPULAUNCHSINKINDEXCOMPUTATIONSPASS

Definition at line 621 of file Passes.h.

◆ GEN_PASS_REGISTRATION_GPUMAPPARALLELLOOPSPASS

#define GEN_PASS_REGISTRATION_GPUMAPPARALLELLOOPSPASS

Definition at line 622 of file Passes.h.

◆ GEN_PASS_REGISTRATION_GPUMODULETOBINARYPASS

#define GEN_PASS_REGISTRATION_GPUMODULETOBINARYPASS

Definition at line 623 of file Passes.h.