Merge branch 'users/xlauko/cir-cast-op-format' into users/xlauko/cir-opt-test-verify-roundtrip

Author: xlauko

clang/docs/OpenMPSupport.rst

@@ -580,9 +580,12 @@ implementation.
 | need_device_addr modifier for adjust_args clause            | :part:`partial`           | :none:`unclaimed`         | Parsing/Sema: https://github.com/llvm/llvm-project/pull/143442           |
 |                                                             |                           |                           |               https://github.com/llvm/llvm-project/pull/149586           |
 +-------------------------------------------------------------+---------------------------+---------------------------+--------------------------------------------------------------------------+
-| Prescriptive num_threads                                    | :part:`In Progress`       | :none:`unclaimed`         | ro-i                                                                     |
+| Prescriptive num_threads                                    | :good:`done`              | :none:`unclaimed`         |  https://github.com/llvm/llvm-project/pull/160659                        |
+|                                                             |                           |                           |  https://github.com/llvm/llvm-project/pull/146403                        |
+|                                                             |                           |                           |  https://github.com/llvm/llvm-project/pull/146404                        |
+|                                                             |                           |                           |  https://github.com/llvm/llvm-project/pull/146405                        |
 +-------------------------------------------------------------+---------------------------+---------------------------+--------------------------------------------------------------------------+
-| Message and severity clauses                                | :part:`In Progress`       | :none:`unclaimed`         | ro-i                                                                     |
+| Message and severity clauses                                | :good:`done`              | :none:`unclaimed`         |  https://github.com/llvm/llvm-project/pull/146093                        |
 +-------------------------------------------------------------+---------------------------+---------------------------+--------------------------------------------------------------------------+
 | Local clause on declare target                              | :part:`In Progress`       | :none:`unclaimed`         |                                                                          |
 +-------------------------------------------------------------+---------------------------+---------------------------+--------------------------------------------------------------------------+

clang/include/clang/AST/HLSLResource.h

@@ -0,0 +1,78 @@
+//===- HLSLResource.h - Routines for HLSL resources and bindings ----------===//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+//
+// This file provides shared routines to help analyze HLSL resources and
+// theirs bindings during Sema and CodeGen.
+//
+//===----------------------------------------------------------------------===//
+
+#ifndef LLVM_CLANG_AST_HLSLRESOURCE_H
+#define LLVM_CLANG_AST_HLSLRESOURCE_H
+
+#include "clang/AST/ASTContext.h"
+#include "clang/AST/Attr.h"
+#include "clang/AST/Attrs.inc"
+#include "clang/AST/DeclBase.h"
+#include "clang/Basic/TargetInfo.h"
+#include "clang/Support/Compiler.h"
+#include "llvm/Support/raw_ostream.h"
+
+namespace clang {
+
+class HLSLResourceBindingAttr;
+class HLSLRVkBindingAttr;
+
+namespace hlsl {
+
+struct ResourceBindingAttrs {
+  HLSLResourceBindingAttr *RegBinding;
+  HLSLVkBindingAttr *VkBinding;
+
+  ResourceBindingAttrs(const Decl *D) {
+    RegBinding = D->getAttr<HLSLResourceBindingAttr>();
+    bool IsSpirv = D->getASTContext().getTargetInfo().getTriple().isSPIRV();
+    VkBinding = IsSpirv ? D->getAttr<HLSLVkBindingAttr>() : nullptr;
+  }
+
+  bool hasBinding() const { return RegBinding || VkBinding; }
+  bool isExplicit() const {
+    return (RegBinding && RegBinding->hasRegisterSlot()) || VkBinding;
+  }
+
+  unsigned getSlot() const {
+    assert(isExplicit() && "no explicit binding");
+    if (VkBinding)
+      return VkBinding->getBinding();
+    if (RegBinding && RegBinding->hasRegisterSlot())
+      return RegBinding->getSlotNumber();
+    llvm_unreachable("no explicit binding");
+  }
+
+  unsigned getSpace() const {
+    if (VkBinding)
+      return VkBinding->getSet();
+    if (RegBinding)
+      return RegBinding->getSpaceNumber();
+    return 0;
+  }
+
+  bool hasImplicitOrderID() const {
+    return RegBinding && RegBinding->hasImplicitBindingOrderID();
+  }
+
+  unsigned getImplicitOrderID() const {
+    assert(hasImplicitOrderID());
+    return RegBinding->getImplicitBindingOrderID();
+  }
+};
+
+} // namespace hlsl
+
+} // namespace clang
+
+#endif // LLVM_CLANG_AST_HLSLRESOURCE_H

