Pipelines.h

//===-- Pipelines.h -- FIR pass pipelines -----------------------*- C++ -*-===//
//
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
// See https://llvm.org/LICENSE.txt for license information.
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
//
//===----------------------------------------------------------------------===//

 
#ifndef FORTRAN_OPTIMIZER_PASSES_PIPELINES_H
#define FORTRAN_OPTIMIZER_PASSES_PIPELINES_H
 
#include "flang/Optimizer/CodeGen/CodeGen.h"
#include "flang/Optimizer/HLFIR/Passes.h"
#include "flang/Optimizer/OpenMP/Passes.h"
#include "flang/Optimizer/Passes/CommandLineOpts.h"
#include "flang/Optimizer/Transforms/Passes.h"
#include "flang/Tools/CrossToolHelpers.h"
#include "mlir/Conversion/ReconcileUnrealizedCasts/ReconcileUnrealizedCasts.h"
#include "mlir/Conversion/SCFToControlFlow/SCFToControlFlow.h"
#include "mlir/Dialect/GPU/IR/GPUDialect.h"
#include "mlir/Dialect/LLVMIR/LLVMAttrs.h"
#include "mlir/Pass/PassManager.h"
#include "mlir/Transforms/GreedyPatternRewriteDriver.h"
#include "mlir/Transforms/Passes.h"
#include "llvm/Frontend/Debug/Options.h"
#include "llvm/Passes/OptimizationLevel.h"
#include "llvm/Support/CommandLine.h"
 
namespace fir {
 
class GlobalOp;
 
using PassConstructor = std::unique_ptr<mlir::Pass>();
 
template <typename F, typename OP>
void addNestedPassToOps(mlir::PassManager &pm, F ctor) {
  pm.addNestedPass<OP>(ctor());
}
 
template <typename F, typename OP, typename... OPS,
          typename = std::enable_if_t<sizeof...(OPS) != 0>>
void addNestedPassToOps(mlir::PassManager &pm, F ctor) {
  addNestedPassToOps<F, OP>(pm, ctor);
  addNestedPassToOps<F, OPS...>(pm, ctor);
}

template <typename F>
void addPassConditionally(mlir::PassManager &pm, llvm::cl::opt<bool> &disabled,
                          F ctor) {
  if (!disabled)
    pm.addPass(ctor());
}
 
template <typename OP, typename F>
void addNestedPassConditionally(mlir::PassManager &pm,
                                llvm::cl::opt<bool> &disabled, F ctor) {
  if (!disabled)
    pm.addNestedPass<OP>(ctor());
}
 
template <typename F>
void addNestedPassToAllTopLevelOperations(mlir::PassManager &pm, F ctor) {
  addNestedPassToOps<F, mlir::func::FuncOp, mlir::omp::DeclareMapperOp,
                     mlir::omp::DeclareReductionOp, mlir::omp::PrivateClauseOp,
                     fir::GlobalOp>(pm, ctor);
}
 
template <typename F>
void addNestedPassToAllTopLevelOperationsConditionally(
    mlir::PassManager &pm, llvm::cl::opt<bool> &disabled, F ctor) {
  if (!disabled)
    addNestedPassToAllTopLevelOperations<F>(pm, ctor);
}
 
template <typename F>
void addPassToGPUModuleOperations(mlir::PassManager &pm, F ctor) {
  mlir::OpPassManager &nestPM = pm.nest<mlir::gpu::GPUModuleOp>();
  nestPM.addNestedPass<mlir::func::FuncOp>(ctor());
  nestPM.addNestedPass<mlir::gpu::GPUFuncOp>(ctor());
}

void addCanonicalizerPassWithoutRegionSimplification(mlir::OpPassManager &pm);
 
void addCfgConversionPass(mlir::PassManager &pm,
                          const MLIRToLLVMPassPipelineConfig &config);
 
void addMemoryAllocationOpt(mlir::PassManager &pm);
 
void addAllocationPlacement(mlir::PassManager &pm, bool stackArrays);
 
void addCodeGenRewritePass(mlir::PassManager &pm, bool preserveDeclare);
 
void addTargetRewritePass(mlir::PassManager &pm);
 
mlir::LLVM::DIEmissionKind
getEmissionKind(llvm::codegenoptions::DebugInfoKind kind);
 
void addBoxedProcedurePass(mlir::PassManager &pm,
                           bool enableSafeTrampoline = false);
 
void addExternalNameConversionPass(mlir::PassManager &pm,
                                   bool appendUnderscore = true);
 
void addCompilerGeneratedNamesConversionPass(mlir::PassManager &pm);
 
void addDebugInfoPass(mlir::PassManager &pm,
                      const MLIRToLLVMPassPipelineConfig &config,
                      llvm::StringRef inputFilename);

FIRToLLVMPassOptions
getFIRToLLVMPassOptions(const MLIRToLLVMPassPipelineConfig &config);
 
void addFIRToLLVMPass(mlir::PassManager &pm,
                      const MLIRToLLVMPassPipelineConfig &config);
 
void addLLVMDialectToLLVMPass(mlir::PassManager &pm, llvm::raw_ostream &output);

void registerDefaultInlinerPass(MLIRToLLVMPassPipelineConfig &config);

void registerFlangPipelinePasses();

void createDefaultFIROptimizerPassPipeline(mlir::PassManager &pm,
                                           MLIRToLLVMPassPipelineConfig &pc);

enum class EnableOpenMP { None, Simd, Full };

void createHLFIRToFIRPassPipeline(mlir::PassManager &pm,
                                  EnableOpenMP enableOpenMP,
                                  const MLIRToLLVMPassPipelineConfig &config);
 
struct OpenMPFIRPassPipelineOpts {
  bool isTargetDevice;

  Fortran::frontend::CodeGenOptions::DoConcurrentMappingKind
      doConcurrentMappingKind;
};

void createOpenMPFIRPassPipeline(mlir::PassManager &pm,
                                 OpenMPFIRPassPipelineOpts opts);
 
#if !defined(FLANG_EXCLUDE_CODEGEN)
void createDebugPasses(mlir::PassManager &pm,
                       const MLIRToLLVMPassPipelineConfig &config,
                       llvm::StringRef inputFilename);
 
void createDefaultFIRCodeGenPassPipeline(mlir::PassManager &pm,
                                         MLIRToLLVMPassPipelineConfig config,
                                         llvm::StringRef inputFilename = {});

void createMLIRToLLVMPassPipeline(mlir::PassManager &pm,
                                  MLIRToLLVMPassPipelineConfig &config,
                                  llvm::StringRef inputFilename = {});
#undef FLANG_EXCLUDE_CODEGEN
#endif
 
} // namespace fir
 
#endif // FORTRAN_OPTIMIZER_PASSES_PIPELINES_H