[OpenMP][OMPIRBuilder] Add support for order(concurrent) to OMPIRBuilder for SIMD directive

[domada]

Summary

If order(concurrent) clause is specified, then the iterations of SIMD loop can be executed concurrently.

This patch adds support for LLVM IR codegen via OMPIRBuilder for SIMD loop with order(concurrent) clause. The functionality added to OMPIRBuilder is similar to the functionality implemented in CodeGenFunction::EmitOMPSimdInit.

[domada]

Patch rebased

[jdoerfert]

I think the rest looks good.

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

213

I think this belongs into llvm/include/llvm/Frontend/OpenMP/OMPConstants.h either explicitly or via the llvm/Frontend/OpenMP/OMPKinds.def logic.

[domada]

Remove custom enum for mapping order(concurrent)

[jdoerfert]

LG

clang/lib/CodeGen/CGStmtOpenMP.cpp

if (ContBlock) {
     CGF.EmitBranch(ContBlock);
     CGF.EmitBlock(ContBlock, true);
   }
 }
 
 static bool isSupportedByOpenMPIRBuilder(const OMPSimdDirective &S) {
   // Check for unsupported clauses
   for (OMPClause *C : S.clauses()) {
-    // Currently only simdlen and safelen clauses are supported
-    if (!(isa<OMPSimdlenClause>(C) || isa<OMPSafelenClause>(C)))
+    // Currently only order, simdlen and safelen clauses are supported
+    if (!(isa<OMPSimdlenClause>(C) || isa<OMPSafelenClause>(C) ||
+          isa<OMPOrderClause>(C)))
       return false;
   }
 
   // Check if we have a statement with the ordered directive.
   // Visit the statement hierarchy to find a compound statement
   // with a ordered directive in it.
   if (const auto *CanonLoop = dyn_cast<OMPCanonicalLoop>(S.getRawStmt())) {
     if (const Stmt *SyntacticalLoop = CanonLoop->getLoopStmt()) {
auto &&CodeGenIRBuilder = [this, &S, UseOMPIRBuilder](CodeGenFunction &CGF,
         llvm::ConstantInt *Safelen = nullptr;
         if (const auto *C = S.getSingleClause<OMPSafelenClause>()) {
           RValue Len =
               this->EmitAnyExpr(C->getSafelen(), AggValueSlot::ignored(),
                                 /*ignoreResult=*/true);
           auto *Val = cast<llvm::ConstantInt>(Len.getScalarVal());
           Safelen = Val;
         }
+        llvm::omp::OrderKind Order = llvm::omp::OrderKind::OMP_ORDER_unknown;
+        if (const auto *C = S.getSingleClause<OMPOrderClause>()) {
+          if (C->getKind() == OpenMPOrderClauseKind ::OMPC_ORDER_concurrent) {
+            Order = llvm::omp::OrderKind::OMP_ORDER_concurrent;
+          }
+        }
         // Add simd metadata to the collapsed loop. Do not generate
         // another loop for if clause. Support for if clause is done earlier.
-        OMPBuilder.applySimd(CLI, /*IfCond*/ nullptr, Simdlen, Safelen);
+        OMPBuilder.applySimd(CLI, /*IfCond*/ nullptr, Order, Simdlen, Safelen);
         return;
       }
     };
     {
       auto LPCRegion =
           CGOpenMPRuntime::LastprivateConditionalRAII::disable(*this, S);
       OMPLexicalScope Scope(*this, S, OMPD_unknown);
       CGM.getOpenMPRuntime().emitInlinedDirective(*this, OMPD_simd,

clang/test/OpenMP/irbuilder_safelen_order_concurrent.cpp

+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --check-globals
+// RUN: %clang_cc1 -no-opaque-pointers -fopenmp-enable-irbuilder -verify -fopenmp -fopenmp-version=50 -x c++ -triple x86_64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
+// expected-no-diagnostics
+
+struct S {
+  int a, b;
+};
+
+struct P {
+  int a, b;
+};
+
+// CHECK-LABEL: @_Z6simplePfS_Pi(
+// CHECK-NEXT:  entry:
+// CHECK-NEXT:    [[A_ADDR:%.*]] = alloca float*, align 8
+// CHECK-NEXT:    [[B_ADDR:%.*]] = alloca float*, align 8
+// CHECK-NEXT:    [[C_ADDR:%.*]] = alloca i32*, align 8
+// CHECK-NEXT:    [[S:%.*]] = alloca [[STRUCT_S:%.*]], align 4
+// CHECK-NEXT:    [[P:%.*]] = alloca %struct.S*, align 8
+// CHECK-NEXT:    [[PP:%.*]] = alloca [[STRUCT_P:%.*]], align 4
+// CHECK-NEXT:    [[I:%.*]] = alloca i32, align 4
+// CHECK-NEXT:    [[AGG_CAPTURED:%.*]] = alloca [[STRUCT_ANON:%.*]], align 8
+// CHECK-NEXT:    [[AGG_CAPTURED1:%.*]] = alloca [[STRUCT_ANON_0:%.*]], align 4
+// CHECK-NEXT:    [[DOTCOUNT_ADDR:%.*]] = alloca i32, align 4
+// CHECK-NEXT:    [[J:%.*]] = alloca i32, align 4
+// CHECK-NEXT:    [[AGG_CAPTURED8:%.*]] = alloca [[STRUCT_ANON_1:%.*]], align 8
+// CHECK-NEXT:    [[AGG_CAPTURED9:%.*]] = alloca [[STRUCT_ANON_2:%.*]], align 4
+// CHECK-NEXT:    [[DOTCOUNT_ADDR10:%.*]] = alloca i32, align 4
+// CHECK-NEXT:    store float* [[A:%.*]], float** [[A_ADDR]], align 8
+// CHECK-NEXT:    store float* [[B:%.*]], float** [[B_ADDR]], align 8
+// CHECK-NEXT:    store i32* [[C:%.*]], i32** [[C_ADDR]], align 8
+// CHECK-NEXT:    store i32 3, i32* [[I]], align 4
+// CHECK-NEXT:    [[TMP0:%.*]] = getelementptr inbounds [[STRUCT_ANON]], %struct.anon* [[AGG_CAPTURED]], i32 0, i32 0
+// CHECK-NEXT:    store i32* [[I]], i32** [[TMP0]], align 8
+// CHECK-NEXT:    [[TMP1:%.*]] = getelementptr inbounds [[STRUCT_ANON_0]], %struct.anon.0* [[AGG_CAPTURED1]], i32 0, i32 0
+// CHECK-NEXT:    [[TMP2:%.*]] = load i32, i32* [[I]], align 4
+// CHECK-NEXT:    store i32 [[TMP2]], i32* [[TMP1]], align 4
+// CHECK-NEXT:    call void @__captured_stmt(i32* [[DOTCOUNT_ADDR]], %struct.anon* [[AGG_CAPTURED]])
+// CHECK-NEXT:    [[DOTCOUNT:%.*]] = load i32, i32* [[DOTCOUNT_ADDR]], align 4
+// CHECK-NEXT:    br label [[OMP_LOOP_PREHEADER:%.*]]
+// CHECK:       omp_loop.preheader:
+// CHECK-NEXT:    br label [[OMP_LOOP_HEADER:%.*]]
+// CHECK:       omp_loop.header:
+// CHECK-NEXT:    [[OMP_LOOP_IV:%.*]] = phi i32 [ 0, [[OMP_LOOP_PREHEADER]] ], [ [[OMP_LOOP_NEXT:%.*]], [[OMP_LOOP_INC:%.*]] ]
+// CHECK-NEXT:    br label [[OMP_LOOP_COND:%.*]]
+// CHECK:       omp_loop.cond:
+// CHECK-NEXT:    [[OMP_LOOP_CMP:%.*]] = icmp ult i32 [[OMP_LOOP_IV]], [[DOTCOUNT]]
+// CHECK-NEXT:    br i1 [[OMP_LOOP_CMP]], label [[OMP_LOOP_BODY:%.*]], label [[OMP_LOOP_EXIT:%.*]]
+// CHECK:       omp_loop.body:
+// CHECK-NEXT:    call void @__captured_stmt.1(i32* [[I]], i32 [[OMP_LOOP_IV]], %struct.anon.0* [[AGG_CAPTURED1]]), !llvm.access.group [[ACC_GRP3:![0-9]+]]
+// CHECK-NEXT:    [[TMP3:%.*]] = load float*, float** [[B_ADDR]], align 8, !llvm.access.group [[ACC_GRP3]]
+// CHECK-NEXT:    [[TMP4:%.*]] = load i32, i32* [[I]], align 4, !llvm.access.group [[ACC_GRP3]]
+// CHECK-NEXT:    [[IDXPROM:%.*]] = sext i32 [[TMP4]] to i64
+// CHECK-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds float, float* [[TMP3]], i64 [[IDXPROM]]
+// CHECK-NEXT:    [[TMP5:%.*]] = load float, float* [[ARRAYIDX]], align 4, !llvm.access.group [[ACC_GRP3]]
+// CHECK-NEXT:    [[A2:%.*]] = getelementptr inbounds [[STRUCT_S]], %struct.S* [[S]], i32 0, i32 0
+// CHECK-NEXT:    [[TMP6:%.*]] = load i32, i32* [[A2]], align 4, !llvm.access.group [[ACC_GRP3]]
+// CHECK-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP6]] to float
+// CHECK-NEXT:    [[ADD:%.*]] = fadd float [[TMP5]], [[CONV]]
+// CHECK-NEXT:    [[TMP7:%.*]] = load %struct.S*, %struct.S** [[P]], align 8, !llvm.access.group [[ACC_GRP3]]
+// CHECK-NEXT:    [[A3:%.*]] = getelementptr inbounds [[STRUCT_S]], %struct.S* [[TMP7]], i32 0, i32 0
+// CHECK-NEXT:    [[TMP8:%.*]] = load i32, i32* [[A3]], align 4, !llvm.access.group [[ACC_GRP3]]
+// CHECK-NEXT:    [[CONV4:%.*]] = sitofp i32 [[TMP8]] to float
+// CHECK-NEXT:    [[ADD5:%.*]] = fadd float [[ADD]], [[CONV4]]
+// CHECK-NEXT:    [[TMP9:%.*]] = load float*, float** [[A_ADDR]], align 8, !llvm.access.group [[ACC_GRP3]]
+// CHECK-NEXT:    [[TMP10:%.*]] = load i32, i32* [[I]], align 4, !llvm.access.group [[ACC_GRP3]]
+// CHECK-NEXT:    [[IDXPROM6:%.*]] = sext i32 [[TMP10]] to i64
+// CHECK-NEXT:    [[ARRAYIDX7:%.*]] = getelementptr inbounds float, float* [[TMP9]], i64 [[IDXPROM6]]
+// CHECK-NEXT:    store float [[ADD5]], float* [[ARRAYIDX7]], align 4, !llvm.access.group [[ACC_GRP3]]
+// CHECK-NEXT:    br label [[OMP_LOOP_INC]]
+// CHECK:       omp_loop.inc:
+// CHECK-NEXT:    [[OMP_LOOP_NEXT]] = add nuw i32 [[OMP_LOOP_IV]], 1
+// CHECK-NEXT:    br label [[OMP_LOOP_HEADER]], !llvm.loop [[LOOP4:![0-9]+]]
+// CHECK:       omp_loop.exit:
+// CHECK-NEXT:    br label [[OMP_LOOP_AFTER:%.*]]
+// CHECK:       omp_loop.after:
+// CHECK-NEXT:    store i32 3, i32* [[J]], align 4
+// CHECK-NEXT:    [[TMP11:%.*]] = getelementptr inbounds [[STRUCT_ANON_1]], %struct.anon.1* [[AGG_CAPTURED8]], i32 0, i32 0
+// CHECK-NEXT:    store i32* [[J]], i32** [[TMP11]], align 8
+// CHECK-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [[STRUCT_ANON_2]], %struct.anon.2* [[AGG_CAPTURED9]], i32 0, i32 0
+// CHECK-NEXT:    [[TMP13:%.*]] = load i32, i32* [[J]], align 4
+// CHECK-NEXT:    store i32 [[TMP13]], i32* [[TMP12]], align 4
+// CHECK-NEXT:    call void @__captured_stmt.2(i32* [[DOTCOUNT_ADDR10]], %struct.anon.1* [[AGG_CAPTURED8]])
+// CHECK-NEXT:    [[DOTCOUNT11:%.*]] = load i32, i32* [[DOTCOUNT_ADDR10]], align 4
+// CHECK-NEXT:    br label [[OMP_LOOP_PREHEADER12:%.*]]
+// CHECK:       omp_loop.preheader12:
+// CHECK-NEXT:    br label [[OMP_LOOP_HEADER13:%.*]]
+// CHECK:       omp_loop.header13:
+// CHECK-NEXT:    [[OMP_LOOP_IV19:%.*]] = phi i32 [ 0, [[OMP_LOOP_PREHEADER12]] ], [ [[OMP_LOOP_NEXT21:%.*]], [[OMP_LOOP_INC16:%.*]] ]
+// CHECK-NEXT:    br label [[OMP_LOOP_COND14:%.*]]
+// CHECK:       omp_loop.cond14:
+// CHECK-NEXT:    [[OMP_LOOP_CMP20:%.*]] = icmp ult i32 [[OMP_LOOP_IV19]], [[DOTCOUNT11]]
+// CHECK-NEXT:    br i1 [[OMP_LOOP_CMP20]], label [[OMP_LOOP_BODY15:%.*]], label [[OMP_LOOP_EXIT17:%.*]]
+// CHECK:       omp_loop.body15:
+// CHECK-NEXT:    call void @__captured_stmt.3(i32* [[J]], i32 [[OMP_LOOP_IV19]], %struct.anon.2* [[AGG_CAPTURED9]]), !llvm.access.group [[ACC_GRP8:![0-9]+]]
+// CHECK-NEXT:    [[A22:%.*]] = getelementptr inbounds [[STRUCT_P]], %struct.P* [[PP]], i32 0, i32 0
+// CHECK-NEXT:    [[TMP14:%.*]] = load i32, i32* [[A22]], align 4, !llvm.access.group [[ACC_GRP8]]
+// CHECK-NEXT:    [[TMP15:%.*]] = load i32*, i32** [[C_ADDR]], align 8, !llvm.access.group [[ACC_GRP8]]
+// CHECK-NEXT:    [[TMP16:%.*]] = load i32, i32* [[J]], align 4, !llvm.access.group [[ACC_GRP8]]
+// CHECK-NEXT:    [[IDXPROM23:%.*]] = sext i32 [[TMP16]] to i64
+// CHECK-NEXT:    [[ARRAYIDX24:%.*]] = getelementptr inbounds i32, i32* [[TMP15]], i64 [[IDXPROM23]]
+// CHECK-NEXT:    store i32 [[TMP14]], i32* [[ARRAYIDX24]], align 4, !llvm.access.group [[ACC_GRP8]]
+// CHECK-NEXT:    br label [[OMP_LOOP_INC16]]
+// CHECK:       omp_loop.inc16:
+// CHECK-NEXT:    [[OMP_LOOP_NEXT21]] = add nuw i32 [[OMP_LOOP_IV19]], 1
+// CHECK-NEXT:    br label [[OMP_LOOP_HEADER13]], !llvm.loop [[LOOP9:![0-9]+]]
+// CHECK:       omp_loop.exit17:
+// CHECK-NEXT:    br label [[OMP_LOOP_AFTER18:%.*]]
+// CHECK:       omp_loop.after18:
+// CHECK-NEXT:    ret void
+//
+void simple(float *a, float *b, int *c) {
+  S s, *p;
+  P pp;
+#pragma omp simd safelen(3) order(concurrent)
+  for (int i = 3; i < 32; i += 5) {
+    a[i] = b[i] + s.a + p->a;
+  }
+
+#pragma omp simd
+  for (int j = 3; j < 32; j += 5) {
+    c[j] = pp.a;
+  }
+}
+//.
+// CHECK: attributes #0 = { mustprogress noinline nounwind optnone "frame-pointer"="none" "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-features"="+cx8,+mmx,+sse,+sse2,+x87" }
+// CHECK: attributes #1 = { noinline nounwind optnone "frame-pointer"="none" "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-features"="+cx8,+mmx,+sse,+sse2,+x87" }
+//.
+// CHECK: !0 = !{i32 1, !"wchar_size", i32 4}
+// CHECK: !1 = !{i32 7, !"openmp", i32 50}
+// CHECK: !3 = distinct !{}
+// CHECK: !4 = distinct !{!4, !5, !6, !7}
+// CHECK: !5 = !{!"llvm.loop.parallel_accesses", !3}
+// CHECK: !6 = !{!"llvm.loop.vectorize.enable", i1 true}
+// CHECK: !7 = !{!"llvm.loop.vectorize.width", i32 3}
+// CHECK: !8 = distinct !{}
+// CHECK: !9 = distinct !{!9, !10, !6}
+// CHECK: !10 = !{!"llvm.loop.parallel_accesses", !8}
+//.

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

LocationDescription(const IRBuilderBase &IRB)
         : IP(IRB.saveIP()), DL(IRB.getCurrentDebugLocation()) {}
     LocationDescription(const InsertPointTy &IP) : IP(IP) {}
     LocationDescription(const InsertPointTy &IP, const DebugLoc &DL)
         : IP(IP), DL(DL) {}
     InsertPointTy IP;
     DebugLoc DL;
   };
 
   /// Emitter methods for OpenMP directives.
   ///
   ///{
 
   /// Generator for '#omp barrier'
   ///
   /// \param Loc The location where the barrier directive was encountered.
   /// \param DK The kind of directive that caused the barrier.
   /// \param ForceSimpleCall Flag to force a simple (=non-cancellation) barrier.
public:
 
   /// Add metadata to simd-ize a loop. If IfCond is not nullptr, the loop
   /// is cloned. The metadata which prevents vectorization is added to
   /// to the cloned loop. The cloned loop is executed when ifCond is evaluated
   /// to false.
   ///
   /// \param Loop    The loop to simd-ize.
   /// \param IfCond  The value which corresponds to the if clause condition.
+  /// \param Order   The enum to map order clause
   /// \param Simdlen The Simdlen length to apply to the simd loop.
   /// \param Safelen The Safelen length to apply to the simd loop.
-  void applySimd(CanonicalLoopInfo *Loop, Value *IfCond, ConstantInt *Simdlen,
-                 ConstantInt *Safelen);
+  void applySimd(CanonicalLoopInfo *Loop, Value *IfCond, omp::OrderKind Order,
+                 ConstantInt *Simdlen, ConstantInt *Safelen);
 
   /// Generator for '#omp flush'
   ///
   /// \param Loc The location where the flush directive was encountered
   void createFlush(const LocationDescription &Loc);
 
   /// Generator for '#omp taskwait'
   ///

llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp

for (BasicBlock *Block : L->getBlocks()) {
     VMap[Block] = NewBB;
     NewBlocks.push_back(NewBB);
   }
   remapInstructionsInBlocks(NewBlocks, VMap);
   Builder.CreateBr(NewBlocks.front());
 }
 
 void OpenMPIRBuilder::applySimd(CanonicalLoopInfo *CanonicalLoop, Value *IfCond,
-                                ConstantInt *Simdlen, ConstantInt *Safelen) {
+                                OrderKind Order, ConstantInt *Simdlen,
+                                ConstantInt *Safelen) {
   LLVMContext &Ctx = Builder.getContext();
 
   Function *F = CanonicalLoop->getFunction();
 
   // TODO: We should not rely on pass manager. Currently we use pass manager
   // only for getting llvm::Loop which corresponds to given CanonicalLoopInfo
   // object. We should have a method  which returns all blocks between
   // CanonicalLoopInfo::getHeader() and CanonicalLoopInfo::getAfter()
for (BasicBlock *Block : L->getBlocks()) {
     Reachable.insert(Block);
   }
 
   SmallVector<Metadata *> LoopMDList;
 
   // In presence of finite 'safelen', it may be unsafe to mark all
   // the memory instructions parallel, because loop-carried
   // dependences of 'safelen' iterations are possible.
-  if (Safelen == nullptr) {
+  // If clause order(concurrent) is specified then the memory instructions
+  // are marked parallel even if 'safelen' is finite.
+  if ((Safelen == nullptr) || (Order == OrderKind::OMP_ORDER_concurrent)) {
     // Add access group metadata to memory-access instructions.
     MDNode *AccessGroup = MDNode::getDistinct(Ctx, {});
     for (BasicBlock *BB : Reachable)
       addSimdMetadata(BB, AccessGroup, LI);
     // TODO:  If the loop has existing parallel access metadata, have
     // to combine two lists.
     LoopMDList.push_back(MDNode::get(
         Ctx, {MDString::get(Ctx, "llvm.loop.parallel_accesses"), AccessGroup}));

llvm/unittests/Frontend/OpenMPIRBuilderTest.cpp

 }
 
 TEST_F(OpenMPIRBuilderTest, ApplySimd) {
   OpenMPIRBuilder OMPBuilder(*M);
 
   CanonicalLoopInfo *CLI = buildSingleLoopFunction(DL, OMPBuilder, 32);
 
   // Simd-ize the loop.
-  OMPBuilder.applySimd(CLI, /* IfCond */ nullptr, /* Simdlen */ nullptr,
+  OMPBuilder.applySimd(CLI, /* IfCond */ nullptr, OrderKind::OMP_ORDER_unknown,
+                       /* Simdlen */ nullptr,
                        /* Safelen */ nullptr);
 
   OMPBuilder.finalize();
   EXPECT_FALSE(verifyModule(*M, &errs()));
 
   PassBuilder PB;
   FunctionAnalysisManager FAM;
   PB.registerFunctionAnalyses(FAM);

 }
 
 TEST_F(OpenMPIRBuilderTest, ApplySimdlen) {
   OpenMPIRBuilder OMPBuilder(*M);
 
   CanonicalLoopInfo *CLI = buildSingleLoopFunction(DL, OMPBuilder, 32);
 
   // Simd-ize the loop.
-  OMPBuilder.applySimd(CLI, /* IfCond */ nullptr,
+  OMPBuilder.applySimd(CLI, /* IfCond */ nullptr, OrderKind::OMP_ORDER_unknown,
                        ConstantInt::get(Type::getInt32Ty(Ctx), 3),
                        /* Safelen */ nullptr);
 
   OMPBuilder.finalize();
   EXPECT_FALSE(verifyModule(*M, &errs()));
 
   PassBuilder PB;
   FunctionAnalysisManager FAM;
TEST_F(OpenMPIRBuilderTest, ApplySimdlen) {
   // Check for llvm.access.group metadata attached to the printf
   // function in the loop body.
   BasicBlock *LoopBody = CLI->getBody();
   EXPECT_TRUE(any_of(*LoopBody, [](Instruction &I) {
     return I.getMetadata("llvm.access.group") != nullptr;
   }));
 }
 
+TEST_F(OpenMPIRBuilderTest, ApplySafelenOrderConcurrent) {
+  OpenMPIRBuilder OMPBuilder(*M);
+
+  CanonicalLoopInfo *CLI = buildSingleLoopFunction(DL, OMPBuilder, 32);
+
+  // Simd-ize the loop.
+  OMPBuilder.applySimd(
+      CLI, /* IfCond */ nullptr, OrderKind::OMP_ORDER_concurrent,
+      /* Simdlen */ nullptr, ConstantInt::get(Type::getInt32Ty(Ctx), 3));
+
+  OMPBuilder.finalize();
+  EXPECT_FALSE(verifyModule(*M, &errs()));
+
+  PassBuilder PB;
+  FunctionAnalysisManager FAM;
+  PB.registerFunctionAnalyses(FAM);
+  LoopInfo &LI = FAM.getResult<LoopAnalysis>(*F);
+
+  const std::vector<Loop *> &TopLvl = LI.getTopLevelLoops();
+  EXPECT_EQ(TopLvl.size(), 1u);
+
+  Loop *L = TopLvl.front();
+  // Parallel metadata shoudl be attached because of presence of
+  // the order(concurrent) OpenMP clause
+  EXPECT_TRUE(findStringMetadataForLoop(L, "llvm.loop.parallel_accesses"));
+  EXPECT_TRUE(getBooleanLoopAttribute(L, "llvm.loop.vectorize.enable"));
+  EXPECT_EQ(getIntLoopAttribute(L, "llvm.loop.vectorize.width"), 3);
+
+  // Check for llvm.access.group metadata attached to the printf
+  // function in the loop body.
+  BasicBlock *LoopBody = CLI->getBody();
+  EXPECT_TRUE(any_of(*LoopBody, [](Instruction &I) {
+    return I.getMetadata("llvm.access.group") != nullptr;
+  }));
+}
+
 TEST_F(OpenMPIRBuilderTest, ApplySafelen) {
   OpenMPIRBuilder OMPBuilder(*M);
 
   CanonicalLoopInfo *CLI = buildSingleLoopFunction(DL, OMPBuilder, 32);
 
   // Simd-ize the loop.
-  OMPBuilder.applySimd(CLI, /* IfCond */ nullptr,
+  OMPBuilder.applySimd(CLI, /* IfCond */ nullptr, OrderKind::OMP_ORDER_unknown,
                        /* Simdlen */ nullptr,
                        ConstantInt::get(Type::getInt32Ty(Ctx), 3));
 
   OMPBuilder.finalize();
   EXPECT_FALSE(verifyModule(*M, &errs()));
 
   PassBuilder PB;
   FunctionAnalysisManager FAM;

 }
 
 TEST_F(OpenMPIRBuilderTest, ApplySimdlenSafelen) {
   OpenMPIRBuilder OMPBuilder(*M);
 
   CanonicalLoopInfo *CLI = buildSingleLoopFunction(DL, OMPBuilder, 32);
 
   // Simd-ize the loop.
-  OMPBuilder.applySimd(CLI, /* IfCond */ nullptr,
+  OMPBuilder.applySimd(CLI, /* IfCond */ nullptr, OrderKind::OMP_ORDER_unknown,
                        ConstantInt::get(Type::getInt32Ty(Ctx), 2),
                        ConstantInt::get(Type::getInt32Ty(Ctx), 3));
 
   OMPBuilder.finalize();
   EXPECT_FALSE(verifyModule(*M, &errs()));
 
   PassBuilder PB;
   FunctionAnalysisManager FAM;
TEST_F(OpenMPIRBuilderTest, ApplySimdLoopIf) {
   LoadInst *Load1 = Builder.CreateLoad(Alloc1->getAllocatedType(), Alloc1);
   LoadInst *Load2 = Builder.CreateLoad(Alloc2->getAllocatedType(), Alloc2);
 
   Value *IfCmp = Builder.CreateICmpNE(Load1, Load2);
 
   CanonicalLoopInfo *CLI = buildSingleLoopFunction(DL, OMPBuilder, 32);
 
   // Simd-ize the loop with if condition
-  OMPBuilder.applySimd(CLI, IfCmp, ConstantInt::get(Type::getInt32Ty(Ctx), 3),
+  OMPBuilder.applySimd(CLI, IfCmp, OrderKind::OMP_ORDER_unknown,
+                       ConstantInt::get(Type::getInt32Ty(Ctx), 3),
                        /* Safelen */ nullptr);
 
   OMPBuilder.finalize();
   EXPECT_FALSE(verifyModule(*M, &errs()));
 
   PassBuilder PB;
   FunctionAnalysisManager FAM;
   PB.registerFunctionAnalyses(FAM);

mlir/lib/Target/LLVMIR/Dialect/OpenMP/OpenMPToLLVMIRTranslation.cpp

convertOmpSimdLoop(Operation &opInst, llvm::IRBuilderBase &builder,
   llvm::ConstantInt *simdlen = nullptr;
   if (llvm::Optional<uint64_t> simdlenVar = loop.getSimdlen())
     simdlen = builder.getInt64(simdlenVar.value());
 
   llvm::ConstantInt *safelen = nullptr;
   if (llvm::Optional<uint64_t> safelenVar = loop.getSafelen())
     safelen = builder.getInt64(safelenVar.value());
 
-  ompBuilder->applySimd(loopInfo,
-                        loop.getIfExpr()
-                            ? moduleTranslation.lookupValue(loop.getIfExpr())
-                            : nullptr,
-                        simdlen, safelen);
+  ompBuilder->applySimd(
+      loopInfo,
+      loop.getIfExpr() ? moduleTranslation.lookupValue(loop.getIfExpr())
+      llvm::omp::OrderKind::OMP_ORDER_unknown, simdlen, safelen);
 
   builder.restoreIP(afterIP);
   return success();
 }
 
 /// Convert an Atomic Ordering attribute to llvm::AtomicOrdering.
 llvm::AtomicOrdering
 convertAtomicOrdering(Optional<omp::ClauseMemoryOrderKind> ao) {