clang/include/clang/CIR/MissingFeatures.h

@@ -248,7 +248,6 @@ struct MissingFeatures {
   static bool metaDataNode() { return false; }
   static bool moduleNameHash() { return false; }
   static bool msabi() { return false; }
-  static bool needsGlobalCtorDtor() { return false; }
   static bool nrvo() { return false; }
   static bool objCBlocks() { return false; }
   static bool objCGC() { return false; }

clang/include/clang/Driver/CommonArgs.h

@@ -304,6 +304,11 @@ std::string complexRangeKindToStr(LangOptions::ComplexRangeKind Range);
 // Render a frontend option corresponding to ComplexRangeKind.
 std::string renderComplexRangeOption(LangOptions::ComplexRangeKind Range);
 
+// Set the complex range and output a warning as needed.
+void setComplexRange(const Driver &D, StringRef NewOpt,
+                     LangOptions::ComplexRangeKind NewRange, StringRef &LastOpt,
+                     LangOptions::ComplexRangeKind &Range);
+
 } // end namespace tools
 } // end namespace driver
 } // end namespace clang

clang/lib/CIR/CodeGen/CIRGenCXX.cpp

@@ -15,10 +15,89 @@
 
 #include "clang/AST/GlobalDecl.h"
 #include "clang/CIR/MissingFeatures.h"
+#include "llvm/Support/SaveAndRestore.h"
 
 using namespace clang;
 using namespace clang::CIRGen;
 
