98 lines · cpp
1//===-- Optimizer/Support/DataLayout.cpp ----------------------------------===//2//3// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.4// See https://llvm.org/LICENSE.txt for license information.5// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception6//7//===----------------------------------------------------------------------===//8 9#include "flang/Optimizer/Support/DataLayout.h"10#include "flang/Optimizer/Dialect/Support/FIRContext.h"11#include "flang/Optimizer/Support/FatalError.h"12#include "mlir/Dialect/DLTI/DLTI.h"13#include "mlir/Dialect/GPU/IR/GPUDialect.h"14#include "mlir/Dialect/LLVMIR/LLVMDialect.h"15#include "mlir/IR/BuiltinOps.h"16#include "mlir/Interfaces/DataLayoutInterfaces.h"17#include "mlir/Support/LLVM.h"18#include "mlir/Target/LLVMIR/Import.h"19#include "llvm/IR/DataLayout.h"20#include "llvm/MC/TargetRegistry.h"21#include "llvm/Support/TargetSelect.h"22#include "llvm/Target/TargetMachine.h"23 24namespace {25template <typename ModOpTy>26static void setDataLayout(ModOpTy mlirModule, const llvm::DataLayout &dl) {27 mlir::MLIRContext *context = mlirModule.getContext();28 mlirModule->setAttr(29 mlir::LLVM::LLVMDialect::getDataLayoutAttrName(),30 mlir::StringAttr::get(context, dl.getStringRepresentation()));31 mlir::DataLayoutSpecInterface dlSpec = mlir::translateDataLayout(dl, context);32 mlirModule->setAttr(mlir::DLTIDialect::kDataLayoutAttrName, dlSpec);33}34 35template <typename ModOpTy>36static void setDataLayoutFromAttributes(ModOpTy mlirModule,37 bool allowDefaultLayout) {38 if (mlirModule.getDataLayoutSpec())39 return; // Already set.40 if (auto dataLayoutString =41 mlirModule->template getAttrOfType<mlir::StringAttr>(42 mlir::LLVM::LLVMDialect::getDataLayoutAttrName())) {43 llvm::DataLayout llvmDataLayout(dataLayoutString);44 fir::support::setMLIRDataLayout(mlirModule, llvmDataLayout);45 return;46 }47 if (!allowDefaultLayout)48 return;49 llvm::DataLayout llvmDataLayout("");50 fir::support::setMLIRDataLayout(mlirModule, llvmDataLayout);51}52 53template <typename ModOpTy>54static std::optional<mlir::DataLayout>55getOrSetDataLayout(ModOpTy mlirModule, bool allowDefaultLayout) {56 if (!mlirModule.getDataLayoutSpec())57 fir::support::setMLIRDataLayoutFromAttributes(mlirModule,58 allowDefaultLayout);59 if (!mlirModule.getDataLayoutSpec() &&60 !mlir::isa<mlir::gpu::GPUModuleOp>(mlirModule))61 return std::nullopt;62 return mlir::DataLayout(mlirModule);63}64 65} // namespace66 67void fir::support::setMLIRDataLayout(mlir::ModuleOp mlirModule,68 const llvm::DataLayout &dl) {69 setDataLayout(mlirModule, dl);70}71 72void fir::support::setMLIRDataLayout(mlir::gpu::GPUModuleOp mlirModule,73 const llvm::DataLayout &dl) {74 setDataLayout(mlirModule, dl);75}76 77void fir::support::setMLIRDataLayoutFromAttributes(mlir::ModuleOp mlirModule,78 bool allowDefaultLayout) {79 setDataLayoutFromAttributes(mlirModule, allowDefaultLayout);80}81 82void fir::support::setMLIRDataLayoutFromAttributes(83 mlir::gpu::GPUModuleOp mlirModule, bool allowDefaultLayout) {84 setDataLayoutFromAttributes(mlirModule, allowDefaultLayout);85}86 87std::optional<mlir::DataLayout>88fir::support::getOrSetMLIRDataLayout(mlir::ModuleOp mlirModule,89 bool allowDefaultLayout) {90 return getOrSetDataLayout(mlirModule, allowDefaultLayout);91}92 93std::optional<mlir::DataLayout>94fir::support::getOrSetMLIRDataLayout(mlir::gpu::GPUModuleOp mlirModule,95 bool allowDefaultLayout) {96 return getOrSetDataLayout(mlirModule, allowDefaultLayout);97}98