Lines Matching +defs:mlir +defs:mode
14 #include "mlir/Conversion/GPUToNVVM/GPUToNVVMPass.h"
16 #include "mlir/Conversion/ArithToLLVM/ArithToLLVM.h"
17 #include "mlir/Conversion/ControlFlowToLLVM/ControlFlowToLLVM.h"
18 #include "mlir/Conversion/ConvertToLLVM/ToLLVMInterface.h"
19 #include "mlir/Conversion/FuncToLLVM/ConvertFuncToLLVM.h"
20 #include "mlir/Conversion/GPUCommon/GPUCommonPass.h"
21 #include "mlir/Conversion/GPUToNVVM/GPUToNVVM.h"
22 #include "mlir/Conversion/LLVMCommon/ConversionTarget.h"
23 #include "mlir/Conversion/LLVMCommon/LoweringOptions.h"
24 #include "mlir/Conversion/LLVMCommon/TypeConverter.h"
25 #include "mlir/Conversion/MemRefToLLVM/MemRefToLLVM.h"
26 #include "mlir/Conversion/VectorToLLVM/ConvertVectorToLLVM.h"
27 #include "mlir/Dialect/ControlFlow/IR/ControlFlow.h"
28 #include "mlir/Dialect/ControlFlow/IR/ControlFlowOps.h"
29 #include "mlir/Dialect/Func/IR/FuncOps.h"
30 #include "mlir/Dialect/GPU/IR/GPUDialect.h"
31 #include "mlir/Dialect/GPU/Transforms/Passes.h"
32 #include "mlir/Dialect/LLVMIR/NVVMDialect.h"
33 #include "mlir/Dialect/Math/IR/Math.h"
34 #include "mlir/Dialect/MemRef/IR/MemRef.h"
35 #include "mlir/Dialect/NVGPU/IR/NVGPUDialect.h"
36 #include "mlir/Transforms/DialectConversion.h"
37 #include "mlir/Transforms/GreedyPatternRewriteDriver.h"
44 namespace mlir {
46 #include "mlir/Conversion/Passes.h.inc"
47 } // namespace mlir
49 using namespace mlir;
53 /// Convert gpu dialect shfl mode enum to the equivalent nvvm one.
54 static NVVM::ShflKind convertShflKind(gpu::ShuffleMode mode) {
55 switch (mode) {
65 llvm_unreachable("unknown shuffle mode");
69 convertReduxKind(gpu::AllReduceOperation mode) {
70 switch (mode) {
120 std::optional<NVVM::ReduxKind> mode = convertReduxKind(op.getOp());
121 if (!mode.has_value())
123 op, "unsupported reduction mode for redux");
130 mode.value(), offset);
398 void mlir::configureGpuToNVVMConversionLegality(ConversionTarget &target) {
400 target.addLegalDialect<::mlir::LLVM::LLVMDialect>();
401 target.addLegalDialect<::mlir::NVVM::NVVMDialect>();
413 void mlir::configureGpuToNVVMTypeConverter(LLVMTypeConverter &converter) {
465 void mlir::populateGpuSubgroupReduceOpLoweringPattern(
470 void mlir::populateGpuToNVVMConversionPatterns(
620 void mlir::NVVM::registerConvertGpuToNVVMInterface(DialectRegistry ®istry) {