+static void emitDeclInit(CIRGenFunction &cgf, const VarDecl *varDecl,
+                         cir::GlobalOp globalOp) {
+  assert((varDecl->hasGlobalStorage() ||
+          (varDecl->hasLocalStorage() &&
+           cgf.getContext().getLangOpts().OpenCLCPlusPlus)) &&
+         "VarDecl must have global or local (in the case of OpenCL) storage!");
+  assert(!varDecl->getType()->isReferenceType() &&
+         "Should not call emitDeclInit on a reference!");
+
+  CIRGenBuilderTy &builder = cgf.getBuilder();
+
+  // Set up the ctor region.
+  mlir::OpBuilder::InsertionGuard guard(builder);
+  mlir::Block *block = builder.createBlock(&globalOp.getCtorRegion());
+  CIRGenFunction::LexicalScope lexScope{cgf, globalOp.getLoc(),
+                                        builder.getInsertionBlock()};
+  lexScope.setAsGlobalInit();
+  builder.setInsertionPointToStart(block);
+
+  Address declAddr(cgf.cgm.getAddrOfGlobalVar(varDecl),
+                   cgf.cgm.getASTContext().getDeclAlign(varDecl));
+
+  QualType type = varDecl->getType();
+  LValue lv = cgf.makeAddrLValue(declAddr, type);
+
+  const Expr *init = varDecl->getInit();
+  switch (CIRGenFunction::getEvaluationKind(type)) {
+  case cir::TEK_Scalar:
+    assert(!cir::MissingFeatures::objCGC());
+    cgf.emitScalarInit(init, cgf.getLoc(varDecl->getLocation()), lv, false);
+    break;
+  case cir::TEK_Complex:
+    cgf.cgm.errorNYI(varDecl->getSourceRange(), "complex global initializer");
+    break;
+  case cir::TEK_Aggregate:
+    assert(!cir::MissingFeatures::aggValueSlotGC());
+    cgf.emitAggExpr(init,
+                    AggValueSlot::forLValue(lv, AggValueSlot::IsDestructed,
+                                            AggValueSlot::IsNotAliased,
+                                            AggValueSlot::DoesNotOverlap));
+    break;
+  }
+
+  // Finish the ctor region.
+  builder.setInsertionPointToEnd(block);
+  cir::YieldOp::create(builder, globalOp.getLoc());
+}
+
+static void emitDeclDestroy(CIRGenFunction &cgf, const VarDecl *vd,
+                            cir::GlobalOp addr) {
+  // Honor __attribute__((no_destroy)) and bail instead of attempting
+  // to emit a reference to a possibly nonexistent destructor, which
+  // in turn can cause a crash. This will result in a global constructor
+  // that isn't balanced out by a destructor call as intended by the
+  // attribute. This also checks for -fno-c++-static-destructors and
+  // bails even if the attribute is not present.
+  QualType::DestructionKind dtorKind = vd->needsDestruction(cgf.getContext());
+
+  // FIXME:  __attribute__((cleanup)) ?
+
+  switch (dtorKind) {
+  case QualType::DK_none:
+    return;
+
+  case QualType::DK_cxx_destructor:
+    break;
+
+  case QualType::DK_objc_strong_lifetime:
+  case QualType::DK_objc_weak_lifetime:
+  case QualType::DK_nontrivial_c_struct:
+    // We don't care about releasing objects during process teardown.
+    assert(!vd->getTLSKind() && "should have rejected this");
+    return;
+  }
+
+  cgf.cgm.errorNYI(vd->getSourceRange(), "global with destructor");
+}
+
 cir::FuncOp CIRGenModule::codegenCXXStructor(GlobalDecl gd) {
   const CIRGenFunctionInfo &fnInfo =
       getTypes().arrangeCXXStructorDeclaration(gd);
@@ -38,3 +117,63 @@ cir::FuncOp CIRGenModule::codegenCXXStructor(GlobalDecl gd) {
   assert(!cir::MissingFeatures::opFuncAttributesForDefinition());
   return fn;
 }
+
+// Global variables requiring non-trivial initialization are handled
+// differently in CIR than in classic codegen. Classic codegen emits
+// a global init function (__cxx_global_var_init) and inserts
+// initialization for each global there. In CIR, we attach a ctor
+// region to the global variable and insert the initialization code
+// into the ctor region. This will be moved into the
+// __cxx_global_var_init function during the LoweringPrepare pass.
+void CIRGenModule::emitCXXGlobalVarDeclInit(const VarDecl *varDecl,
+                                            cir::GlobalOp addr,
+                                            bool performInit) {
+  QualType ty = varDecl->getType();
+
+  // TODO: handle address space
+  // The address space of a static local variable (addr) may be different
+  // from the address space of the "this" argument of the constructor. In that
+  // case, we need an addrspacecast before calling the constructor.
+  //
+  // struct StructWithCtor {
+  //   __device__ StructWithCtor() {...}
+  // };
+  // __device__ void foo() {
+  //   __shared__ StructWithCtor s;
+  //   ...
+  // }
+  //
+  // For example, in the above CUDA code, the static local variable s has a
+  // "shared" address space qualifier, but the constructor of StructWithCtor
+  // expects "this" in the "generic" address space.
+  assert(!cir::MissingFeatures::addressSpace());
+
+  // Create a CIRGenFunction to emit the initializer. While this isn't a true
+  // function, the handling works the same way.
+  CIRGenFunction cgf{*this, builder, true};
+  llvm::SaveAndRestore<CIRGenFunction *> savedCGF(curCGF, &cgf);
+  curCGF->curFn = addr;
+
+  CIRGenFunction::SourceLocRAIIObject fnLoc{cgf,
+                                            getLoc(varDecl->getLocation())};
+
+  assert(!cir::MissingFeatures::astVarDeclInterface());
+
+  if (!ty->isReferenceType()) {
+    assert(!cir::MissingFeatures::openMP());
+
+    bool needsDtor = varDecl->needsDestruction(getASTContext()) ==
+                     QualType::DK_cxx_destructor;
+    // PerformInit, constant store invariant / destroy handled below.
+    if (performInit)
+      emitDeclInit(cgf, varDecl, addr);
+
+    if (varDecl->getType().isConstantStorage(getASTContext(), true, !needsDtor))
+      errorNYI(varDecl->getSourceRange(), "global with constant storage");
+    else
+      emitDeclDestroy(cgf, varDecl, addr);
+    return;
+  }
+
+  errorNYI(varDecl->getSourceRange(), "global with reference type");
+}

clang/lib/CIR/CodeGen/CIRGenDeclCXX.cpp

@@ -0,0 +1,28 @@
+//===----------------------------------------------------------------------===//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+//
+// This contains code dealing with code generation of C++ declarations
+//
+//===----------------------------------------------------------------------===//
+
+#include "CIRGenModule.h"
+#include "clang/AST/Attr.h"
+#include "clang/Basic/LangOptions.h"
+
+using namespace clang;
+using namespace clang::CIRGen;
+
+void CIRGenModule::emitCXXGlobalVarDeclInitFunc(const VarDecl *vd,
+                                                cir::GlobalOp addr,
+                                                bool performInit) {
+  assert(!cir::MissingFeatures::cudaSupport());
+
+  assert(!cir::MissingFeatures::deferredCXXGlobalInit());
+
+  emitCXXGlobalVarDeclInit(vd, addr, performInit);
+}

clang/lib/CIR/CodeGen/CIRGenExprConstant.cpp

@@ -775,7 +775,9 @@ class ConstExprEmitter
   }
 
   mlir::Attribute VisitCXXConstructExpr(CXXConstructExpr *e, QualType ty) {
-    cgm.errorNYI(e->getBeginLoc(), "ConstExprEmitter::VisitCXXConstructExpr");
+    if (!e->getConstructor()->isTrivial())
+      return nullptr;
+    cgm.errorNYI(e->getBeginLoc(), "trivial constructor const handling");
     return {};
   }
 

