[Clang][OpenMP] Add the codegen support for `atomic compare`

[tianshilei1992]

Summary

This patch adds the codegen support for atomic compare in clang.

[tianshilei1992]

rebase and finish the functionality

[tianshilei1992]

remove unrelated changes

[tianshilei1992]

rebase

[tianshilei1992]

rebase

[jdoerfert]

clang/lib/CodeGen/CGStmtOpenMP.cpp

6023

leftover

clang/test/OpenMP/atomic_ast_print.cpp

1 ↗

(On Diff #405393)

this is a parser/sema test, not codegen. Add to the other patch.

llvm/include/llvm/Frontend/OpenMP/OMPIRBuilder.h

1360

move into the other patch

[tianshilei1992]

rebase and make preparation for codegen in clang

[tianshilei1992]

remove leftovers

[tianshilei1992]

add tests

[tianshilei1992]

clang/test/OpenMP/atomic_compare_codegen.cpp

1 ↗

(On Diff #406310)

The test case itself was generated from a Python script.

3 ↗

(On Diff #406310)

This test cannot be done because clang crashed. I filed a bug https://github.com/llvm/llvm-project/issues/53601.

6 ↗

(On Diff #406310)

CodeGen for -fopen-simd is wrong. It doesn't emit corresponding atomic operations. I'll fix it.

llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp

3496 ↗

(On Diff #406310)

This change will be separated.

Actually I got a question about whether we want support equality comparison for floating point values. In the patch, we don't support '<' or '>' for floating point values because we don't have corresponding atomic instruction for them. For '==', we cast them to integers and then compare the two integers. However, We all know it doesn't make too much sense to compare two floating point variables, such as x == d. The comparison of floating point variables should always do like x - d < epsilon. However, like we said before, that comparison cannot be done because of lack of instruction. So I'm not sure if we want to support floating point variables at all.

[tianshilei1992]

fix clang crash and rebase

[tianshilei1992]

rebase and remove all SIMD-ONLY0 check lines as they are useless

[tianshilei1992]

clang/test/OpenMP/atomic_compare_codegen.cpp

1990 ↗

(On Diff #406603)

I think the store here is redundant. Is it because I'm using CGF.EmitScalarExpr?

1990 ↗

(On Diff #406603)

Oh, shoot. load here, instead of store.

[tianshilei1992]

clang/lib/CodeGen/CGStmtOpenMP.cpp

6041

Using D->IgnoreImpCasts() can make sure to avoid the case that char is casted to int in binary operation. However, say, if user writes the following code:

int x;
#pragma omp atomic compare
  x = x > 1.01 ? 1.01 : x;

1.01 here will be casted to 1 by clang, and a warning will be emitted. Because we ignore the implicit cast, in Sema, it is taken as floating point value. However, we already told user that it is casted to 1, which is a little weird to emit an error then.

[tianshilei1992]

remove support for floating point values

[tianshilei1992]

ping

[ABataev]

clang/test/OpenMP/atomic_compare_codegen.cpp

6 ↗

(On Diff #406310)

The logic behind -fopenmp-simd is to emit the code only for simd specific instructions.

[tianshilei1992]

clang/test/OpenMP/atomic_compare_codegen.cpp

6 ↗

(On Diff #406310)

Right. // SIMD-ONLY0-NOT: {{__kmpc|__tgt}} is used here, same as other cases.

[tianshilei1992]

rebase

[ABataev]

LG with a nit

clang/include/clang/AST/StmtOpenMP.h

2970

Fix comment

[tianshilei1992]

In D118632#3336094, @ABataev wrote:

LG with a nit

Actually I got two question in the inline comments. Can you please take a look?

[ABataev]

In D118632#3336133, @tianshilei1992 wrote:

In D118632#3336094, @ABataev wrote:

LG with a nit

Actually I got two question in the inline comments. Can you please take a look?

Could you point me?

[tianshilei1992]

In D118632#3336145, @ABataev wrote:

In D118632#3336133, @tianshilei1992 wrote:

In D118632#3336094, @ABataev wrote:

LG with a nit

Actually I got two question in the inline comments. Can you please take a look?

Could you point me?

Sure.

clang/lib/CodeGen/CGStmtOpenMP.cpp

6041

@ABataev Here.

clang/test/OpenMP/atomic_compare_codegen.cpp

1990 ↗

(On Diff #406603)

And here. @ABataev

[ABataev]

clang/lib/CodeGen/CGStmtOpenMP.cpp

6041

Sorry, did not understand it was a question. Could you provide a bit more background, did not quite understand problem?

clang/test/OpenMP/atomic_compare_codegen.cpp

1990 ↗

(On Diff #406603)

Yes, because of EmitScalarExpr

[tianshilei1992]

clang/test/OpenMP/atomic_compare_codegen.cpp

1990 ↗

(On Diff #406603)

Can we somehow avoid the load here?

[ABataev]

clang/test/OpenMP/atomic_compare_codegen.cpp

1990 ↗

(On Diff #406603)

Do you need EmitScalarExpr?

[tianshilei1992]

clang/test/OpenMP/atomic_compare_codegen.cpp

1990 ↗

(On Diff #406603)

So it is a rvalue scalar here. What alternative do we have then?

[tianshilei1992]

fix comments

[tianshilei1992]

clang/lib/CodeGen/CGStmtOpenMP.cpp

6041

I have a better explain in another inline comment in Sema.

clang/lib/Sema/SemaOpenMP.cpp

11165

Here we do D->IgnoreImpCasts(), as shown in the code. The reason is, if we have two chars, say char a, b, and when a > b, clang also inserts an implicit cast from char to int for a and b. We want the actual type here otherwise in the IR builder, the type of D doesn't match X's, causing issues.
However, that IgnoreImpCasts here can cause another issue. Let's say we have the following code:

int x;
#pragma omp atomic compare
  x = x > 1.01 ? 1.01 : x;

clang will also insert an implicit cast from floating point value to integer for the scalar value 1.01 (corresponding to D in the code), and also emits a warning saying the floating point value has been cast to integer or something like that. Because we have the IgnoreImpCasts now, it will fail the type check because D now is of floating point type, and we will emit an error to user.
For users, it might be confusing because on one hand we emit a warning saying the floating point value has been casted, and on the other hand in Sema we tell users it is expected an integer. Users might be thinking, you already cast it to integer, why do you tell me it's a floating point value again?

[ABataev]

clang/lib/CodeGen/CGStmtOpenMP.cpp

6041

If you don't need it, do not call it. Why do you need this rvalue?

clang/lib/Sema/SemaOpenMP.cpp

11165

Yiu have the resulting type from the whole expression. You van do explicit cast to this resulting type to avoid problems, if I got it correctly

[tianshilei1992]

clang/test/OpenMP/atomic_compare_codegen.cpp

1990 ↗

(On Diff #406603)

I have to get it passed to IRBuilder, which accepts a llvm::Value *.

[ABataev]

clang/test/OpenMP/atomic_compare_codegen.cpp

1990 ↗

(On Diff #406603)

Modify irbuilder to not require it in some cases.

[tianshilei1992]

clang/test/OpenMP/atomic_compare_codegen.cpp

1990 ↗

(On Diff #406603)

You got me wrong. I need to call a function to convert a rvalue Express * to llvm::Value *. CGF.EmitScalarExpr can do it. Actually now I feel the load here can not be avoided as we want a scalar value instead of a pointer.

[tianshilei1992]

final touch before landing

clang/include/clang/AST/StmtOpenMP.h

explicit OMPAtomicDirective()
       : OMPExecutableDirective(OMPAtomicDirectiveClass, llvm::omp::OMPD_atomic,
                                SourceLocation(), SourceLocation()) {}
 
   enum DataPositionTy : size_t {
     POS_X = 0,
     POS_V,
     POS_E,
     POS_UpdateExpr,
+    POS_D,
+    POS_Cond,
   };
 
   /// Set 'x' part of the associated expression/statement.
   void setX(Expr *X) { Data->getChildren()[DataPositionTy::POS_X] = X; }
   /// Set helper expression of the form
   /// 'OpaqueValueExpr(x) binop OpaqueValueExpr(expr)' or
   /// 'OpaqueValueExpr(expr) binop OpaqueValueExpr(x)'.
   void setUpdateExpr(Expr *UE) {
     Data->getChildren()[DataPositionTy::POS_UpdateExpr] = UE;
   }
   /// Set 'v' part of the associated expression/statement.
   void setV(Expr *V) { Data->getChildren()[DataPositionTy::POS_V] = V; }
   /// Set 'expr' part of the associated expression/statement.
   void setExpr(Expr *E) { Data->getChildren()[DataPositionTy::POS_E] = E; }
+  /// Set 'd' part of the associated expression/statement.
+  void setD(Expr *D) { Data->getChildren()[DataPositionTy::POS_D] = D; }
+  /// Set conditional expression in `atomic compare`.
+  void setCond(Expr *C) { Data->getChildren()[DataPositionTy::POS_Cond] = C; }
 
 public:
   /// Creates directive with a list of \a Clauses and 'x', 'v' and 'expr'
   /// parts of the atomic construct (see Section 2.12.6, atomic Construct, for
   /// detailed description of 'x', 'v' and 'expr').
   ///
   /// \param C AST context.
   /// \param StartLoc Starting location of the directive kind.
   /// \param EndLoc Ending Location of the directive.
   /// \param Clauses List of clauses.
   /// \param AssociatedStmt Statement, associated with the directive.
   /// \param X 'x' part of the associated expression/statement.
   /// \param V 'v' part of the associated expression/statement.
   /// \param E 'expr' part of the associated expression/statement.
   /// \param UE Helper expression of the form
   /// 'OpaqueValueExpr(x) binop OpaqueValueExpr(expr)' or
   /// 'OpaqueValueExpr(expr) binop OpaqueValueExpr(x)'.
+  /// \param D 'd' part of the associated expression/statement.
+  /// \param Cond Conditional expression in `atomic compare` construct.
   /// \param IsXLHSInRHSPart true if \a UE has the first form and false if the
   /// second.
   /// \param IsPostfixUpdate true if original value of 'x' must be stored in
   /// 'v', not an updated one.
   static OMPAtomicDirective *
   Create(const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
          ArrayRef<OMPClause *> Clauses, Stmt *AssociatedStmt, Expr *X, Expr *V,
-         Expr *E, Expr *UE, bool IsXLHSInRHSPart, bool IsPostfixUpdate);
+         Expr *E, Expr *UE, Expr *D, Expr *Cond, bool IsXLHSInRHSPart,
+         bool IsPostfixUpdate);
 
   /// Creates an empty directive with the place for \a NumClauses
   /// clauses.
   ///
   /// \param C AST context.
   /// \param NumClauses Number of clauses.
   ///
   static OMPAtomicDirective *CreateEmpty(const ASTContext &C,
public:
   }
   /// Get 'expr' part of the associated expression/statement.
   Expr *getExpr() {
     return cast_or_null<Expr>(Data->getChildren()[DataPositionTy::POS_E]);
   }
   const Expr *getExpr() const {
     return cast_or_null<Expr>(Data->getChildren()[DataPositionTy::POS_E]);
   }
+  /// Get 'd' part of the associated expression/statement.
+  Expr *getD() {
+    return cast_or_null<Expr>(Data->getChildren()[DataPositionTy::POS_D]);
+  }
+  Expr *getD() const {
+    return cast_or_null<Expr>(Data->getChildren()[DataPositionTy::POS_D]);
+  }
+  /// Get
+  Expr *getCondExpr() {
+    return cast_or_null<Expr>(Data->getChildren()[DataPositionTy::POS_Cond]);
+  }
+  Expr *getCondExpr() const {
+    return cast_or_null<Expr>(Data->getChildren()[DataPositionTy::POS_Cond]);
+  }
 
   static bool classof(const Stmt *T) {
     return T->getStmtClass() == OMPAtomicDirectiveClass;
   }
 };
 
 /// This represents '#pragma omp target' directive.
 ///

clang/lib/AST/StmtOpenMP.cpp

 OMPOrderedDirective *OMPOrderedDirective::CreateEmpty(const ASTContext &C,
                                                       unsigned NumClauses,
                                                       bool IsStandalone,
                                                       EmptyShell) {
   return createEmptyDirective<OMPOrderedDirective>(C, NumClauses,
                                                    !IsStandalone);
 }
 
-OMPAtomicDirective *OMPAtomicDirective::Create(
-    const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
-    ArrayRef<OMPClause *> Clauses, Stmt *AssociatedStmt, Expr *X, Expr *V,
-    Expr *E, Expr *UE, bool IsXLHSInRHSPart, bool IsPostfixUpdate) {
+OMPAtomicDirective *
+OMPAtomicDirective::Create(const ASTContext &C, SourceLocation StartLoc,
+                           SourceLocation EndLoc, ArrayRef<OMPClause *> Clauses,
+                           Stmt *AssociatedStmt, Expr *X, Expr *V, Expr *E,
+                           Expr *UE, Expr *D, Expr *Cond, bool IsXLHSInRHSPart,
+                           bool IsPostfixUpdate) {
   auto *Dir = createDirective<OMPAtomicDirective>(
-      C, Clauses, AssociatedStmt, /*NumChildren=*/4, StartLoc, EndLoc);
+      C, Clauses, AssociatedStmt, /*NumChildren=*/6, StartLoc, EndLoc);
   Dir->setX(X);
   Dir->setV(V);
   Dir->setExpr(E);
   Dir->setUpdateExpr(UE);
+  Dir->setD(D);
+  Dir->setCond(Cond);
   Dir->IsXLHSInRHSPart = IsXLHSInRHSPart;
   Dir->IsPostfixUpdate = IsPostfixUpdate;
   return Dir;
 }
 
 OMPAtomicDirective *OMPAtomicDirective::CreateEmpty(const ASTContext &C,
                                                     unsigned NumClauses,
                                                     EmptyShell) {

clang/lib/CodeGen/CGStmtOpenMP.cpp

case llvm::AtomicOrdering::Monotonic:
       break;
     case llvm::AtomicOrdering::NotAtomic:
     case llvm::AtomicOrdering::Unordered:
       llvm_unreachable("Unexpected ordering.");
     }
   }
 }
 
+static void emitOMPAtomicCompareExpr(CodeGenFunction &CGF,
+                                     llvm::AtomicOrdering AO, const Expr *X,
+                                     const Expr *E, const Expr *D,
+                                     const Expr *CE, bool IsXBinopExpr,
+                                     SourceLocation Loc) {
+
+  llvm::OpenMPIRBuilder &OMPBuilder =
+      CGF.CGM.getOpenMPRuntime().getOMPBuilder();
+  // llvm::OpenMPIRBuilder::InsertPointTy AllocaIP(
+  //     CGF.AllocaInsertPt->getParent(), CGF.AllocaInsertPt->getIterator());
+
+  OMPAtomicCompareOp Op;
+  assert(isa<BinaryOperator>(CE) && "CE is not a BinaryOperator");
+  switch (cast<BinaryOperator>(CE)->getOpcode()) {
+  case BO_EQ:
+    Op = OMPAtomicCompareOp::EQ;
+    break;
+  case BO_LT:
+    Op = OMPAtomicCompareOp::MIN;
+    break;
+  case BO_GT:
+    Op = OMPAtomicCompareOp::MAX;
+    break;
+  default:
+    llvm_unreachable("unsupported atomic compare binary operator");
+  }
+
+  LValue XLVal = CGF.EmitLValue(X);
+  llvm::Value *XPtr = XLVal.getPointer(CGF);
+  llvm::Value *EVal = CGF.EmitScalarExpr(E);
+  llvm::Value *DVal = D ? CGF.EmitScalarExpr(D) : nullptr;
+
+  llvm::OpenMPIRBuilder::AtomicOpValue XOpVal{
+      XPtr, XPtr->getType()->getPointerElementType(),
+      X->getType().isVolatileQualified(),
+      X->getType()->hasSignedIntegerRepresentation()};
+
+  CGF.Builder.restoreIP(OMPBuilder.createAtomicCompare(
+      CGF.Builder, XOpVal, EVal, DVal, AO, Op, IsXBinopExpr));
+}
+
 static void emitOMPAtomicExpr(CodeGenFunction &CGF, OpenMPClauseKind Kind,
                               llvm::AtomicOrdering AO, bool IsPostfixUpdate,
                               const Expr *X, const Expr *V, const Expr *E,
-                              const Expr *UE, bool IsXLHSInRHSPart,
-                              SourceLocation Loc) {
+                              const Expr *UE, const Expr *D, const Expr *CE,
+                              bool IsXLHSInRHSPart, SourceLocation Loc) {
   switch (Kind) {
   case OMPC_read:
     emitOMPAtomicReadExpr(CGF, AO, X, V, Loc);
     break;
   case OMPC_write:
     emitOMPAtomicWriteExpr(CGF, AO, X, E, Loc);
     break;
   case OMPC_unknown:
   case OMPC_update:
     emitOMPAtomicUpdateExpr(CGF, AO, X, E, UE, IsXLHSInRHSPart, Loc);
     break;
   case OMPC_capture:
     emitOMPAtomicCaptureExpr(CGF, AO, IsPostfixUpdate, V, X, E, UE,
                              IsXLHSInRHSPart, Loc);
     break;
   case OMPC_compare:
-    // Do nothing here as we already emit an error.
+    emitOMPAtomicCompareExpr(CGF, AO, X, E, D, CE, IsXLHSInRHSPart, Loc);
     break;
   case OMPC_if:
   case OMPC_final:
   case OMPC_num_threads:
   case OMPC_private:
   case OMPC_firstprivate:
   case OMPC_lastprivate:
   case OMPC_reduction:
if (DefaultOrder == llvm::AtomicOrdering::Monotonic ||
         AO = llvm::AtomicOrdering::Acquire;
       }
     }
   }
 
   LexicalScope Scope(*this, S.getSourceRange());
   EmitStopPoint(S.getAssociatedStmt());
   emitOMPAtomicExpr(*this, Kind, AO, S.isPostfixUpdate(), S.getX(), S.getV(),
-                    S.getExpr(), S.getUpdateExpr(), S.isXLHSInRHSPart(),
-                    S.getBeginLoc());
+                    S.getExpr(), S.getUpdateExpr(), S.getD(), S.getCondExpr(),
+                    S.isXLHSInRHSPart(), S.getBeginLoc());
 }
 
 static void emitCommonOMPTargetDirective(CodeGenFunction &CGF,
                                          const OMPExecutableDirective &S,
                                          const RegionCodeGenTy &CodeGen) {
   assert(isOpenMPTargetExecutionDirective(S.getDirectiveKind()));
   CodeGenModule &CGM = CGF.CGM;
 

clang/lib/Sema/SemaOpenMP.cpp

if (Cond->getOpcode() != BO_EQ && Cond->getOpcode() != BO_LT &&
     ErrorInfo.Error = ErrorTy::InvalidBinaryOp;
     ErrorInfo.ErrorLoc = ErrorInfo.NoteLoc = Cond->getExprLoc();
     ErrorInfo.ErrorRange = ErrorInfo.NoteRange = Cond->getSourceRange();
     return false;
   }
 
   if (Cond->getOpcode() == BO_EQ) {
     C = Cond;
     D = CO->getTrueExpr();
     if (checkIfTwoExprsAreSame(ContextRef, X, Cond->getLHS())) {
       E = Cond->getRHS();
     } else if (checkIfTwoExprsAreSame(ContextRef, X, Cond->getRHS())) {
       E = Cond->getLHS();
     } else {
       ErrorInfo.Error = ErrorTy::InvalidComparison;
       ErrorInfo.ErrorLoc = ErrorInfo.NoteLoc = Cond->getExprLoc();
       ErrorInfo.ErrorRange = ErrorInfo.NoteRange = Cond->getSourceRange();
StmtResult Sema::ActOnOpenMPAtomicDirective(ArrayRef<OMPClause *> Clauses,
   Stmt *Body = AStmt;
   if (auto *EWC = dyn_cast<ExprWithCleanups>(Body))
     Body = EWC->getSubExpr();
 
   Expr *X = nullptr;
   Expr *V = nullptr;
   Expr *E = nullptr;
   Expr *UE = nullptr;
+  Expr *D = nullptr;
+  Expr *CE = nullptr;
   bool IsXLHSInRHSPart = false;
   bool IsPostfixUpdate = false;
   // OpenMP [2.12.6, atomic Construct]
   // In the next expressions:
   // * x and v (as applicable) are both l-value expressions with scalar type.
   // * During the execution of an atomic region, multiple syntactic
   // occurrences of x must designate the same storage location.
   // * Neither of v and expr (as applicable) may access the storage location
if (AtomicKind == OMPC_read) {
     OpenMPAtomicCompareChecker Checker(*this);
     if (!Checker.checkStmt(Body, ErrorInfo)) {
       Diag(ErrorInfo.ErrorLoc, diag::err_omp_atomic_compare)
           << ErrorInfo.ErrorRange;
       Diag(ErrorInfo.NoteLoc, diag::note_omp_atomic_compare)
           << ErrorInfo.Error << ErrorInfo.NoteRange;
       return StmtError();
     }
-    // TODO: For now we emit an error here and in emitOMPAtomicExpr we ignore
-    // code gen.
-    unsigned DiagID = Diags.getCustomDiagID(
-        DiagnosticsEngine::Error, "atomic compare is not supported for now");
-    Diag(AtomicKindLoc, DiagID);
+    X = Checker.getX();
+    E = Checker.getE();
+    D = Checker.getD();
+    CE = Checker.getCond();
+    // We reuse this bool variable to tell if it is in the form 'x ordop expr'.
+    IsXLHSInRHSPart = Checker.isXBinopExpr();
   }
 
   setFunctionHasBranchProtectedScope();
 
   return OMPAtomicDirective::Create(Context, StartLoc, EndLoc, Clauses, AStmt,
-                                    X, V, E, UE, IsXLHSInRHSPart,
+                                    X, V, E, UE, D, CE, IsXLHSInRHSPart,
                                     IsPostfixUpdate);
 }
 
 StmtResult Sema::ActOnOpenMPTargetDirective(ArrayRef<OMPClause *> Clauses,
                                             Stmt *AStmt,
                                             SourceLocation StartLoc,
                                             SourceLocation EndLoc) {
   if (!AStmt)

llvm/include/llvm/Frontend/OpenMP/OMPIRBuilder.h

public:
   /// x = e == x ? d : x; (this one is not in the spec)
   /// cond-update-stmt:
   /// if (x ordop expr) { x = expr; }
   /// if (expr ordop x) { x = expr; }
   /// if (x == e) { x = d; }
   /// if (e == x) { x = d; } (this one is not in the spec)
   ///
   /// \param Loc          The insert and source location description.
-  /// \param AllocIP      Instruction to create AllocaInst before.
   /// \param X            The target atomic pointer to be updated.
   /// \param E            The expected value ('e') for forms that use an
   ///                     equality comparison or an expression ('expr') for
   ///                     forms that use 'ordop' (logically an atomic maximum or
   ///                     minimum).
   /// \param D            The desired value for forms that use an equality
   ///                     comparison. If forms that use 'ordop', it should be
   ///                     \p nullptr.