clang/lib/CIR/CodeGen/CIRGenFunction.cpp

@@ -342,10 +342,12 @@ void CIRGenFunction::LexicalScope::cleanup() {
 cir::ReturnOp CIRGenFunction::LexicalScope::emitReturn(mlir::Location loc) {
   CIRGenBuilderTy &builder = cgf.getBuilder();
 
-  if (!cgf.curFn.getFunctionType().hasVoidReturn()) {
+  auto fn = dyn_cast<cir::FuncOp>(cgf.curFn);
+  assert(fn && "emitReturn from non-function");
+  if (!fn.getFunctionType().hasVoidReturn()) {
     // Load the value from `__retval` and return it via the `cir.return` op.
     auto value = builder.create<cir::LoadOp>(
-        loc, cgf.curFn.getFunctionType().getReturnType(), *cgf.fnRetAlloca);
+        loc, fn.getFunctionType().getReturnType(), *cgf.fnRetAlloca);
     return builder.create<cir::ReturnOp>(loc,
                                          llvm::ArrayRef(value.getResult()));
   }
@@ -459,7 +461,9 @@ void CIRGenFunction::startFunction(GlobalDecl gd, QualType returnType,
     const auto *md = cast<CXXMethodDecl>(d);
     if (md->getParent()->isLambda() && md->getOverloadedOperator() == OO_Call) {
       // We're in a lambda.
-      curFn.setLambda(true);
+      auto fn = dyn_cast<cir::FuncOp>(curFn);
+      assert(fn && "lambda in non-function region");
+      fn.setLambda(true);
 
       // Figure out the captures.
       md->getParent()->getCaptureFields(lambdaCaptureFields,

clang/lib/CIR/CodeGen/CIRGenFunction.h

@@ -98,8 +98,10 @@ class CIRGenFunction : public CIRGenTypeCache {
   /// This is the inner-most code context, which includes blocks.
   const clang::Decl *curCodeDecl = nullptr;
 
-  /// The function for which code is currently being generated.
-  cir::FuncOp curFn;
+  /// The current function or global initializer that is generated code for.
+  /// This is usually a cir::FuncOp, but it can also be a cir::GlobalOp for
+  /// global initializers.
+  mlir::Operation *curFn = nullptr;
 
   using DeclMapTy = llvm::DenseMap<const clang::Decl *, Address>;
   /// This keeps track of the CIR allocas or globals for local C
@@ -116,7 +118,11 @@ class CIRGenFunction : public CIRGenTypeCache {
   CIRGenModule &getCIRGenModule() { return cgm; }
   const CIRGenModule &getCIRGenModule() const { return cgm; }
 
-  mlir::Block *getCurFunctionEntryBlock() { return &curFn.getRegion().front(); }
+  mlir::Block *getCurFunctionEntryBlock() {
+    // We currently assume this isn't called for a global initializer.
+    auto fn = mlir::cast<cir::FuncOp>(curFn);
+    return &fn.getRegion().front();
+  }
 
   /// Sanitizers enabled for this function.
   clang::SanitizerSet sanOpts;