[OpenMPIRBuilder] Implement static-chunked workshare-loop schedules.

[Meinersbur]

Summary

Add applyStaticChunkedWorkshareLoop method implementing static schedule when chunk-size is specified. Unlike a static schedule without chunk-size (where chunk-size is chosen by the runtime such that each thread receives one chunk), we need two nested loops: one for looping over the iterations of a chunk, and a second for looping over all chunks assigned to the threads.

This patch includes the following related changes:

• Adapt applyWorkshareLoop to triage between the schedule types, now possible since all schedules have been implemented. The default schedule is assumed to be non-chunked static, as without OpenMPIRBuilder.
• Remove the chunk parameter from applyStaticWorkshareLoop, it is ignored by the runtime. Change the value for the value passed to the init function to 0, as without OpenMPIRBuilder.
• Refactor CanonicalLoopInfo::setTripCount and CanonicalLoopInfo::mapIndVar as used by both, applyStaticWorkshareLoop and applyStaticChunkedWorkshareLoop.
• Enable Clang to use the OpenMPIRBuilder in the presence of the schedule clause.

[Meinersbur]

• Simicolon after LLVM_FALLTHROUGH
• Changes requested by @peixin

[Meinersbur]

• Remove unused code
• Fix mlir test

[peixin]

@Meinersbur Please rebase on main. The function "getPreheader()" was moved into OMPIRBuilder.h.

[Meinersbur]

• Rebase
• clang-format

[Meinersbur]

In D114413#3154936, @peixin wrote:

@Meinersbur Please rebase on main. The function "getPreheader()" was moved into OMPIRBuilder.h.

I rebased, but I am not sure what you are referring to. getPreheader() always was in OMPIRBuilder.h

[peixin]

In D114413#3159095, @Meinersbur wrote:

In D114413#3154936, @peixin wrote:

@Meinersbur Please rebase on main. The function "getPreheader()" was moved into OMPIRBuilder.h.

I rebased, but I am not sure what you are referring to. getPreheader() always was in OMPIRBuilder.h

getPreheader() was in OMPIRBuilder.cpp before you rebase in your last update here. That's why I let you rebase since I failed to git apply your last patch in main branch. It's not important now. Please forget about that.

[Meinersbur]

In D114413#3160044, @peixin wrote:

In D114413#3159095, @Meinersbur wrote:

In D114413#3154936, @peixin wrote:

getPreheader() was in OMPIRBuilder.cpp before you rebase in your last update here. That's why I let you rebase since I failed to git apply your last patch in main branch. It's not important now. Please forget about that.

D114368 (which this patch depends on) moves getPreheder() to the .cpp files (because it has become more than a simple getter)

[peixin]

In D114413#3166979, @Meinersbur wrote:

In D114413#3160044, @peixin wrote:

In D114413#3159095, @Meinersbur wrote:

In D114413#3154936, @peixin wrote:

getPreheader() was in OMPIRBuilder.cpp before you rebase in your last update here. That's why I let you rebase since I failed to git apply your last patch in main branch. It's not important now. Please forget about that.

D114368 (which this patch depends on) moves getPreheder() to the .cpp files (because it has become more than a simple getter)

Thanks a lot. Now I get it.

[peixin]

Can you check the following example by applying this patch on fir-dev?

program main
  integer :: i, N = 10
  real :: x = 0

  !$omp do schedule(static, 2)
  do i = 3, N
    x = x + i
  end do
  !$omp end do

  print *, x
end

Test running result:

$ gfortran test.f90 -fopenmp && ./a.out
   52.0000000    
$ bbc -fopenmp -emit-fir test.f90
$ tco test.mlir -o test.ll
$ clang++ -lFortran_main -lFortranRuntime -lFortranDecimal -lomp -o a.out test.ll
$ ./a.out
 7.

When you change "schedule(static, 2)" into "schedule(static, 1)", the running result is 3.0.

llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp

1676

Can you move "Value *ThreadNum = getOrCreateThreadID(SrcLoc);" after "Builder.CreateStore(One, PStride);" in order that the "kmpc_global_thread_num" call is right before the "kmpc_static_init" call to keep consistence with others?

1765

Please remove the space between "ScheduleKind" and "OMP_SCHEDULE_Default"? Also for the following switch cases.

[peixin]

@Meinersbur Here is the c++ code test. Without the chunk size specified, the running result using OMPIRBuilder is correct.

#include<iostream>

int main() {
  int i, N = 10;
  float x = 0.0;

  #pragma omp for schedule(static, 2)
  for(i = 3; i <= N; i++) {
    x = x + i;
  }

  std::cout << "x = " << x << std::endl;

  return 0;
}

$ clang++ chunk-1.cpp -fopenmp -fopenmp-enable-irbuilder && ./a.out
x = 7
$ clang++ chunk-1.cpp -fopenmp && ./a.out
x = 52

[peixin]

BTW, please rebase on main. There is one conflict about function getOrCreateSrcLocStr.

[Meinersbur]

• Rebase

Still to do: fix bug report

[Meinersbur]

libomp implemented a special case that I was not considering: When not in a parallel environment, it would ignore the chunksize and execute the entire iteration space as a single chunk. That is, we cannot assume that the chunk loop iteration count is chunksize must compute it from the lower and upper bound returned from the runtime call.

Thanks for the reproducer!

[peixin]

Can you rebase this? I cannot apply this patch on current main branch.

[Meinersbur]

Rebase

[peixin]

When I investigated the edge cases you mentioned in D116292. Found one unsupported case as follows

#include <climits>
#include <iostream>
using namespace std;

void func(unsigned long long lb, unsigned long long ub, unsigned long long step) {
  unsigned long long i;
  #pragma omp for schedule(static, 1)
  for (i = lb; i > ub; i -= step) {
    cout << i << endl;
  }
}

int main() {
  unsigned long long lb, ub, step;
  lb = ULLONG_MAX;
  ub = ULLONG_MAX / 10;
  step = ULLONG_MAX / 10;
  cout << "lb: " << lb << endl;
  cout << "ub: " << ub << endl;
  cout << "step: " << step << endl;

  func(lb, ub, step);

  cout << endl;
  return 0;
}

$ clang++ temp.cpp -fopenmp && ./a.out
lb: 18446744073709551615
ub: 1844674407370955161
step: 1844674407370955161
18446744073709551615
16602069666338596454
14757395258967641293
12912720851596686132
11068046444225730971
9223372036854775810
7378697629483820649
5534023222112865488
3689348814741910327
1844674407370955166
$ clang++ temp.cpp -fopenmp -fopenmp-enable-irbuilder
clang-14: /home/qpx/compilers/llvm-community/static-chunk-codegen/llvm-project/llvm/lib/IR/Instructions.cpp:506: void llvm::CallInst::init(llvm::FunctionType*, llvm::Value*, llvm::ArrayRef<llvm::Value*>, llvm::ArrayRef<llvm::OperandBundleDefT<llvm::Value*> >, const llvm::Twine&): Assertion `(i >= FTy->getNumParams() || FTy->getParamType(i) == Args[i]->getType()) && "Calling a function with a bad signature!"' failed.
PLEASE submit a bug report to https://github.com/llvm/llvm-project/issues/ and include the crash backtrace, preprocessed source, and associated run script.
Stack dump:

This is also for schedule(static).

[Meinersbur]

• Fix chunksize/tripcount type mismatch
• Test with different iv sizes

[Meinersbur]

@peixin Thanks for testing edge cases. You hit multiple issues:

1. Chunksize was i32, combining it with a i64 induction variable caused an error. Fixed the latest update of this patch.
2. OpenMPIRBuilder currently doesn't work really with exceptions. See D115216 for a start of a discussion with #pragma omp parallel. Support for irregular exits (exceptions, cancellation, destructors) out of CanonicalLoopInfo is what I was working on recently. Use -fno-exceptions to work around.
3. There is an off-by-one error that I already fixed in my development branch. Upstream patch here: D118542

Result with these fixes for me is:

lb: 18446744073709551615
ub: 1844674407370955161
step: 1844674407370955161
18446744073709551615
16602069666338596454
14757395258967641293
12912720851596686132
11068046444225730971
9223372036854775810
7378697629483820649
5534023222112865488
3689348814741910327
1844674407370955166

Note that this does not involve __kmpc_doacross_init code in libomp you pointed-to in D116292. This uses __kmpc_for_static_init calls of which there are 4 variants for (signed/unsigned x 32/64 bits). To do __kmpc_doacross_init correctly, it would also need at least have variants for signed/unsigned (or one working internally with signed 128 bits).

[peixin]

Thanks for the fix. The fix of off-by-one issue looks ok to me. Will continue reviewing other parts in one week due to the Spring Festival in China.

[peixin]

Except for three nits. LGTM.

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

1445

Nit: integer -> 64-bit integer?

llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp

1676

This comment is not addressed.

1765

The extra space is not removed.

[Meinersbur]

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

1445

not necessarily, we do not require a specific integer size. For instance, __kmpc_for_static_init_4u takes a 32-bit integer. It is up to the applyXYZ function to zext/trunc it when necessary.

[peixin]

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

1445

Got it. Thanks.

clang/lib/CodeGen/CGStmtOpenMP.cpp

HasLastprivates = CGF.EmitOMPWorksharingLoop(S, S.getEnsureUpperBound(),
                                                  emitDispatchForLoopBounds);
   }
   return HasLastprivates;
 }
 
 static bool isSupportedByOpenMPIRBuilder(const OMPForDirective &S) {
   if (S.hasCancel())
     return false;
-  for (OMPClause *C : S.clauses())
-    if (!isa<OMPNowaitClause>(C))
-      return false;
+  for (OMPClause *C : S.clauses()) {
+    if (isa<OMPNowaitClause>(C))
+      continue;
+
+    if (auto *SC = dyn_cast<OMPScheduleClause>(C)) {
+      if (SC->getFirstScheduleModifier() != OMPC_SCHEDULE_unknown)
+        return false;
+      if (SC->getSecondScheduleModifier() != OMPC_SCHEDULE_unknown)
+        return false;
+      switch (SC->getScheduleKind()) {
+      case OMPC_SCHEDULE_auto:
+      case OMPC_SCHEDULE_dynamic:
+      case OMPC_SCHEDULE_runtime:
+      case OMPC_SCHEDULE_guided:
+      case OMPC_SCHEDULE_static:
+        continue;
+      case OMPC_SCHEDULE_unknown:
+        return false;
+      }
+    }
+
+  }
 
   return true;
 }
 
+static llvm::omp::ScheduleKind
+convertClauseKindToSchedKind(OpenMPScheduleClauseKind ScheduleClauseKind) {
+  switch (ScheduleClauseKind) {
+  case OMPC_SCHEDULE_unknown:
+    return llvm::omp::OMP_SCHEDULE_Default;
+  case OMPC_SCHEDULE_auto:
+    return llvm::omp::OMP_SCHEDULE_Auto;
+  case OMPC_SCHEDULE_dynamic:
+    return llvm::omp::OMP_SCHEDULE_Dynamic;
+  case OMPC_SCHEDULE_guided:
+    return llvm::omp::OMP_SCHEDULE_Guided;
+  case OMPC_SCHEDULE_runtime:
+    return llvm::omp::OMP_SCHEDULE_Runtime;
+  case OMPC_SCHEDULE_static:
+    return llvm::omp::OMP_SCHEDULE_Static;
+  }
+  llvm_unreachable("Unhandled schedule kind");
+}
+
 void CodeGenFunction::EmitOMPForDirective(const OMPForDirective &S) {
   bool HasLastprivates = false;
   bool UseOMPIRBuilder =
       CGM.getLangOpts().OpenMPIRBuilder && isSupportedByOpenMPIRBuilder(S);
   auto &&CodeGen = [this, &S, &HasLastprivates,
                     UseOMPIRBuilder](CodeGenFunction &CGF, PrePostActionTy &) {
     // Use the OpenMPIRBuilder if enabled.
     if (UseOMPIRBuilder) {
+      bool NeedsBarrier = !S.getSingleClause<OMPNowaitClause>();
+
+      llvm::omp::ScheduleKind SchedKind = llvm::omp::OMP_SCHEDULE_Default;
+      llvm::Value *ChunkSize = nullptr;
+      if (auto *SchedClause = S.getSingleClause<OMPScheduleClause>()) {
+        SchedKind =
+            convertClauseKindToSchedKind(SchedClause->getScheduleKind());
+        if (const Expr *ChunkSizeExpr = SchedClause->getChunkSize())
+          ChunkSize = EmitScalarExpr(ChunkSizeExpr);
+      }
+
       // Emit the associated statement and get its loop representation.
       const Stmt *Inner = S.getRawStmt();
       llvm::CanonicalLoopInfo *CLI =
           EmitOMPCollapsedCanonicalLoopNest(Inner, 1);
 
-      bool NeedsBarrier = !S.getSingleClause<OMPNowaitClause>();
       llvm::OpenMPIRBuilder &OMPBuilder =
           CGM.getOpenMPRuntime().getOMPBuilder();
       llvm::OpenMPIRBuilder::InsertPointTy AllocaIP(
           AllocaInsertPt->getParent(), AllocaInsertPt->getIterator());
       OMPBuilder.applyWorkshareLoop(Builder.getCurrentDebugLocation(), CLI,
-                                    AllocaIP, NeedsBarrier);
+                                    AllocaIP, NeedsBarrier, SchedKind,
+                                    ChunkSize);
       return;
     }
 
     HasLastprivates = emitWorksharingDirective(CGF, S, S.hasCancel());
   };
   {
     auto LPCRegion =
         CGOpenMPRuntime::LastprivateConditionalRAII::disable(*this, S);

clang/test/OpenMP/cancel_codegen.cpp

 // CHECK3-NEXT:    br label [[OMP_PAR_EXIT_SPLIT:%.*]]
 // CHECK3:       omp.par.exit.split:
 // CHECK3-NEXT:    br label [[OMP_SECTION_LOOP_PREHEADER:%.*]]
 // CHECK3:       omp_section_loop.preheader:
 // CHECK3-NEXT:    store i32 0, i32* [[P_LOWERBOUND]], align 4
 // CHECK3-NEXT:    store i32 0, i32* [[P_UPPERBOUND]], align 4
 // CHECK3-NEXT:    store i32 1, i32* [[P_STRIDE]], align 4
 // CHECK3-NEXT:    [[OMP_GLOBAL_THREAD_NUM11:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1]])
-// CHECK3-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM11]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 1)
+// CHECK3-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM11]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 0)
 // CHECK3-NEXT:    [[TMP0:%.*]] = load i32, i32* [[P_LOWERBOUND]], align 4
 // CHECK3-NEXT:    [[TMP1:%.*]] = load i32, i32* [[P_UPPERBOUND]], align 4
 // CHECK3-NEXT:    [[TMP2:%.*]] = sub i32 [[TMP1]], [[TMP0]]
 // CHECK3-NEXT:    [[TMP3:%.*]] = add i32 [[TMP2]], 1
 // CHECK3-NEXT:    br label [[OMP_SECTION_LOOP_HEADER:%.*]]
 // CHECK3:       omp_section_loop.header:
 // CHECK3-NEXT:    [[OMP_SECTION_LOOP_IV:%.*]] = phi i32 [ 0, [[OMP_SECTION_LOOP_PREHEADER]] ], [ [[OMP_SECTION_LOOP_NEXT:%.*]], [[OMP_SECTION_LOOP_INC:%.*]] ]
 // CHECK3-NEXT:    br label [[OMP_SECTION_LOOP_COND:%.*]]

 // CHECK3-NEXT:    br label [[OMP_SECTIONS_END:%.*]]
 // CHECK3:       omp_sections.end:
 // CHECK3-NEXT:    br label [[OMP_SECTION_LOOP_PREHEADER13:%.*]]
 // CHECK3:       omp_section_loop.preheader13:
 // CHECK3-NEXT:    store i32 0, i32* [[P_LOWERBOUND28]], align 4
 // CHECK3-NEXT:    store i32 1, i32* [[P_UPPERBOUND29]], align 4
 // CHECK3-NEXT:    store i32 1, i32* [[P_STRIDE30]], align 4
 // CHECK3-NEXT:    [[OMP_GLOBAL_THREAD_NUM31:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1]])
-// CHECK3-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM31]], i32 34, i32* [[P_LASTITER27]], i32* [[P_LOWERBOUND28]], i32* [[P_UPPERBOUND29]], i32* [[P_STRIDE30]], i32 1, i32 1)
+// CHECK3-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM31]], i32 34, i32* [[P_LASTITER27]], i32* [[P_LOWERBOUND28]], i32* [[P_UPPERBOUND29]], i32* [[P_STRIDE30]], i32 1, i32 0)
 // CHECK3-NEXT:    [[TMP7:%.*]] = load i32, i32* [[P_LOWERBOUND28]], align 4
 // CHECK3-NEXT:    [[TMP8:%.*]] = load i32, i32* [[P_UPPERBOUND29]], align 4
 // CHECK3-NEXT:    [[TMP9:%.*]] = sub i32 [[TMP8]], [[TMP7]]
 // CHECK3-NEXT:    [[TMP10:%.*]] = add i32 [[TMP9]], 1
 // CHECK3-NEXT:    br label [[OMP_SECTION_LOOP_HEADER14:%.*]]
 // CHECK3:       omp_section_loop.header14:
 // CHECK3-NEXT:    [[OMP_SECTION_LOOP_IV20:%.*]] = phi i32 [ 0, [[OMP_SECTION_LOOP_PREHEADER13]] ], [ [[OMP_SECTION_LOOP_NEXT22:%.*]], [[OMP_SECTION_LOOP_INC17:%.*]] ]
 // CHECK3-NEXT:    br label [[OMP_SECTION_LOOP_COND15:%.*]]

 // CHECK4-NEXT:    br label [[OMP_PAR_EXIT_SPLIT:%.*]]
 // CHECK4:       omp.par.exit.split:
 // CHECK4-NEXT:    br label [[OMP_SECTION_LOOP_PREHEADER:%.*]]
 // CHECK4:       omp_section_loop.preheader:
 // CHECK4-NEXT:    store i32 0, i32* [[P_LOWERBOUND]], align 4
 // CHECK4-NEXT:    store i32 0, i32* [[P_UPPERBOUND]], align 4
 // CHECK4-NEXT:    store i32 1, i32* [[P_STRIDE]], align 4
 // CHECK4-NEXT:    [[OMP_GLOBAL_THREAD_NUM11:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1]])
-// CHECK4-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM11]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 1)
+// CHECK4-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM11]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 0)
 // CHECK4-NEXT:    [[TMP0:%.*]] = load i32, i32* [[P_LOWERBOUND]], align 4
 // CHECK4-NEXT:    [[TMP1:%.*]] = load i32, i32* [[P_UPPERBOUND]], align 4
 // CHECK4-NEXT:    [[TMP2:%.*]] = sub i32 [[TMP1]], [[TMP0]]
 // CHECK4-NEXT:    [[TMP3:%.*]] = add i32 [[TMP2]], 1
 // CHECK4-NEXT:    br label [[OMP_SECTION_LOOP_HEADER:%.*]]
 // CHECK4:       omp_section_loop.header:
 // CHECK4-NEXT:    [[OMP_SECTION_LOOP_IV:%.*]] = phi i32 [ 0, [[OMP_SECTION_LOOP_PREHEADER]] ], [ [[OMP_SECTION_LOOP_NEXT:%.*]], [[OMP_SECTION_LOOP_INC:%.*]] ]
 // CHECK4-NEXT:    br label [[OMP_SECTION_LOOP_COND:%.*]]

 // CHECK4-NEXT:    br label [[OMP_SECTIONS_END:%.*]]
 // CHECK4:       omp_sections.end:
 // CHECK4-NEXT:    br label [[OMP_SECTION_LOOP_PREHEADER13:%.*]]
 // CHECK4:       omp_section_loop.preheader13:
 // CHECK4-NEXT:    store i32 0, i32* [[P_LOWERBOUND28]], align 4
 // CHECK4-NEXT:    store i32 1, i32* [[P_UPPERBOUND29]], align 4
 // CHECK4-NEXT:    store i32 1, i32* [[P_STRIDE30]], align 4
 // CHECK4-NEXT:    [[OMP_GLOBAL_THREAD_NUM31:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1]])
-// CHECK4-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM31]], i32 34, i32* [[P_LASTITER27]], i32* [[P_LOWERBOUND28]], i32* [[P_UPPERBOUND29]], i32* [[P_STRIDE30]], i32 1, i32 1)
+// CHECK4-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM31]], i32 34, i32* [[P_LASTITER27]], i32* [[P_LOWERBOUND28]], i32* [[P_UPPERBOUND29]], i32* [[P_STRIDE30]], i32 1, i32 0)
 // CHECK4-NEXT:    [[TMP7:%.*]] = load i32, i32* [[P_LOWERBOUND28]], align 4
 // CHECK4-NEXT:    [[TMP8:%.*]] = load i32, i32* [[P_UPPERBOUND29]], align 4
 // CHECK4-NEXT:    [[TMP9:%.*]] = sub i32 [[TMP8]], [[TMP7]]
 // CHECK4-NEXT:    [[TMP10:%.*]] = add i32 [[TMP9]], 1
 // CHECK4-NEXT:    br label [[OMP_SECTION_LOOP_HEADER14:%.*]]
 // CHECK4:       omp_section_loop.header14:
 // CHECK4-NEXT:    [[OMP_SECTION_LOOP_IV20:%.*]] = phi i32 [ 0, [[OMP_SECTION_LOOP_PREHEADER13]] ], [ [[OMP_SECTION_LOOP_NEXT22:%.*]], [[OMP_SECTION_LOOP_INC17:%.*]] ]
 // CHECK4-NEXT:    br label [[OMP_SECTION_LOOP_COND15:%.*]]

 // CHECK9-NEXT:    br label [[OMP_PAR_EXIT_SPLIT:%.*]]
 // CHECK9:       omp.par.exit.split:
 // CHECK9-NEXT:    br label [[OMP_SECTION_LOOP_PREHEADER:%.*]]
 // CHECK9:       omp_section_loop.preheader:
 // CHECK9-NEXT:    store i32 0, i32* [[P_LOWERBOUND]], align 4
 // CHECK9-NEXT:    store i32 0, i32* [[P_UPPERBOUND]], align 4
 // CHECK9-NEXT:    store i32 1, i32* [[P_STRIDE]], align 4
 // CHECK9-NEXT:    [[OMP_GLOBAL_THREAD_NUM11:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1]])
-// CHECK9-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM11]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 1)
+// CHECK9-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM11]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 0)
 // CHECK9-NEXT:    [[TMP0:%.*]] = load i32, i32* [[P_LOWERBOUND]], align 4
 // CHECK9-NEXT:    [[TMP1:%.*]] = load i32, i32* [[P_UPPERBOUND]], align 4
 // CHECK9-NEXT:    [[TMP2:%.*]] = sub i32 [[TMP1]], [[TMP0]]
 // CHECK9-NEXT:    [[TMP3:%.*]] = add i32 [[TMP2]], 1
 // CHECK9-NEXT:    br label [[OMP_SECTION_LOOP_HEADER:%.*]]
 // CHECK9:       omp_section_loop.header:
 // CHECK9-NEXT:    [[OMP_SECTION_LOOP_IV:%.*]] = phi i32 [ 0, [[OMP_SECTION_LOOP_PREHEADER]] ], [ [[OMP_SECTION_LOOP_NEXT:%.*]], [[OMP_SECTION_LOOP_INC:%.*]] ]
 // CHECK9-NEXT:    br label [[OMP_SECTION_LOOP_COND:%.*]]

 // CHECK9-NEXT:    br label [[OMP_SECTIONS_END:%.*]]
 // CHECK9:       omp_sections.end:
 // CHECK9-NEXT:    br label [[OMP_SECTION_LOOP_PREHEADER13:%.*]]
 // CHECK9:       omp_section_loop.preheader13:
 // CHECK9-NEXT:    store i32 0, i32* [[P_LOWERBOUND28]], align 4
 // CHECK9-NEXT:    store i32 1, i32* [[P_UPPERBOUND29]], align 4
 // CHECK9-NEXT:    store i32 1, i32* [[P_STRIDE30]], align 4
 // CHECK9-NEXT:    [[OMP_GLOBAL_THREAD_NUM31:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1]])
-// CHECK9-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM31]], i32 34, i32* [[P_LASTITER27]], i32* [[P_LOWERBOUND28]], i32* [[P_UPPERBOUND29]], i32* [[P_STRIDE30]], i32 1, i32 1)
+// CHECK9-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM31]], i32 34, i32* [[P_LASTITER27]], i32* [[P_LOWERBOUND28]], i32* [[P_UPPERBOUND29]], i32* [[P_STRIDE30]], i32 1, i32 0)
 // CHECK9-NEXT:    [[TMP7:%.*]] = load i32, i32* [[P_LOWERBOUND28]], align 4
 // CHECK9-NEXT:    [[TMP8:%.*]] = load i32, i32* [[P_UPPERBOUND29]], align 4
 // CHECK9-NEXT:    [[TMP9:%.*]] = sub i32 [[TMP8]], [[TMP7]]
 // CHECK9-NEXT:    [[TMP10:%.*]] = add i32 [[TMP9]], 1
 // CHECK9-NEXT:    br label [[OMP_SECTION_LOOP_HEADER14:%.*]]
 // CHECK9:       omp_section_loop.header14:
 // CHECK9-NEXT:    [[OMP_SECTION_LOOP_IV20:%.*]] = phi i32 [ 0, [[OMP_SECTION_LOOP_PREHEADER13]] ], [ [[OMP_SECTION_LOOP_NEXT22:%.*]], [[OMP_SECTION_LOOP_INC17:%.*]] ]
 // CHECK9-NEXT:    br label [[OMP_SECTION_LOOP_COND15:%.*]]

 // CHECK10-NEXT:    br label [[OMP_PAR_EXIT_SPLIT:%.*]]
 // CHECK10:       omp.par.exit.split:
 // CHECK10-NEXT:    br label [[OMP_SECTION_LOOP_PREHEADER:%.*]]
 // CHECK10:       omp_section_loop.preheader:
 // CHECK10-NEXT:    store i32 0, i32* [[P_LOWERBOUND]], align 4
 // CHECK10-NEXT:    store i32 0, i32* [[P_UPPERBOUND]], align 4
 // CHECK10-NEXT:    store i32 1, i32* [[P_STRIDE]], align 4
 // CHECK10-NEXT:    [[OMP_GLOBAL_THREAD_NUM11:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1]])
-// CHECK10-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM11]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 1)
+// CHECK10-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM11]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 0)
 // CHECK10-NEXT:    [[TMP0:%.*]] = load i32, i32* [[P_LOWERBOUND]], align 4
 // CHECK10-NEXT:    [[TMP1:%.*]] = load i32, i32* [[P_UPPERBOUND]], align 4
 // CHECK10-NEXT:    [[TMP2:%.*]] = sub i32 [[TMP1]], [[TMP0]]
 // CHECK10-NEXT:    [[TMP3:%.*]] = add i32 [[TMP2]], 1
 // CHECK10-NEXT:    br label [[OMP_SECTION_LOOP_HEADER:%.*]]
 // CHECK10:       omp_section_loop.header:
 // CHECK10-NEXT:    [[OMP_SECTION_LOOP_IV:%.*]] = phi i32 [ 0, [[OMP_SECTION_LOOP_PREHEADER]] ], [ [[OMP_SECTION_LOOP_NEXT:%.*]], [[OMP_SECTION_LOOP_INC:%.*]] ]
 // CHECK10-NEXT:    br label [[OMP_SECTION_LOOP_COND:%.*]]

 // CHECK10-NEXT:    br label [[OMP_SECTIONS_END:%.*]]
 // CHECK10:       omp_sections.end:
 // CHECK10-NEXT:    br label [[OMP_SECTION_LOOP_PREHEADER13:%.*]]
 // CHECK10:       omp_section_loop.preheader13:
 // CHECK10-NEXT:    store i32 0, i32* [[P_LOWERBOUND28]], align 4
 // CHECK10-NEXT:    store i32 1, i32* [[P_UPPERBOUND29]], align 4
 // CHECK10-NEXT:    store i32 1, i32* [[P_STRIDE30]], align 4
 // CHECK10-NEXT:    [[OMP_GLOBAL_THREAD_NUM31:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1]])
-// CHECK10-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM31]], i32 34, i32* [[P_LASTITER27]], i32* [[P_LOWERBOUND28]], i32* [[P_UPPERBOUND29]], i32* [[P_STRIDE30]], i32 1, i32 1)
+// CHECK10-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM31]], i32 34, i32* [[P_LASTITER27]], i32* [[P_LOWERBOUND28]], i32* [[P_UPPERBOUND29]], i32* [[P_STRIDE30]], i32 1, i32 0)
 // CHECK10-NEXT:    [[TMP7:%.*]] = load i32, i32* [[P_LOWERBOUND28]], align 4
 // CHECK10-NEXT:    [[TMP8:%.*]] = load i32, i32* [[P_UPPERBOUND29]], align 4
 // CHECK10-NEXT:    [[TMP9:%.*]] = sub i32 [[TMP8]], [[TMP7]]
 // CHECK10-NEXT:    [[TMP10:%.*]] = add i32 [[TMP9]], 1
 // CHECK10-NEXT:    br label [[OMP_SECTION_LOOP_HEADER14:%.*]]
 // CHECK10:       omp_section_loop.header14:
 // CHECK10-NEXT:    [[OMP_SECTION_LOOP_IV20:%.*]] = phi i32 [ 0, [[OMP_SECTION_LOOP_PREHEADER13]] ], [ [[OMP_SECTION_LOOP_NEXT22:%.*]], [[OMP_SECTION_LOOP_INC17:%.*]] ]
 // CHECK10-NEXT:    br label [[OMP_SECTION_LOOP_COND15:%.*]]

clang/test/OpenMP/irbuilder_for_iterator.cpp

 // CHECK-NEXT:    [[DOTCOUNT:%.*]] = load i64, i64* [[DOTCOUNT_ADDR]], align 8
 // CHECK-NEXT:    br label [[OMP_LOOP_PREHEADER:%.*]]
 // CHECK:       omp_loop.preheader:
 // CHECK-NEXT:    store i64 0, i64* [[P_LOWERBOUND]], align 8
 // CHECK-NEXT:    [[TMP2:%.*]] = sub i64 [[DOTCOUNT]], 1
 // CHECK-NEXT:    store i64 [[TMP2]], i64* [[P_UPPERBOUND]], align 8
 // CHECK-NEXT:    store i64 1, i64* [[P_STRIDE]], align 8
 // CHECK-NEXT:    [[OMP_GLOBAL_THREAD_NUM:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1:[0-9]+]])
-// CHECK-NEXT:    call void @__kmpc_for_static_init_8u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* [[P_LASTITER]], i64* [[P_LOWERBOUND]], i64* [[P_UPPERBOUND]], i64* [[P_STRIDE]], i64 1, i64 1)
+// CHECK-NEXT:    call void @__kmpc_for_static_init_8u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* [[P_LASTITER]], i64* [[P_LOWERBOUND]], i64* [[P_UPPERBOUND]], i64* [[P_STRIDE]], i64 1, i64 0)
 // CHECK-NEXT:    [[TMP3:%.*]] = load i64, i64* [[P_LOWERBOUND]], align 8
 // CHECK-NEXT:    [[TMP4:%.*]] = load i64, i64* [[P_UPPERBOUND]], align 8
 // CHECK-NEXT:    [[TMP5:%.*]] = sub i64 [[TMP4]], [[TMP3]]
 // CHECK-NEXT:    [[TMP6:%.*]] = add i64 [[TMP5]], 1
 // CHECK-NEXT:    br label [[OMP_LOOP_HEADER:%.*]]
 // CHECK:       omp_loop.header:
 // CHECK-NEXT:    [[OMP_LOOP_IV:%.*]] = phi i64 [ 0, [[OMP_LOOP_PREHEADER]] ], [ [[OMP_LOOP_NEXT:%.*]], [[OMP_LOOP_INC:%.*]] ]
 // CHECK-NEXT:    br label [[OMP_LOOP_COND:%.*]]

clang/test/OpenMP/irbuilder_for_rangefor.cpp

 // CHECK-NEXT:    [[DOTCOUNT:%.*]] = load i64, i64* [[DOTCOUNT_ADDR]], align 8
 // CHECK-NEXT:    br label [[OMP_LOOP_PREHEADER:%.*]]
 // CHECK:       omp_loop.preheader:
 // CHECK-NEXT:    store i64 0, i64* [[P_LOWERBOUND]], align 8
 // CHECK-NEXT:    [[TMP5:%.*]] = sub i64 [[DOTCOUNT]], 1
 // CHECK-NEXT:    store i64 [[TMP5]], i64* [[P_UPPERBOUND]], align 8
 // CHECK-NEXT:    store i64 1, i64* [[P_STRIDE]], align 8
 // CHECK-NEXT:    [[OMP_GLOBAL_THREAD_NUM:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1:[0-9]+]])
-// CHECK-NEXT:    call void @__kmpc_for_static_init_8u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* [[P_LASTITER]], i64* [[P_LOWERBOUND]], i64* [[P_UPPERBOUND]], i64* [[P_STRIDE]], i64 1, i64 1)
+// CHECK-NEXT:    call void @__kmpc_for_static_init_8u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* [[P_LASTITER]], i64* [[P_LOWERBOUND]], i64* [[P_UPPERBOUND]], i64* [[P_STRIDE]], i64 1, i64 0)
 // CHECK-NEXT:    [[TMP6:%.*]] = load i64, i64* [[P_LOWERBOUND]], align 8
 // CHECK-NEXT:    [[TMP7:%.*]] = load i64, i64* [[P_UPPERBOUND]], align 8
 // CHECK-NEXT:    [[TMP8:%.*]] = sub i64 [[TMP7]], [[TMP6]]
 // CHECK-NEXT:    [[TMP9:%.*]] = add i64 [[TMP8]], 1
 // CHECK-NEXT:    br label [[OMP_LOOP_HEADER:%.*]]
 // CHECK:       omp_loop.header:
 // CHECK-NEXT:    [[OMP_LOOP_IV:%.*]] = phi i64 [ 0, [[OMP_LOOP_PREHEADER]] ], [ [[OMP_LOOP_NEXT:%.*]], [[OMP_LOOP_INC:%.*]] ]
 // CHECK-NEXT:    br label [[OMP_LOOP_COND:%.*]]

clang/test/OpenMP/irbuilder_for_unsigned.c

 // CHECK-NEXT:    [[DOTCOUNT:%.*]] = load i32, i32* [[DOTCOUNT_ADDR]], align 4
 // CHECK-NEXT:    br label [[OMP_LOOP_PREHEADER:%.*]]
 // CHECK:       omp_loop.preheader:
 // CHECK-NEXT:    store i32 0, i32* [[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    [[TMP3:%.*]] = sub i32 [[DOTCOUNT]], 1
 // CHECK-NEXT:    store i32 [[TMP3]], i32* [[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    store i32 1, i32* [[P_STRIDE]], align 4
 // CHECK-NEXT:    [[OMP_GLOBAL_THREAD_NUM:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1:[0-9]+]])
-// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 1)
+// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* [[P_LASTITER]], i32* [[P_LOWERBOUND]], i32* [[P_UPPERBOUND]], i32* [[P_STRIDE]], i32 1, i32 0)
 // CHECK-NEXT:    [[TMP4:%.*]] = load i32, i32* [[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    [[TMP5:%.*]] = load i32, i32* [[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    [[TMP6:%.*]] = sub i32 [[TMP5]], [[TMP4]]
 // CHECK-NEXT:    [[TMP7:%.*]] = add i32 [[TMP6]], 1
 // 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:%.*]]

clang/test/OpenMP/irbuilder_for_unsigned_auto.c

+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs
+// RUN: %clang_cc1 -fopenmp-enable-irbuilder -verify -fopenmp -fopenmp-version=45 -x c++ -triple x86_64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
+// expected-no-diagnostics
+
+#ifndef HEADER
+#define HEADER
+
+// CHECK-LABEL: define {{.*}}@workshareloop_unsigned_auto(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[A_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[B_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[C_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[D_ADDR:.+]] = alloca float*, align 8
+// 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:    %[[P_LASTITER:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_LOWERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_UPPERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_STRIDE:.+]] = 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 float* %[[C:.+]], float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    store float* %[[D:.+]], float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    store i32 33, 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-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER]]:
+// CHECK-NEXT:    store i32 1, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    store i32 %[[DOTCOUNT]], i32* %[[P_UPPERBOUND]], align 4
+// CHECK-NEXT:    store i32 1, i32* %[[P_STRIDE]], align 4
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    call void @__kmpc_dispatch_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 38, i32 1, i32 %[[DOTCOUNT]], i32 1, i32 1)
+// CHECK-NEXT:    br label %[[OMP_LOOP_PREHEADER_OUTER_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_HEADER:.*]]:
+// CHECK-NEXT:    %[[OMP_LOOP_IV:.+]] = phi i32 [ %[[LB:.+]], %[[OMP_LOOP_PREHEADER_OUTER_COND]] ], [ %[[OMP_LOOP_NEXT:.+]], %[[OMP_LOOP_INC:.+]] ]
+// CHECK-NEXT:    br label %[[OMP_LOOP_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_COND]]:
+// CHECK-NEXT:    %[[UB:.+]] = load i32, i32* %[[P_UPPERBOUND]], align 4
+// CHECK-NEXT:    %[[OMP_LOOP_CMP:.+]] = icmp ult i32 %[[OMP_LOOP_IV]], %[[UB]]
+// CHECK-NEXT:    br i1 %[[OMP_LOOP_CMP]], label %[[OMP_LOOP_BODY:.+]], label %[[OMP_LOOP_PREHEADER_OUTER_COND]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_BODY]]:
+// CHECK-NEXT:    call void @__captured_stmt.1(i32* %[[I]], i32 %[[OMP_LOOP_IV]], %struct.anon.0* %[[AGG_CAPTURED1]])
+// CHECK-NEXT:    %[[TMP3:.+]] = load float*, float** %[[B_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM:.+]] = zext i32 %[[TMP4]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX:.+]] = getelementptr inbounds float, float* %[[TMP3]], i64 %[[IDXPROM]]
+// CHECK-NEXT:    %[[TMP5:.+]] = load float, float* %[[ARRAYIDX]], align 4
+// CHECK-NEXT:    %[[TMP6:.+]] = load float*, float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP7:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM2:.+]] = zext i32 %[[TMP7]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX3:.+]] = getelementptr inbounds float, float* %[[TMP6]], i64 %[[IDXPROM2]]
+// CHECK-NEXT:    %[[TMP8:.+]] = load float, float* %[[ARRAYIDX3]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = fmul float %[[TMP5]], %[[TMP8]]
+// CHECK-NEXT:    %[[TMP9:.+]] = load float*, float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP10:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM4:.+]] = zext i32 %[[TMP10]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX5:.+]] = getelementptr inbounds float, float* %[[TMP9]], i64 %[[IDXPROM4]]
+// CHECK-NEXT:    %[[TMP11:.+]] = load float, float* %[[ARRAYIDX5]], align 4
+// CHECK-NEXT:    %[[MUL6:.+]] = fmul float %[[MUL]], %[[TMP11]]
+// CHECK-NEXT:    %[[TMP12:.+]] = load float*, float** %[[A_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP13:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM7:.+]] = zext i32 %[[TMP13]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX8:.+]] = getelementptr inbounds float, float* %[[TMP12]], i64 %[[IDXPROM7]]
+// CHECK-NEXT:    store float %[[MUL6]], float* %[[ARRAYIDX8]], align 4
+// CHECK-NEXT:    br label %[[OMP_LOOP_INC]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_INC]]:
+// CHECK-NEXT:    %[[OMP_LOOP_NEXT]] = add nuw i32 %[[OMP_LOOP_IV]], 1
+// CHECK-NEXT:    br label %[[OMP_LOOP_HEADER]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_EXIT:.*]]:
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM9:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    call void @__kmpc_barrier(%struct.ident_t* @2, i32 %[[OMP_GLOBAL_THREAD_NUM9]])
+// CHECK-NEXT:    br label %[[OMP_LOOP_AFTER:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_AFTER]]:
+// CHECK-NEXT:    ret void
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER_OUTER_COND]]:
+// CHECK-NEXT:    %[[TMP14:.+]] = call i32 @__kmpc_dispatch_next_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]])
+// CHECK-NEXT:    %[[TMP15:.+]] = icmp ne i32 %[[TMP14]], 0
+// CHECK-NEXT:    %[[TMP16:.+]] = load i32, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    %[[LB]] = sub i32 %[[TMP16]], 1
+// CHECK-NEXT:    br i1 %[[TMP15]], label %[[OMP_LOOP_HEADER]], label %[[OMP_LOOP_EXIT]]
+// CHECK-NEXT:  }
+
+extern "C" void workshareloop_unsigned_auto(float *a, float *b, float *c, float *d) {
+#pragma omp for schedule(auto)
+  for (unsigned i = 33; i < 32000000; i += 7) {
+    a[i] = b[i] * c[i] * d[i];
+  }
+}
+
+#endif // HEADER
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[DISTANCE_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon*, align 8
+// CHECK-NEXT:    %[[DOTSTART:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTOP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTEP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    store i32* %[[DISTANCE:.+]], i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store %struct.anon* %[[__CONTEXT:.+]], %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon*, %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon, %struct.anon* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32*, i32** %[[TMP1]], align 8
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[TMP2]], align 4
+// CHECK-NEXT:    store i32 %[[TMP3]], i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    store i32 32000000, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    store i32 7, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[TMP5:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[CMP:.+]] = icmp ult i32 %[[TMP4]], %[[TMP5]]
+// CHECK-NEXT:    br i1 %[[CMP]], label %[[COND_TRUE:.+]], label %[[COND_FALSE:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_TRUE]]:
+// CHECK-NEXT:    %[[TMP6:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[TMP7:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[SUB:.+]] = sub i32 %[[TMP6]], %[[TMP7]]
+// CHECK-NEXT:    %[[TMP8:.+]] = load i32, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[DIV:.+]] = udiv i32 %[[SUB]], %[[TMP8]]
+// CHECK-NEXT:    br label %[[COND_END:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_FALSE]]:
+// CHECK-NEXT:    br label %[[COND_END]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_END]]:
+// CHECK-NEXT:    %[[COND:.+]] = phi i32 [ %[[DIV]], %[[COND_TRUE]] ], [ 0, %[[COND_FALSE]] ]
+// CHECK-NEXT:    %[[TMP9:.+]] = load i32*, i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[COND]], i32* %[[TMP9]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt.1(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[LOOPVAR_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[LOGICAL_ADDR:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon.0*, align 8
+// CHECK-NEXT:    store i32* %[[LOOPVAR:.+]], i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[LOGICAL:.+]], i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    store %struct.anon.0* %[[__CONTEXT:.+]], %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon.0*, %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon.0, %struct.anon.0* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32, i32* %[[TMP1]], align 4
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = mul i32 7, %[[TMP3]]
+// CHECK-NEXT:    %[[ADD:.+]] = add i32 %[[TMP2]], %[[MUL]]
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32*, i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[ADD]], i32* %[[TMP4]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK: ![[META0:[0-9]+]] = !{i32 1, !"wchar_size", i32 4}
+// CHECK: ![[META1:[0-9]+]] = !{i32 7, !"openmp", i32 45}
+// CHECK: ![[META2:[0-9]+]] =

clang/test/OpenMP/irbuilder_for_unsigned_dynamic.c

+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs
+// RUN: %clang_cc1 -fopenmp-enable-irbuilder -verify -fopenmp -fopenmp-version=45 -x c++ -triple x86_64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
+// expected-no-diagnostics
+
+#ifndef HEADER
+#define HEADER
+
+// CHECK-LABEL: define {{.*}}@workshareloop_unsigned_dynamic(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[A_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[B_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[C_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[D_ADDR:.+]] = alloca float*, align 8
+// 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:    %[[P_LASTITER:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_LOWERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_UPPERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_STRIDE:.+]] = 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 float* %[[C:.+]], float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    store float* %[[D:.+]], float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    store i32 33, 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-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER]]:
+// CHECK-NEXT:    store i32 1, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    store i32 %[[DOTCOUNT]], i32* %[[P_UPPERBOUND]], align 4
+// CHECK-NEXT:    store i32 1, i32* %[[P_STRIDE]], align 4
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    call void @__kmpc_dispatch_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 35, i32 1, i32 %[[DOTCOUNT]], i32 1, i32 1)
+// CHECK-NEXT:    br label %[[OMP_LOOP_PREHEADER_OUTER_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_HEADER:.*]]:
+// CHECK-NEXT:    %[[OMP_LOOP_IV:.+]] = phi i32 [ %[[LB:.+]], %[[OMP_LOOP_PREHEADER_OUTER_COND]] ], [ %[[OMP_LOOP_NEXT:.+]], %[[OMP_LOOP_INC:.+]] ]
+// CHECK-NEXT:    br label %[[OMP_LOOP_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_COND]]:
+// CHECK-NEXT:    %[[UB:.+]] = load i32, i32* %[[P_UPPERBOUND]], align 4
+// CHECK-NEXT:    %[[OMP_LOOP_CMP:.+]] = icmp ult i32 %[[OMP_LOOP_IV]], %[[UB]]
+// CHECK-NEXT:    br i1 %[[OMP_LOOP_CMP]], label %[[OMP_LOOP_BODY:.+]], label %[[OMP_LOOP_PREHEADER_OUTER_COND]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_BODY]]:
+// CHECK-NEXT:    call void @__captured_stmt.1(i32* %[[I]], i32 %[[OMP_LOOP_IV]], %struct.anon.0* %[[AGG_CAPTURED1]])
+// CHECK-NEXT:    %[[TMP3:.+]] = load float*, float** %[[B_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM:.+]] = zext i32 %[[TMP4]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX:.+]] = getelementptr inbounds float, float* %[[TMP3]], i64 %[[IDXPROM]]
+// CHECK-NEXT:    %[[TMP5:.+]] = load float, float* %[[ARRAYIDX]], align 4
+// CHECK-NEXT:    %[[TMP6:.+]] = load float*, float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP7:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM2:.+]] = zext i32 %[[TMP7]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX3:.+]] = getelementptr inbounds float, float* %[[TMP6]], i64 %[[IDXPROM2]]
+// CHECK-NEXT:    %[[TMP8:.+]] = load float, float* %[[ARRAYIDX3]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = fmul float %[[TMP5]], %[[TMP8]]
+// CHECK-NEXT:    %[[TMP9:.+]] = load float*, float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP10:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM4:.+]] = zext i32 %[[TMP10]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX5:.+]] = getelementptr inbounds float, float* %[[TMP9]], i64 %[[IDXPROM4]]
+// CHECK-NEXT:    %[[TMP11:.+]] = load float, float* %[[ARRAYIDX5]], align 4
+// CHECK-NEXT:    %[[MUL6:.+]] = fmul float %[[MUL]], %[[TMP11]]
+// CHECK-NEXT:    %[[TMP12:.+]] = load float*, float** %[[A_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP13:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM7:.+]] = zext i32 %[[TMP13]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX8:.+]] = getelementptr inbounds float, float* %[[TMP12]], i64 %[[IDXPROM7]]
+// CHECK-NEXT:    store float %[[MUL6]], float* %[[ARRAYIDX8]], align 4
+// CHECK-NEXT:    br label %[[OMP_LOOP_INC]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_INC]]:
+// CHECK-NEXT:    %[[OMP_LOOP_NEXT]] = add nuw i32 %[[OMP_LOOP_IV]], 1
+// CHECK-NEXT:    br label %[[OMP_LOOP_HEADER]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_EXIT:.*]]:
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM9:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    call void @__kmpc_barrier(%struct.ident_t* @2, i32 %[[OMP_GLOBAL_THREAD_NUM9]])
+// CHECK-NEXT:    br label %[[OMP_LOOP_AFTER:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_AFTER]]:
+// CHECK-NEXT:    ret void
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER_OUTER_COND]]:
+// CHECK-NEXT:    %[[TMP14:.+]] = call i32 @__kmpc_dispatch_next_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]])
+// CHECK-NEXT:    %[[TMP15:.+]] = icmp ne i32 %[[TMP14]], 0
+// CHECK-NEXT:    %[[TMP16:.+]] = load i32, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    %[[LB]] = sub i32 %[[TMP16]], 1
+// CHECK-NEXT:    br i1 %[[TMP15]], label %[[OMP_LOOP_HEADER]], label %[[OMP_LOOP_EXIT]]
+// CHECK-NEXT:  }
+
+extern "C" void workshareloop_unsigned_dynamic(float *a, float *b, float *c, float *d) {
+#pragma omp for schedule(dynamic)
+  for (unsigned i = 33; i < 32000000; i += 7) {
+    a[i] = b[i] * c[i] * d[i];
+  }
+}
+
+#endif // HEADER
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[DISTANCE_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon*, align 8
+// CHECK-NEXT:    %[[DOTSTART:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTOP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTEP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    store i32* %[[DISTANCE:.+]], i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store %struct.anon* %[[__CONTEXT:.+]], %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon*, %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon, %struct.anon* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32*, i32** %[[TMP1]], align 8
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[TMP2]], align 4
+// CHECK-NEXT:    store i32 %[[TMP3]], i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    store i32 32000000, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    store i32 7, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[TMP5:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[CMP:.+]] = icmp ult i32 %[[TMP4]], %[[TMP5]]
+// CHECK-NEXT:    br i1 %[[CMP]], label %[[COND_TRUE:.+]], label %[[COND_FALSE:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_TRUE]]:
+// CHECK-NEXT:    %[[TMP6:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[TMP7:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[SUB:.+]] = sub i32 %[[TMP6]], %[[TMP7]]
+// CHECK-NEXT:    %[[TMP8:.+]] = load i32, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[DIV:.+]] = udiv i32 %[[SUB]], %[[TMP8]]
+// CHECK-NEXT:    br label %[[COND_END:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_FALSE]]:
+// CHECK-NEXT:    br label %[[COND_END]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_END]]:
+// CHECK-NEXT:    %[[COND:.+]] = phi i32 [ %[[DIV]], %[[COND_TRUE]] ], [ 0, %[[COND_FALSE]] ]
+// CHECK-NEXT:    %[[TMP9:.+]] = load i32*, i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[COND]], i32* %[[TMP9]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt.1(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[LOOPVAR_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[LOGICAL_ADDR:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon.0*, align 8
+// CHECK-NEXT:    store i32* %[[LOOPVAR:.+]], i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[LOGICAL:.+]], i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    store %struct.anon.0* %[[__CONTEXT:.+]], %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon.0*, %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon.0, %struct.anon.0* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32, i32* %[[TMP1]], align 4
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = mul i32 7, %[[TMP3]]
+// CHECK-NEXT:    %[[ADD:.+]] = add i32 %[[TMP2]], %[[MUL]]
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32*, i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[ADD]], i32* %[[TMP4]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK: ![[META0:[0-9]+]] = !{i32 1, !"wchar_size", i32 4}
+// CHECK: ![[META1:[0-9]+]] = !{i32 7, !"openmp", i32 45}
+// CHECK: ![[META2:[0-9]+]] =

clang/test/OpenMP/irbuilder_for_unsigned_dynamic_chunked.c

+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs
+// RUN: %clang_cc1 -fopenmp-enable-irbuilder -verify -fopenmp -fopenmp-version=45 -x c++ -triple x86_64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
+// expected-no-diagnostics
+
+#ifndef HEADER
+#define HEADER
+
+// CHECK-LABEL: define {{.*}}@workshareloop_unsigned_dynamic_chunked(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[A_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[B_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[C_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[D_ADDR:.+]] = alloca float*, align 8
+// 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:    %[[P_LASTITER:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_LOWERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_UPPERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_STRIDE:.+]] = 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 float* %[[C:.+]], float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    store float* %[[D:.+]], float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    store i32 33, 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-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER]]:
+// CHECK-NEXT:    store i32 1, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    store i32 %[[DOTCOUNT]], i32* %[[P_UPPERBOUND]], align 4
+// CHECK-NEXT:    store i32 1, i32* %[[P_STRIDE]], align 4
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    call void @__kmpc_dispatch_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 35, i32 1, i32 %[[DOTCOUNT]], i32 1, i32 5)
+// CHECK-NEXT:    br label %[[OMP_LOOP_PREHEADER_OUTER_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_HEADER:.*]]:
+// CHECK-NEXT:    %[[OMP_LOOP_IV:.+]] = phi i32 [ %[[LB:.+]], %[[OMP_LOOP_PREHEADER_OUTER_COND]] ], [ %[[OMP_LOOP_NEXT:.+]], %[[OMP_LOOP_INC:.+]] ]
+// CHECK-NEXT:    br label %[[OMP_LOOP_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_COND]]:
+// CHECK-NEXT:    %[[UB:.+]] = load i32, i32* %[[P_UPPERBOUND]], align 4
+// CHECK-NEXT:    %[[OMP_LOOP_CMP:.+]] = icmp ult i32 %[[OMP_LOOP_IV]], %[[UB]]
+// CHECK-NEXT:    br i1 %[[OMP_LOOP_CMP]], label %[[OMP_LOOP_BODY:.+]], label %[[OMP_LOOP_PREHEADER_OUTER_COND]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_BODY]]:
+// CHECK-NEXT:    call void @__captured_stmt.1(i32* %[[I]], i32 %[[OMP_LOOP_IV]], %struct.anon.0* %[[AGG_CAPTURED1]])
+// CHECK-NEXT:    %[[TMP3:.+]] = load float*, float** %[[B_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM:.+]] = zext i32 %[[TMP4]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX:.+]] = getelementptr inbounds float, float* %[[TMP3]], i64 %[[IDXPROM]]
+// CHECK-NEXT:    %[[TMP5:.+]] = load float, float* %[[ARRAYIDX]], align 4
+// CHECK-NEXT:    %[[TMP6:.+]] = load float*, float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP7:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM2:.+]] = zext i32 %[[TMP7]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX3:.+]] = getelementptr inbounds float, float* %[[TMP6]], i64 %[[IDXPROM2]]
+// CHECK-NEXT:    %[[TMP8:.+]] = load float, float* %[[ARRAYIDX3]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = fmul float %[[TMP5]], %[[TMP8]]
+// CHECK-NEXT:    %[[TMP9:.+]] = load float*, float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP10:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM4:.+]] = zext i32 %[[TMP10]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX5:.+]] = getelementptr inbounds float, float* %[[TMP9]], i64 %[[IDXPROM4]]
+// CHECK-NEXT:    %[[TMP11:.+]] = load float, float* %[[ARRAYIDX5]], align 4
+// CHECK-NEXT:    %[[MUL6:.+]] = fmul float %[[MUL]], %[[TMP11]]
+// CHECK-NEXT:    %[[TMP12:.+]] = load float*, float** %[[A_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP13:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM7:.+]] = zext i32 %[[TMP13]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX8:.+]] = getelementptr inbounds float, float* %[[TMP12]], i64 %[[IDXPROM7]]
+// CHECK-NEXT:    store float %[[MUL6]], float* %[[ARRAYIDX8]], align 4
+// CHECK-NEXT:    br label %[[OMP_LOOP_INC]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_INC]]:
+// CHECK-NEXT:    %[[OMP_LOOP_NEXT]] = add nuw i32 %[[OMP_LOOP_IV]], 1
+// CHECK-NEXT:    br label %[[OMP_LOOP_HEADER]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_EXIT:.*]]:
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM9:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    call void @__kmpc_barrier(%struct.ident_t* @2, i32 %[[OMP_GLOBAL_THREAD_NUM9]])
+// CHECK-NEXT:    br label %[[OMP_LOOP_AFTER:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_AFTER]]:
+// CHECK-NEXT:    ret void
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER_OUTER_COND]]:
+// CHECK-NEXT:    %[[TMP14:.+]] = call i32 @__kmpc_dispatch_next_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]])
+// CHECK-NEXT:    %[[TMP15:.+]] = icmp ne i32 %[[TMP14]], 0
+// CHECK-NEXT:    %[[TMP16:.+]] = load i32, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    %[[LB]] = sub i32 %[[TMP16]], 1
+// CHECK-NEXT:    br i1 %[[TMP15]], label %[[OMP_LOOP_HEADER]], label %[[OMP_LOOP_EXIT]]
+// CHECK-NEXT:  }
+
+extern "C" void workshareloop_unsigned_dynamic_chunked(float *a, float *b, float *c, float *d) {
+#pragma omp for schedule(dynamic, 5)
+  for (unsigned i = 33; i < 32000000; i += 7) {
+    a[i] = b[i] * c[i] * d[i];
+  }
+}
+
+#endif // HEADER
+
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[DISTANCE_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon*, align 8
+// CHECK-NEXT:    %[[DOTSTART:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTOP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTEP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    store i32* %[[DISTANCE:.+]], i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store %struct.anon* %[[__CONTEXT:.+]], %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon*, %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon, %struct.anon* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32*, i32** %[[TMP1]], align 8
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[TMP2]], align 4
+// CHECK-NEXT:    store i32 %[[TMP3]], i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    store i32 32000000, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    store i32 7, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[TMP5:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[CMP:.+]] = icmp ult i32 %[[TMP4]], %[[TMP5]]
+// CHECK-NEXT:    br i1 %[[CMP]], label %[[COND_TRUE:.+]], label %[[COND_FALSE:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_TRUE]]:
+// CHECK-NEXT:    %[[TMP6:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[TMP7:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[SUB:.+]] = sub i32 %[[TMP6]], %[[TMP7]]
+// CHECK-NEXT:    %[[TMP8:.+]] = load i32, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[DIV:.+]] = udiv i32 %[[SUB]], %[[TMP8]]
+// CHECK-NEXT:    br label %[[COND_END:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_FALSE]]:
+// CHECK-NEXT:    br label %[[COND_END]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_END]]:
+// CHECK-NEXT:    %[[COND:.+]] = phi i32 [ %[[DIV]], %[[COND_TRUE]] ], [ 0, %[[COND_FALSE]] ]
+// CHECK-NEXT:    %[[TMP9:.+]] = load i32*, i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[COND]], i32* %[[TMP9]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt.1(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[LOOPVAR_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[LOGICAL_ADDR:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon.0*, align 8
+// CHECK-NEXT:    store i32* %[[LOOPVAR:.+]], i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[LOGICAL:.+]], i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    store %struct.anon.0* %[[__CONTEXT:.+]], %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon.0*, %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon.0, %struct.anon.0* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32, i32* %[[TMP1]], align 4
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = mul i32 7, %[[TMP3]]
+// CHECK-NEXT:    %[[ADD:.+]] = add i32 %[[TMP2]], %[[MUL]]
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32*, i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[ADD]], i32* %[[TMP4]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK: ![[META0:[0-9]+]] = !{i32 1, !"wchar_size", i32 4}
+// CHECK: ![[META1:[0-9]+]] = !{i32 7, !"openmp", i32 45}
+// CHECK: ![[META2:[0-9]+]] =

clang/test/OpenMP/irbuilder_for_unsigned_runtime.c

+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs
+// RUN: %clang_cc1 -fopenmp-enable-irbuilder -verify -fopenmp -fopenmp-version=45 -x c++ -triple x86_64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
+// expected-no-diagnostics
+
+#ifndef HEADER
+#define HEADER
+
+// CHECK-LABEL: define {{.*}}@workshareloop_unsigned_runtime(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[A_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[B_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[C_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[D_ADDR:.+]] = alloca float*, align 8
+// 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:    %[[P_LASTITER:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_LOWERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_UPPERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_STRIDE:.+]] = 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 float* %[[C:.+]], float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    store float* %[[D:.+]], float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    store i32 33, 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-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER]]:
+// CHECK-NEXT:    store i32 1, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    store i32 %[[DOTCOUNT]], i32* %[[P_UPPERBOUND]], align 4
+// CHECK-NEXT:    store i32 1, i32* %[[P_STRIDE]], align 4
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    call void @__kmpc_dispatch_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 37, i32 1, i32 %[[DOTCOUNT]], i32 1, i32 1)
+// CHECK-NEXT:    br label %[[OMP_LOOP_PREHEADER_OUTER_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_HEADER:.*]]:
+// CHECK-NEXT:    %[[OMP_LOOP_IV:.+]] = phi i32 [ %[[LB:.+]], %[[OMP_LOOP_PREHEADER_OUTER_COND]] ], [ %[[OMP_LOOP_NEXT:.+]], %[[OMP_LOOP_INC:.+]] ]
+// CHECK-NEXT:    br label %[[OMP_LOOP_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_COND]]:
+// CHECK-NEXT:    %[[UB:.+]] = load i32, i32* %[[P_UPPERBOUND]], align 4
+// CHECK-NEXT:    %[[OMP_LOOP_CMP:.+]] = icmp ult i32 %[[OMP_LOOP_IV]], %[[UB]]
+// CHECK-NEXT:    br i1 %[[OMP_LOOP_CMP]], label %[[OMP_LOOP_BODY:.+]], label %[[OMP_LOOP_PREHEADER_OUTER_COND]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_BODY]]:
+// CHECK-NEXT:    call void @__captured_stmt.1(i32* %[[I]], i32 %[[OMP_LOOP_IV]], %struct.anon.0* %[[AGG_CAPTURED1]])
+// CHECK-NEXT:    %[[TMP3:.+]] = load float*, float** %[[B_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM:.+]] = zext i32 %[[TMP4]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX:.+]] = getelementptr inbounds float, float* %[[TMP3]], i64 %[[IDXPROM]]
+// CHECK-NEXT:    %[[TMP5:.+]] = load float, float* %[[ARRAYIDX]], align 4
+// CHECK-NEXT:    %[[TMP6:.+]] = load float*, float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP7:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM2:.+]] = zext i32 %[[TMP7]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX3:.+]] = getelementptr inbounds float, float* %[[TMP6]], i64 %[[IDXPROM2]]
+// CHECK-NEXT:    %[[TMP8:.+]] = load float, float* %[[ARRAYIDX3]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = fmul float %[[TMP5]], %[[TMP8]]
+// CHECK-NEXT:    %[[TMP9:.+]] = load float*, float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP10:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM4:.+]] = zext i32 %[[TMP10]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX5:.+]] = getelementptr inbounds float, float* %[[TMP9]], i64 %[[IDXPROM4]]
+// CHECK-NEXT:    %[[TMP11:.+]] = load float, float* %[[ARRAYIDX5]], align 4
+// CHECK-NEXT:    %[[MUL6:.+]] = fmul float %[[MUL]], %[[TMP11]]
+// CHECK-NEXT:    %[[TMP12:.+]] = load float*, float** %[[A_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP13:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM7:.+]] = zext i32 %[[TMP13]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX8:.+]] = getelementptr inbounds float, float* %[[TMP12]], i64 %[[IDXPROM7]]
+// CHECK-NEXT:    store float %[[MUL6]], float* %[[ARRAYIDX8]], align 4
+// CHECK-NEXT:    br label %[[OMP_LOOP_INC]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_INC]]:
+// CHECK-NEXT:    %[[OMP_LOOP_NEXT]] = add nuw i32 %[[OMP_LOOP_IV]], 1
+// CHECK-NEXT:    br label %[[OMP_LOOP_HEADER]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_EXIT:.*]]:
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM9:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    call void @__kmpc_barrier(%struct.ident_t* @2, i32 %[[OMP_GLOBAL_THREAD_NUM9]])
+// CHECK-NEXT:    br label %[[OMP_LOOP_AFTER:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_AFTER]]:
+// CHECK-NEXT:    ret void
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER_OUTER_COND]]:
+// CHECK-NEXT:    %[[TMP14:.+]] = call i32 @__kmpc_dispatch_next_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]])
+// CHECK-NEXT:    %[[TMP15:.+]] = icmp ne i32 %[[TMP14]], 0
+// CHECK-NEXT:    %[[TMP16:.+]] = load i32, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    %[[LB]] = sub i32 %[[TMP16]], 1
+// CHECK-NEXT:    br i1 %[[TMP15]], label %[[OMP_LOOP_HEADER]], label %[[OMP_LOOP_EXIT]]
+// CHECK-NEXT:  }
+
+extern "C" void workshareloop_unsigned_runtime(float *a, float *b, float *c, float *d) {
+#pragma omp for schedule(runtime)
+  for (unsigned i = 33; i < 32000000; i += 7) {
+    a[i] = b[i] * c[i] * d[i];
+  }
+}
+
+#endif // HEADER
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[DISTANCE_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon*, align 8
+// CHECK-NEXT:    %[[DOTSTART:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTOP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTEP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    store i32* %[[DISTANCE:.+]], i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store %struct.anon* %[[__CONTEXT:.+]], %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon*, %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon, %struct.anon* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32*, i32** %[[TMP1]], align 8
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[TMP2]], align 4
+// CHECK-NEXT:    store i32 %[[TMP3]], i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    store i32 32000000, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    store i32 7, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[TMP5:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[CMP:.+]] = icmp ult i32 %[[TMP4]], %[[TMP5]]
+// CHECK-NEXT:    br i1 %[[CMP]], label %[[COND_TRUE:.+]], label %[[COND_FALSE:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_TRUE]]:
+// CHECK-NEXT:    %[[TMP6:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[TMP7:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[SUB:.+]] = sub i32 %[[TMP6]], %[[TMP7]]
+// CHECK-NEXT:    %[[TMP8:.+]] = load i32, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[DIV:.+]] = udiv i32 %[[SUB]], %[[TMP8]]
+// CHECK-NEXT:    br label %[[COND_END:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_FALSE]]:
+// CHECK-NEXT:    br label %[[COND_END]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_END]]:
+// CHECK-NEXT:    %[[COND:.+]] = phi i32 [ %[[DIV]], %[[COND_TRUE]] ], [ 0, %[[COND_FALSE]] ]
+// CHECK-NEXT:    %[[TMP9:.+]] = load i32*, i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[COND]], i32* %[[TMP9]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt.1(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[LOOPVAR_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[LOGICAL_ADDR:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon.0*, align 8
+// CHECK-NEXT:    store i32* %[[LOOPVAR:.+]], i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[LOGICAL:.+]], i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    store %struct.anon.0* %[[__CONTEXT:.+]], %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon.0*, %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon.0, %struct.anon.0* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32, i32* %[[TMP1]], align 4
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = mul i32 7, %[[TMP3]]
+// CHECK-NEXT:    %[[ADD:.+]] = add i32 %[[TMP2]], %[[MUL]]
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32*, i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[ADD]], i32* %[[TMP4]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK: ![[META0:[0-9]+]] = !{i32 1, !"wchar_size", i32 4}
+// CHECK: ![[META1:[0-9]+]] = !{i32 7, !"openmp", i32 45}
+// CHECK: ![[META2:[0-9]+]] =

clang/test/OpenMP/irbuilder_for_unsigned_static_chunked.c

+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs
+// RUN: %clang_cc1 -fopenmp-enable-irbuilder -verify -fopenmp -fopenmp-version=45 -x c++ -triple x86_64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
+// expected-no-diagnostics
+
+#ifndef HEADER
+#define HEADER
+
+// CHECK-LABEL: define {{.*}}@workshareloop_unsigned_static_chunked(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[A_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[B_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[C_ADDR:.+]] = alloca float*, align 8
+// CHECK-NEXT:    %[[D_ADDR:.+]] = alloca float*, align 8
+// 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:    %[[P_LASTITER:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_LOWERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_UPPERBOUND:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[P_STRIDE:.+]] = 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 float* %[[C:.+]], float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    store float* %[[D:.+]], float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    store i32 33, 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-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER]]:
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    store i32 0, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    %[[TMP3:.+]] = sub i32 %[[DOTCOUNT]], 1
+// CHECK-NEXT:    store i32 %[[TMP3]], i32* %[[P_UPPERBOUND]], align 4
+// CHECK-NEXT:    store i32 1, i32* %[[P_STRIDE]], align 4
+// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 33, i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]], i32 1, i32 5)
+// CHECK-NEXT:    %[[OMP_FIRSTCHUNK_LB:.+]] = load i32, i32* %[[P_LOWERBOUND]], align 4
+// CHECK-NEXT:    %[[OMP_DISPATCH_STRIDE:.+]] = load i32, i32* %[[P_STRIDE]], align 4
+// CHECK-NEXT:    %[[TMP4:.+]] = sub nuw i32 %[[DOTCOUNT]], %[[OMP_FIRSTCHUNK_LB]]
+// CHECK-NEXT:    %[[TMP5:.+]] = icmp ule i32 %[[DOTCOUNT]], %[[OMP_FIRSTCHUNK_LB]]
+// CHECK-NEXT:    %[[TMP6:.+]] = sub i32 %[[TMP4]], 1
+// CHECK-NEXT:    %[[TMP7:.+]] = udiv i32 %[[TMP6]], %[[OMP_DISPATCH_STRIDE]]
+// CHECK-NEXT:    %[[TMP8:.+]] = add i32 %[[TMP7]], 1
+// CHECK-NEXT:    %[[TMP9:.+]] = icmp ule i32 %[[TMP4]], %[[OMP_DISPATCH_STRIDE]]
+// CHECK-NEXT:    %[[TMP10:.+]] = select i1 %[[TMP9]], i32 1, i32 %[[TMP8]]
+// CHECK-NEXT:    %[[OMP_DISPATCH_TRIPCOUNT:.+]] = select i1 %[[TMP5]], i32 0, i32 %[[TMP10]]
+// CHECK-NEXT:    br label %[[OMP_DISPATCH_PREHEADER:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_DISPATCH_PREHEADER]]:
+// CHECK-NEXT:    br label %[[OMP_DISPATCH_HEADER:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_DISPATCH_HEADER]]:
+// CHECK-NEXT:    %[[OMP_DISPATCH_IV:.+]] = phi i32 [ 0, %[[OMP_DISPATCH_PREHEADER]] ], [ %[[OMP_DISPATCH_NEXT:.+]], %[[OMP_DISPATCH_INC:.+]] ]
+// CHECK-NEXT:    br label %[[OMP_DISPATCH_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_DISPATCH_COND]]:
+// CHECK-NEXT:    %[[OMP_DISPATCH_CMP:.+]] = icmp ult i32 %[[OMP_DISPATCH_IV]], %[[OMP_DISPATCH_TRIPCOUNT]]
+// CHECK-NEXT:    br i1 %[[OMP_DISPATCH_CMP]], label %[[OMP_DISPATCH_BODY:.+]], label %[[OMP_DISPATCH_EXIT:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_DISPATCH_BODY]]:
+// CHECK-NEXT:    %[[TMP11:.+]] = mul i32 %[[OMP_DISPATCH_IV]], %[[OMP_DISPATCH_STRIDE]]
+// CHECK-NEXT:    %[[TMP12:.+]] = add i32 %[[TMP11]], %[[OMP_FIRSTCHUNK_LB]]
+// CHECK-NEXT:    br label %[[OMP_LOOP_PREHEADER9:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_DISPATCH_INC]]:
+// CHECK-NEXT:    %[[OMP_DISPATCH_NEXT]] = add nuw i32 %[[OMP_DISPATCH_IV]], 1
+// CHECK-NEXT:    br label %[[OMP_DISPATCH_HEADER]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_DISPATCH_EXIT]]:
+// CHECK-NEXT:    call void @__kmpc_for_static_fini(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]])
+// CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM10:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
+// CHECK-NEXT:    call void @__kmpc_barrier(%struct.ident_t* @2, i32 %[[OMP_GLOBAL_THREAD_NUM10]])
+// CHECK-NEXT:    br label %[[OMP_DISPATCH_AFTER:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_DISPATCH_AFTER]]:
+// CHECK-NEXT:    br label %[[OMP_LOOP_AFTER:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_PREHEADER9]]:
+// CHECK-NEXT:    %[[TMP13:.+]] = add i32 %[[TMP12]], 5
+// CHECK-NEXT:    %[[OMP_CHUNK_IS_LAST:.+]] = icmp uge i32 %[[TMP13]], %[[DOTCOUNT]]
+// CHECK-NEXT:    %[[TMP14:.+]] = sub i32 %[[DOTCOUNT]], %[[TMP12]]
+// CHECK-NEXT:    %[[OMP_CHUNK_TRIPCOUNT:.+]] = select i1 %[[OMP_CHUNK_IS_LAST]], i32 %[[TMP14]], i32 5
+// CHECK-NEXT:    br label %[[OMP_LOOP_HEADER:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_HEADER]]:
+// CHECK-NEXT:    %[[OMP_LOOP_IV:.+]] = phi i32 [ 0, %[[OMP_LOOP_PREHEADER9]] ], [ %[[OMP_LOOP_NEXT:.+]], %[[OMP_LOOP_INC:.+]] ]
+// CHECK-NEXT:    br label %[[OMP_LOOP_COND:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_COND]]:
+// CHECK-NEXT:    %[[OMP_LOOP_CMP:.+]] = icmp ult i32 %[[OMP_LOOP_IV]], %[[OMP_CHUNK_TRIPCOUNT]]
+// CHECK-NEXT:    br i1 %[[OMP_LOOP_CMP]], label %[[OMP_LOOP_BODY:.+]], label %[[OMP_LOOP_EXIT:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_BODY]]:
+// CHECK-NEXT:    %[[TMP15:.+]] = add i32 %[[OMP_LOOP_IV]], %[[TMP12]]
+// CHECK-NEXT:    call void @__captured_stmt.1(i32* %[[I]], i32 %[[TMP15]], %struct.anon.0* %[[AGG_CAPTURED1]])
+// CHECK-NEXT:    %[[TMP16:.+]] = load float*, float** %[[B_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP17:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM:.+]] = zext i32 %[[TMP17]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX:.+]] = getelementptr inbounds float, float* %[[TMP16]], i64 %[[IDXPROM]]
+// CHECK-NEXT:    %[[TMP18:.+]] = load float, float* %[[ARRAYIDX]], align 4
+// CHECK-NEXT:    %[[TMP19:.+]] = load float*, float** %[[C_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP20:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM2:.+]] = zext i32 %[[TMP20]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX3:.+]] = getelementptr inbounds float, float* %[[TMP19]], i64 %[[IDXPROM2]]
+// CHECK-NEXT:    %[[TMP21:.+]] = load float, float* %[[ARRAYIDX3]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = fmul float %[[TMP18]], %[[TMP21]]
+// CHECK-NEXT:    %[[TMP22:.+]] = load float*, float** %[[D_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP23:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM4:.+]] = zext i32 %[[TMP23]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX5:.+]] = getelementptr inbounds float, float* %[[TMP22]], i64 %[[IDXPROM4]]
+// CHECK-NEXT:    %[[TMP24:.+]] = load float, float* %[[ARRAYIDX5]], align 4
+// CHECK-NEXT:    %[[MUL6:.+]] = fmul float %[[MUL]], %[[TMP24]]
+// CHECK-NEXT:    %[[TMP25:.+]] = load float*, float** %[[A_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP26:.+]] = load i32, i32* %[[I]], align 4
+// CHECK-NEXT:    %[[IDXPROM7:.+]] = zext i32 %[[TMP26]] to i64
+// CHECK-NEXT:    %[[ARRAYIDX8:.+]] = getelementptr inbounds float, float* %[[TMP25]], i64 %[[IDXPROM7]]
+// CHECK-NEXT:    store float %[[MUL6]], float* %[[ARRAYIDX8]], align 4
+// CHECK-NEXT:    br label %[[OMP_LOOP_INC]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_INC]]:
+// CHECK-NEXT:    %[[OMP_LOOP_NEXT]] = add nuw i32 %[[OMP_LOOP_IV]], 1
+// CHECK-NEXT:    br label %[[OMP_LOOP_HEADER]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_EXIT]]:
+// CHECK-NEXT:    br label %[[OMP_DISPATCH_INC]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[OMP_LOOP_AFTER]]:
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+extern "C" void workshareloop_unsigned_static_chunked(float *a, float *b, float *c, float *d) {
+#pragma omp for schedule(static, 5)
+  for (unsigned i = 33; i < 32000000; i += 7) {
+    a[i] = b[i] * c[i] * d[i];
+  }
+}
+
+#endif // HEADER
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[DISTANCE_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon*, align 8
+// CHECK-NEXT:    %[[DOTSTART:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTOP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[DOTSTEP:.+]] = alloca i32, align 4
+// CHECK-NEXT:    store i32* %[[DISTANCE:.+]], i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store %struct.anon* %[[__CONTEXT:.+]], %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon*, %struct.anon** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon, %struct.anon* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32*, i32** %[[TMP1]], align 8
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[TMP2]], align 4
+// CHECK-NEXT:    store i32 %[[TMP3]], i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    store i32 32000000, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    store i32 7, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[TMP5:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[CMP:.+]] = icmp ult i32 %[[TMP4]], %[[TMP5]]
+// CHECK-NEXT:    br i1 %[[CMP]], label %[[COND_TRUE:.+]], label %[[COND_FALSE:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_TRUE]]:
+// CHECK-NEXT:    %[[TMP6:.+]] = load i32, i32* %[[DOTSTOP]], align 4
+// CHECK-NEXT:    %[[TMP7:.+]] = load i32, i32* %[[DOTSTART]], align 4
+// CHECK-NEXT:    %[[SUB:.+]] = sub i32 %[[TMP6]], %[[TMP7]]
+// CHECK-NEXT:    %[[TMP8:.+]] = load i32, i32* %[[DOTSTEP]], align 4
+// CHECK-NEXT:    %[[DIV:.+]] = udiv i32 %[[SUB]], %[[TMP8]]
+// CHECK-NEXT:    br label %[[COND_END:.+]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_FALSE]]:
+// CHECK-NEXT:    br label %[[COND_END]]
+// CHECK-EMPTY:
+// CHECK-NEXT:  [[COND_END]]:
+// CHECK-NEXT:    %[[COND:.+]] = phi i32 [ %[[DIV]], %[[COND_TRUE]] ], [ 0, %[[COND_FALSE]] ]
+// CHECK-NEXT:    %[[TMP9:.+]] = load i32*, i32** %[[DISTANCE_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[COND]], i32* %[[TMP9]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK-LABEL: define {{.*}}@__captured_stmt.1(
+// CHECK-NEXT:  [[ENTRY:.*]]:
+// CHECK-NEXT:    %[[LOOPVAR_ADDR:.+]] = alloca i32*, align 8
+// CHECK-NEXT:    %[[LOGICAL_ADDR:.+]] = alloca i32, align 4
+// CHECK-NEXT:    %[[__CONTEXT_ADDR:.+]] = alloca %struct.anon.0*, align 8
+// CHECK-NEXT:    store i32* %[[LOOPVAR:.+]], i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[LOGICAL:.+]], i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    store %struct.anon.0* %[[__CONTEXT:.+]], %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP0:.+]] = load %struct.anon.0*, %struct.anon.0** %[[__CONTEXT_ADDR]], align 8
+// CHECK-NEXT:    %[[TMP1:.+]] = getelementptr inbounds %struct.anon.0, %struct.anon.0* %[[TMP0]], i32 0, i32 0
+// CHECK-NEXT:    %[[TMP2:.+]] = load i32, i32* %[[TMP1]], align 4
+// CHECK-NEXT:    %[[TMP3:.+]] = load i32, i32* %[[LOGICAL_ADDR]], align 4
+// CHECK-NEXT:    %[[MUL:.+]] = mul i32 7, %[[TMP3]]
+// CHECK-NEXT:    %[[ADD:.+]] = add i32 %[[TMP2]], %[[MUL]]
+// CHECK-NEXT:    %[[TMP4:.+]] = load i32*, i32** %[[LOOPVAR_ADDR]], align 8
+// CHECK-NEXT:    store i32 %[[ADD]], i32* %[[TMP4]], align 4
+// CHECK-NEXT:    ret void
+// CHECK-NEXT:  }
+
+// CHECK: ![[META0:[0-9]+]] = !{i32 1, !"wchar_size", i32 4}
+// CHECK: ![[META1:[0-9]+]] = !{i32 7, !"openmp", i32 45}
+// CHECK: ![[META2:[0-9]+]] =

clang/test/OpenMP/irbuilder_nested_parallel_for.c

 // CHECK-NEXT:    br label [[OMP_PAR_OUTLINED_EXIT:%.*]]
 // CHECK:       omp.par.outlined.exit:
 // CHECK-NEXT:    br label [[OMP_PAR_EXIT_SPLIT:%.*]]
 // CHECK:       omp.par.exit.split:
 // CHECK-NEXT:    ret void
 //
 // CHECK-DEBUG-LABEL: @_Z14parallel_for_0v(
 // CHECK-DEBUG-NEXT:  entry:
-// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1:[0-9]+]]), !dbg [[DBG12:![0-9]+]]
+// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1:[0-9]+]]), !dbg [[DBG13:![0-9]+]]
 // CHECK-DEBUG-NEXT:    br label [[OMP_PARALLEL:%.*]]
 // CHECK-DEBUG:       omp_parallel:
-// CHECK-DEBUG-NEXT:    call void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* @[[GLOB1]], i32 0, void (i32*, i32*, ...)* bitcast (void (i32*, i32*)* @_Z14parallel_for_0v..omp_par to void (i32*, i32*, ...)*)), !dbg [[DBG13:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* @[[GLOB1]], i32 0, void (i32*, i32*, ...)* bitcast (void (i32*, i32*)* @_Z14parallel_for_0v..omp_par to void (i32*, i32*, ...)*)), !dbg [[DBG14:![0-9]+]]
 // CHECK-DEBUG-NEXT:    br label [[OMP_PAR_OUTLINED_EXIT:%.*]]
 // CHECK-DEBUG:       omp.par.outlined.exit:
 // CHECK-DEBUG-NEXT:    br label [[OMP_PAR_EXIT_SPLIT:%.*]]
 // CHECK-DEBUG:       omp.par.exit.split:
-// CHECK-DEBUG-NEXT:    ret void, !dbg [[DBG17:![0-9]+]]
+// CHECK-DEBUG-NEXT:    ret void, !dbg [[DBG18:![0-9]+]]
 //
 void parallel_for_0(void) {
 #pragma omp parallel
   {
 #pragma omp for
     for (int i = 0; i < 100; ++i) {
     }
   }

 // CHECK-NEXT:    ret void
 //
 // CHECK-DEBUG-LABEL: @_Z14parallel_for_1Pfid(
 // CHECK-DEBUG-NEXT:  entry:
 // CHECK-DEBUG-NEXT:    [[R_ADDR:%.*]] = alloca float*, align 8
 // CHECK-DEBUG-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4
 // CHECK-DEBUG-NEXT:    [[B_ADDR:%.*]] = alloca double, align 8
 // CHECK-DEBUG-NEXT:    store float* [[R:%.*]], float** [[R_ADDR]], align 8
-// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata float** [[R_ADDR]], metadata [[META71:![0-9]+]], metadata !DIExpression()), !dbg [[DBG72:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata float** [[R_ADDR]], metadata [[META73:![0-9]+]], metadata !DIExpression()), !dbg [[DBG74:![0-9]+]]
 // CHECK-DEBUG-NEXT:    store i32 [[A:%.*]], i32* [[A_ADDR]], align 4
-// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata i32* [[A_ADDR]], metadata [[META73:![0-9]+]], metadata !DIExpression()), !dbg [[DBG74:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata i32* [[A_ADDR]], metadata [[META75:![0-9]+]], metadata !DIExpression()), !dbg [[DBG76:![0-9]+]]
 // CHECK-DEBUG-NEXT:    store double [[B:%.*]], double* [[B_ADDR]], align 8
-// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata double* [[B_ADDR]], metadata [[META75:![0-9]+]], metadata !DIExpression()), !dbg [[DBG76:![0-9]+]]
-// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB6:[0-9]+]]), !dbg [[DBG77:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata double* [[B_ADDR]], metadata [[META77:![0-9]+]], metadata !DIExpression()), !dbg [[DBG78:![0-9]+]]
+// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB6:[0-9]+]]), !dbg [[DBG79:![0-9]+]]
 // CHECK-DEBUG-NEXT:    br label [[OMP_PARALLEL:%.*]]
 // CHECK-DEBUG:       omp_parallel:
-// CHECK-DEBUG-NEXT:    call void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* @[[GLOB6]], i32 3, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i32*, double*, float**)* @_Z14parallel_for_1Pfid..omp_par.4 to void (i32*, i32*, ...)*), i32* [[A_ADDR]], double* [[B_ADDR]], float** [[R_ADDR]]), !dbg [[DBG78:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* @[[GLOB6]], i32 3, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i32*, double*, float**)* @_Z14parallel_for_1Pfid..omp_par.4 to void (i32*, i32*, ...)*), i32* [[A_ADDR]], double* [[B_ADDR]], float** [[R_ADDR]]), !dbg [[DBG80:![0-9]+]]
 // CHECK-DEBUG-NEXT:    br label [[OMP_PAR_OUTLINED_EXIT16:%.*]]
 // CHECK-DEBUG:       omp.par.outlined.exit16:
 // CHECK-DEBUG-NEXT:    br label [[OMP_PAR_EXIT_SPLIT:%.*]]
 // CHECK-DEBUG:       omp.par.exit.split:
-// CHECK-DEBUG-NEXT:    ret void, !dbg [[DBG80:![0-9]+]]
+// CHECK-DEBUG-NEXT:    ret void, !dbg [[DBG82:![0-9]+]]
 //
 void parallel_for_1(float *r, int a, double b) {
 #pragma omp parallel
   {
 #pragma omp parallel
     {
 #pragma omp for
       for (int i = 0; i < 100; ++i) {

 // CHECK-NEXT:    [[DOTCOUNT189:%.*]] = load i32, i32* [[DOTCOUNT_ADDR188]], align 4
 // CHECK-NEXT:    br label [[OMP_LOOP_PREHEADER190:%.*]]
 // CHECK:       omp_loop.preheader190:
 // CHECK-NEXT:    store i32 0, i32* [[P_LOWERBOUND204]], align 4
 // CHECK-NEXT:    [[TMP3:%.*]] = sub i32 [[DOTCOUNT189]], 1
 // CHECK-NEXT:    store i32 [[TMP3]], i32* [[P_UPPERBOUND205]], align 4
 // CHECK-NEXT:    store i32 1, i32* [[P_STRIDE206]], align 4
 // CHECK-NEXT:    [[OMP_GLOBAL_THREAD_NUM207:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB1]])
-// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM207]], i32 34, i32* [[P_LASTITER203]], i32* [[P_LOWERBOUND204]], i32* [[P_UPPERBOUND205]], i32* [[P_STRIDE206]], i32 1, i32 1)
+// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB1]], i32 [[OMP_GLOBAL_THREAD_NUM207]], i32 34, i32* [[P_LASTITER203]], i32* [[P_LOWERBOUND204]], i32* [[P_UPPERBOUND205]], i32* [[P_STRIDE206]], i32 1, i32 0)
 // CHECK-NEXT:    [[TMP4:%.*]] = load i32, i32* [[P_LOWERBOUND204]], align 4
 // CHECK-NEXT:    [[TMP5:%.*]] = load i32, i32* [[P_UPPERBOUND205]], align 4
 // CHECK-NEXT:    [[TMP6:%.*]] = sub i32 [[TMP5]], [[TMP4]]
 // CHECK-NEXT:    [[TMP7:%.*]] = add i32 [[TMP6]], 1
 // CHECK-NEXT:    br label [[OMP_LOOP_HEADER191:%.*]]
 // CHECK:       omp_loop.header191:
 // CHECK-NEXT:    [[OMP_LOOP_IV197:%.*]] = phi i32 [ 0, [[OMP_LOOP_PREHEADER190]] ], [ [[OMP_LOOP_NEXT199:%.*]], [[OMP_LOOP_INC194:%.*]] ]
 // CHECK-NEXT:    br label [[OMP_LOOP_COND192:%.*]]

 // CHECK-DEBUG-NEXT:    [[AGG_CAPTURED186:%.*]] = alloca [[STRUCT_ANON_17:%.*]], align 8
 // CHECK-DEBUG-NEXT:    [[AGG_CAPTURED187:%.*]] = alloca [[STRUCT_ANON_18:%.*]], align 4
 // CHECK-DEBUG-NEXT:    [[DOTCOUNT_ADDR188:%.*]] = alloca i32, align 4
 // CHECK-DEBUG-NEXT:    [[P_LASTITER203:%.*]] = alloca i32, align 4
 // CHECK-DEBUG-NEXT:    [[P_LOWERBOUND204:%.*]] = alloca i32, align 4
 // CHECK-DEBUG-NEXT:    [[P_UPPERBOUND205:%.*]] = alloca i32, align 4
 // CHECK-DEBUG-NEXT:    [[P_STRIDE206:%.*]] = alloca i32, align 4
 // CHECK-DEBUG-NEXT:    store float* [[R:%.*]], float** [[R_ADDR]], align 8
-// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata float** [[R_ADDR]], metadata [[META132:![0-9]+]], metadata !DIExpression()), !dbg [[DBG133:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata float** [[R_ADDR]], metadata [[META134:![0-9]+]], metadata !DIExpression()), !dbg [[DBG135:![0-9]+]]
 // CHECK-DEBUG-NEXT:    store i32 [[A:%.*]], i32* [[A_ADDR]], align 4
-// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata i32* [[A_ADDR]], metadata [[META134:![0-9]+]], metadata !DIExpression()), !dbg [[DBG135:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata i32* [[A_ADDR]], metadata [[META136:![0-9]+]], metadata !DIExpression()), !dbg [[DBG137:![0-9]+]]
 // CHECK-DEBUG-NEXT:    store double [[B:%.*]], double* [[B_ADDR]], align 8
-// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata double* [[B_ADDR]], metadata [[META136:![0-9]+]], metadata !DIExpression()), !dbg [[DBG137:![0-9]+]]
-// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB13:[0-9]+]]), !dbg [[DBG138:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata double* [[B_ADDR]], metadata [[META138:![0-9]+]], metadata !DIExpression()), !dbg [[DBG139:![0-9]+]]
+// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB13:[0-9]+]]), !dbg [[DBG140:![0-9]+]]
 // CHECK-DEBUG-NEXT:    br label [[OMP_PARALLEL:%.*]]
 // CHECK-DEBUG:       omp_parallel:
-// CHECK-DEBUG-NEXT:    call void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* @[[GLOB13]], i32 3, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i32*, double*, float**)* @_Z14parallel_for_2Pfid..omp_par.23 to void (i32*, i32*, ...)*), i32* [[A_ADDR]], double* [[B_ADDR]], float** [[R_ADDR]]), !dbg [[DBG139:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void (%struct.ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%struct.ident_t* @[[GLOB13]], i32 3, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i32*, double*, float**)* @_Z14parallel_for_2Pfid..omp_par.23 to void (i32*, i32*, ...)*), i32* [[A_ADDR]], double* [[B_ADDR]], float** [[R_ADDR]]), !dbg [[DBG141:![0-9]+]]
 // CHECK-DEBUG-NEXT:    br label [[OMP_PAR_OUTLINED_EXIT184:%.*]]
 // CHECK-DEBUG:       omp.par.outlined.exit184:
 // CHECK-DEBUG-NEXT:    br label [[OMP_PAR_EXIT_SPLIT:%.*]]
 // CHECK-DEBUG:       omp.par.exit.split:
-// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata i32* [[I185]], metadata [[META143:![0-9]+]], metadata !DIExpression()), !dbg [[DBG146:![0-9]+]]
-// CHECK-DEBUG-NEXT:    store i32 0, i32* [[I185]], align 4, !dbg [[DBG146]]
-// CHECK-DEBUG-NEXT:    [[TMP0:%.*]] = getelementptr inbounds [[STRUCT_ANON_17]], %struct.anon.17* [[AGG_CAPTURED186]], i32 0, i32 0, !dbg [[DBG147:![0-9]+]]
-// CHECK-DEBUG-NEXT:    store i32* [[I185]], i32** [[TMP0]], align 8, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[TMP1:%.*]] = getelementptr inbounds [[STRUCT_ANON_18]], %struct.anon.18* [[AGG_CAPTURED187]], i32 0, i32 0, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[TMP2:%.*]] = load i32, i32* [[I185]], align 4, !dbg [[DBG148:![0-9]+]]
-// CHECK-DEBUG-NEXT:    store i32 [[TMP2]], i32* [[TMP1]], align 4, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    call void @__captured_stmt.19(i32* [[DOTCOUNT_ADDR188]], %struct.anon.17* [[AGG_CAPTURED186]]), !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[DOTCOUNT189:%.*]] = load i32, i32* [[DOTCOUNT_ADDR188]], align 4, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_PREHEADER190:%.*]], !dbg [[DBG147]]
+// CHECK-DEBUG-NEXT:    call void @llvm.dbg.declare(metadata i32* [[I185]], metadata [[META145:![0-9]+]], metadata !DIExpression()), !dbg [[DBG148:![0-9]+]]
+// CHECK-DEBUG-NEXT:    store i32 0, i32* [[I185]], align 4, !dbg [[DBG148]]
+// CHECK-DEBUG-NEXT:    [[TMP0:%.*]] = getelementptr inbounds [[STRUCT_ANON_17]], %struct.anon.17* [[AGG_CAPTURED186]], i32 0, i32 0, !dbg [[DBG149:![0-9]+]]
+// CHECK-DEBUG-NEXT:    store i32* [[I185]], i32** [[TMP0]], align 8, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[TMP1:%.*]] = getelementptr inbounds [[STRUCT_ANON_18]], %struct.anon.18* [[AGG_CAPTURED187]], i32 0, i32 0, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[TMP2:%.*]] = load i32, i32* [[I185]], align 4, !dbg [[DBG150:![0-9]+]]
+// CHECK-DEBUG-NEXT:    store i32 [[TMP2]], i32* [[TMP1]], align 4, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    call void @__captured_stmt.19(i32* [[DOTCOUNT_ADDR188]], %struct.anon.17* [[AGG_CAPTURED186]]), !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[DOTCOUNT189:%.*]] = load i32, i32* [[DOTCOUNT_ADDR188]], align 4, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_PREHEADER190:%.*]], !dbg [[DBG149]]
 // CHECK-DEBUG:       omp_loop.preheader190:
-// CHECK-DEBUG-NEXT:    store i32 0, i32* [[P_LOWERBOUND204]], align 4, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[TMP3:%.*]] = sub i32 [[DOTCOUNT189]], 1, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    store i32 [[TMP3]], i32* [[P_UPPERBOUND205]], align 4, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    store i32 1, i32* [[P_STRIDE206]], align 4, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM207:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB42:[0-9]+]]), !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB42]], i32 [[OMP_GLOBAL_THREAD_NUM207]], i32 34, i32* [[P_LASTITER203]], i32* [[P_LOWERBOUND204]], i32* [[P_UPPERBOUND205]], i32* [[P_STRIDE206]], i32 1, i32 1), !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[TMP4:%.*]] = load i32, i32* [[P_LOWERBOUND204]], align 4, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[TMP5:%.*]] = load i32, i32* [[P_UPPERBOUND205]], align 4, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[TMP6:%.*]] = sub i32 [[TMP5]], [[TMP4]], !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[TMP7:%.*]] = add i32 [[TMP6]], 1, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_HEADER191:%.*]], !dbg [[DBG147]]
+// CHECK-DEBUG-NEXT:    store i32 0, i32* [[P_LOWERBOUND204]], align 4, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[TMP3:%.*]] = sub i32 [[DOTCOUNT189]], 1, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    store i32 [[TMP3]], i32* [[P_UPPERBOUND205]], align 4, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    store i32 1, i32* [[P_STRIDE206]], align 4, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM207:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB42:[0-9]+]]), !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @[[GLOB42]], i32 [[OMP_GLOBAL_THREAD_NUM207]], i32 34, i32* [[P_LASTITER203]], i32* [[P_LOWERBOUND204]], i32* [[P_UPPERBOUND205]], i32* [[P_STRIDE206]], i32 1, i32 0), !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[TMP4:%.*]] = load i32, i32* [[P_LOWERBOUND204]], align 4, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[TMP5:%.*]] = load i32, i32* [[P_UPPERBOUND205]], align 4, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[TMP6:%.*]] = sub i32 [[TMP5]], [[TMP4]], !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[TMP7:%.*]] = add i32 [[TMP6]], 1, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_HEADER191:%.*]], !dbg [[DBG149]]
 // CHECK-DEBUG:       omp_loop.header191:
-// CHECK-DEBUG-NEXT:    [[OMP_LOOP_IV197:%.*]] = phi i32 [ 0, [[OMP_LOOP_PREHEADER190]] ], [ [[OMP_LOOP_NEXT199:%.*]], [[OMP_LOOP_INC194:%.*]] ], !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_COND192:%.*]], !dbg [[DBG147]]
+// CHECK-DEBUG-NEXT:    [[OMP_LOOP_IV197:%.*]] = phi i32 [ 0, [[OMP_LOOP_PREHEADER190]] ], [ [[OMP_LOOP_NEXT199:%.*]], [[OMP_LOOP_INC194:%.*]] ], !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_COND192:%.*]], !dbg [[DBG149]]
 // CHECK-DEBUG:       omp_loop.cond192:
-// CHECK-DEBUG-NEXT:    [[OMP_LOOP_CMP198:%.*]] = icmp ult i32 [[OMP_LOOP_IV197]], [[TMP7]], !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    br i1 [[OMP_LOOP_CMP198]], label [[OMP_LOOP_BODY193:%.*]], label [[OMP_LOOP_EXIT195:%.*]], !dbg [[DBG147]]
+// CHECK-DEBUG-NEXT:    [[OMP_LOOP_CMP198:%.*]] = icmp ult i32 [[OMP_LOOP_IV197]], [[TMP7]], !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    br i1 [[OMP_LOOP_CMP198]], label [[OMP_LOOP_BODY193:%.*]], label [[OMP_LOOP_EXIT195:%.*]], !dbg [[DBG149]]
 // CHECK-DEBUG:       omp_loop.body193:
-// CHECK-DEBUG-NEXT:    [[TMP8:%.*]] = add i32 [[OMP_LOOP_IV197]], [[TMP4]], !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    call void @__captured_stmt.20(i32* [[I185]], i32 [[TMP8]], %struct.anon.18* [[AGG_CAPTURED187]]), !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[TMP9:%.*]] = load i32, i32* [[A_ADDR]], align 4, !dbg [[DBG149:![0-9]+]]
-// CHECK-DEBUG-NEXT:    [[CONV200:%.*]] = sitofp i32 [[TMP9]] to double, !dbg [[DBG149]]
-// CHECK-DEBUG-NEXT:    [[TMP10:%.*]] = load double, double* [[B_ADDR]], align 8, !dbg [[DBG150:![0-9]+]]
-// CHECK-DEBUG-NEXT:    [[ADD201:%.*]] = fadd double [[CONV200]], [[TMP10]], !dbg [[DBG151:![0-9]+]]
-// CHECK-DEBUG-NEXT:    [[CONV202:%.*]] = fptrunc double [[ADD201]] to float, !dbg [[DBG149]]
-// CHECK-DEBUG-NEXT:    [[TMP11:%.*]] = load float*, float** [[R_ADDR]], align 8, !dbg [[DBG152:![0-9]+]]
-// CHECK-DEBUG-NEXT:    store float [[CONV202]], float* [[TMP11]], align 4, !dbg [[DBG153:![0-9]+]]
-// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_INC194]], !dbg [[DBG147]]
+// CHECK-DEBUG-NEXT:    [[TMP8:%.*]] = add i32 [[OMP_LOOP_IV197]], [[TMP4]], !dbg [[DBG151:![0-9]+]]
+// CHECK-DEBUG-NEXT:    call void @__captured_stmt.20(i32* [[I185]], i32 [[TMP8]], %struct.anon.18* [[AGG_CAPTURED187]]), !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[TMP9:%.*]] = load i32, i32* [[A_ADDR]], align 4, !dbg [[DBG152:![0-9]+]]
+// CHECK-DEBUG-NEXT:    [[CONV200:%.*]] = sitofp i32 [[TMP9]] to double, !dbg [[DBG152]]
+// CHECK-DEBUG-NEXT:    [[TMP10:%.*]] = load double, double* [[B_ADDR]], align 8, !dbg [[DBG151]]
+// CHECK-DEBUG-NEXT:    [[ADD201:%.*]] = fadd double [[CONV200]], [[TMP10]], !dbg [[DBG153:![0-9]+]]
+// CHECK-DEBUG-NEXT:    [[CONV202:%.*]] = fptrunc double [[ADD201]] to float, !dbg [[DBG152]]
+// CHECK-DEBUG-NEXT:    [[TMP11:%.*]] = load float*, float** [[R_ADDR]], align 8, !dbg [[DBG154:![0-9]+]]
+// CHECK-DEBUG-NEXT:    store float [[CONV202]], float* [[TMP11]], align 4, !dbg [[DBG155:![0-9]+]]
+// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_INC194]], !dbg [[DBG149]]
 // CHECK-DEBUG:       omp_loop.inc194:
-// CHECK-DEBUG-NEXT:    [[OMP_LOOP_NEXT199]] = add nuw i32 [[OMP_LOOP_IV197]], 1, !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_HEADER191]], !dbg [[DBG147]]
+// CHECK-DEBUG-NEXT:    [[OMP_LOOP_NEXT199]] = add nuw i32 [[OMP_LOOP_IV197]], 1, !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_HEADER191]], !dbg [[DBG149]]
 // CHECK-DEBUG:       omp_loop.exit195:
-// CHECK-DEBUG-NEXT:    call void @__kmpc_for_static_fini(%struct.ident_t* @[[GLOB42]], i32 [[OMP_GLOBAL_THREAD_NUM207]]), !dbg [[DBG147]]
-// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM208:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB42]]), !dbg [[DBG150]]
-// CHECK-DEBUG-NEXT:    call void @__kmpc_barrier(%struct.ident_t* @[[GLOB43:[0-9]+]], i32 [[OMP_GLOBAL_THREAD_NUM208]]), !dbg [[DBG150]]
-// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_AFTER196:%.*]], !dbg [[DBG147]]
+// CHECK-DEBUG-NEXT:    call void @__kmpc_for_static_fini(%struct.ident_t* @[[GLOB42]], i32 [[OMP_GLOBAL_THREAD_NUM207]]), !dbg [[DBG149]]
+// CHECK-DEBUG-NEXT:    [[OMP_GLOBAL_THREAD_NUM208:%.*]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @[[GLOB42]]), !dbg [[DBG151]]
+// CHECK-DEBUG-NEXT:    call void @__kmpc_barrier(%struct.ident_t* @[[GLOB43:[0-9]+]], i32 [[OMP_GLOBAL_THREAD_NUM208]]), !dbg [[DBG151]]
+// CHECK-DEBUG-NEXT:    br label [[OMP_LOOP_AFTER196:%.*]], !dbg [[DBG149]]
 // CHECK-DEBUG:       omp_loop.after196:
-// CHECK-DEBUG-NEXT:    ret void, !dbg [[DBG154:![0-9]+]]
+// CHECK-DEBUG-NEXT:    ret void, !dbg [[DBG156:![0-9]+]]
 //
 void parallel_for_2(float *r, int a, double b) {
 #pragma omp parallel
   {
 #pragma omp for
     for (int i = 0; i < 100; ++i)
       *r = a + b;
 #pragma omp parallel

clang/test/OpenMP/irbuilder_unroll_partial_factor_for.c

 // CHECK-NEXT:    br label %[[OMP_FLOOR0_PREHEADER:.+]]
 // CHECK-EMPTY:
 // CHECK-NEXT:  [[OMP_FLOOR0_PREHEADER]]:
 // CHECK-NEXT:    store i32 0, i32* %[[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP8:.+]] = sub i32 %[[OMP_FLOOR0_TRIPCOUNT]], 1
 // CHECK-NEXT:    store i32 %[[TMP8]], i32* %[[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    store i32 1, i32* %[[P_STRIDE]], align 4
 // CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
-// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]], i32 1, i32 1)
+// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]], i32 1, i32 0)
 // CHECK-NEXT:    %[[TMP9:.+]] = load i32, i32* %[[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP10:.+]] = load i32, i32* %[[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP11:.+]] = sub i32 %[[TMP10]], %[[TMP9]]
 // CHECK-NEXT:    %[[TMP12:.+]] = add i32 %[[TMP11]], 1
 // CHECK-NEXT:    br label %[[OMP_FLOOR0_HEADER:.+]]
 // CHECK-EMPTY:
 // CHECK-NEXT:  [[OMP_FLOOR0_HEADER]]:
 // CHECK-NEXT:    %[[OMP_FLOOR0_IV:.+]] = phi i32 [ 0, %[[OMP_FLOOR0_PREHEADER]] ], [ %[[OMP_FLOOR0_NEXT:.+]], %[[OMP_FLOOR0_INC:.+]] ]

clang/test/OpenMP/irbuilder_unroll_partial_heuristic_constant_for.c

 // CHECK-NEXT:    br label %[[OMP_FLOOR0_PREHEADER:.+]]
 // CHECK-EMPTY:
 // CHECK-NEXT:  [[OMP_FLOOR0_PREHEADER]]:
 // CHECK-NEXT:    store i32 0, i32* %[[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP7:.+]] = sub i32 %[[OMP_FLOOR0_TRIPCOUNT]], 1
 // CHECK-NEXT:    store i32 %[[TMP7]], i32* %[[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    store i32 1, i32* %[[P_STRIDE]], align 4
 // CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
-// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]], i32 1, i32 1)
+// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]], i32 1, i32 0)
 // CHECK-NEXT:    %[[TMP8:.+]] = load i32, i32* %[[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP9:.+]] = load i32, i32* %[[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP10:.+]] = sub i32 %[[TMP9]], %[[TMP8]]
 // CHECK-NEXT:    %[[TMP11:.+]] = add i32 %[[TMP10]], 1
 // CHECK-NEXT:    br label %[[OMP_FLOOR0_HEADER:.+]]
 // CHECK-EMPTY:
 // CHECK-NEXT:  [[OMP_FLOOR0_HEADER]]:
 // CHECK-NEXT:    %[[OMP_FLOOR0_IV:.+]] = phi i32 [ 0, %[[OMP_FLOOR0_PREHEADER]] ], [ %[[OMP_FLOOR0_NEXT:.+]], %[[OMP_FLOOR0_INC:.+]] ]

clang/test/OpenMP/irbuilder_unroll_partial_heuristic_runtime_for.c

 // CHECK-NEXT:    br label %[[OMP_FLOOR0_PREHEADER:.+]]
 // CHECK-EMPTY:
 // CHECK-NEXT:  [[OMP_FLOOR0_PREHEADER]]:
 // CHECK-NEXT:    store i32 0, i32* %[[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP8:.+]] = sub i32 %[[OMP_FLOOR0_TRIPCOUNT]], 1
 // CHECK-NEXT:    store i32 %[[TMP8]], i32* %[[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    store i32 1, i32* %[[P_STRIDE]], align 4
 // CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
-// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]], i32 1, i32 1)
+// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]], i32 1, i32 0)
 // CHECK-NEXT:    %[[TMP9:.+]] = load i32, i32* %[[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP10:.+]] = load i32, i32* %[[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP11:.+]] = sub i32 %[[TMP10]], %[[TMP9]]
 // CHECK-NEXT:    %[[TMP12:.+]] = add i32 %[[TMP11]], 1
 // CHECK-NEXT:    br label %[[OMP_FLOOR0_HEADER:.+]]
 // CHECK-EMPTY:
 // CHECK-NEXT:  [[OMP_FLOOR0_HEADER]]:
 // CHECK-NEXT:    %[[OMP_FLOOR0_IV:.+]] = phi i32 [ 0, %[[OMP_FLOOR0_PREHEADER]] ], [ %[[OMP_FLOOR0_NEXT:.+]], %[[OMP_FLOOR0_INC:.+]] ]

clang/test/OpenMP/irbuilder_unroll_unroll_partial_factor.c

 // CHECK-NEXT:    br label %[[OMP_FLOOR0_PREHEADER:.+]]
 // CHECK-EMPTY:
 // CHECK-NEXT:  [[OMP_FLOOR0_PREHEADER]]:
 // CHECK-NEXT:    store i32 0, i32* %[[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP7:.+]] = sub i32 %[[OMP_FLOOR0_TRIPCOUNT]], 1
 // CHECK-NEXT:    store i32 %[[TMP7]], i32* %[[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    store i32 1, i32* %[[P_STRIDE]], align 4
 // CHECK-NEXT:    %[[OMP_GLOBAL_THREAD_NUM:.+]] = call i32 @__kmpc_global_thread_num(%struct.ident_t* @1)
-// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]], i32 1, i32 1)
+// CHECK-NEXT:    call void @__kmpc_for_static_init_4u(%struct.ident_t* @1, i32 %[[OMP_GLOBAL_THREAD_NUM]], i32 34, i32* %[[P_LASTITER]], i32* %[[P_LOWERBOUND]], i32* %[[P_UPPERBOUND]], i32* %[[P_STRIDE]], i32 1, i32 0)
 // CHECK-NEXT:    %[[TMP8:.+]] = load i32, i32* %[[P_LOWERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP9:.+]] = load i32, i32* %[[P_UPPERBOUND]], align 4
 // CHECK-NEXT:    %[[TMP10:.+]] = sub i32 %[[TMP9]], %[[TMP8]]
 // CHECK-NEXT:    %[[TMP11:.+]] = add i32 %[[TMP10]], 1
 // CHECK-NEXT:    br label %[[OMP_FLOOR0_HEADER:.+]]
 // CHECK-EMPTY:
 // CHECK-NEXT:  [[OMP_FLOOR0_HEADER]]:
 // CHECK-NEXT:    %[[OMP_FLOOR0_IV:.+]] = phi i32 [ 0, %[[OMP_FLOOR0_PREHEADER]] ], [ %[[OMP_FLOOR0_NEXT:.+]], %[[OMP_FLOOR0_INC:.+]] ]

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

public:
   ///
   /// This takes a \p LoopInfo representing a canonical loop, such as the one
   /// created by \p createCanonicalLoop and emits additional instructions to
   /// turn it into a workshare loop. In particular, it calls to an OpenMP
   /// runtime function in the preheader to obtain the loop bounds to be used in
   /// the current thread, updates the relevant instructions in the canonical
   /// loop and calls to an OpenMP runtime finalization function after the loop.
   ///
-  /// TODO: Workshare loops with static scheduling may contain up to two loops
-  /// that fulfill the requirements of an OpenMP canonical loop. One for
-  /// iterating over all iterations of a chunk and another one for iterating
-  /// over all chunks that are executed on the same thread. Returning
-  /// CanonicalLoopInfo objects representing them may eventually be useful for
-  /// the apply clause planned in OpenMP 6.0, but currently whether these are
-  /// canonical loops is irrelevant.
-  ///
   /// \param DL       Debug location for instructions added for the
   ///                 workshare-loop construct itself.
   /// \param CLI      A descriptor of the canonical loop to workshare.
   /// \param AllocaIP An insertion point for Alloca instructions usable in the
   ///                 preheader of the loop.
   /// \param NeedsBarrier Indicates whether a barrier must be inserted after
   ///                     the loop.
-  /// \param Chunk    The size of loop chunk considered as a unit when
-  ///                 scheduling. If \p nullptr, defaults to 1.
   ///
   /// \returns Point where to insert code after the workshare construct.
   InsertPointTy applyStaticWorkshareLoop(DebugLoc DL, CanonicalLoopInfo *CLI,
                                          InsertPointTy AllocaIP,
+                                         bool NeedsBarrier);
+
+  /// Modifies the canonical loop a statically-scheduled workshare loop with a
+  /// user-specified chunk size.
+  ///
+  /// \param DL           Debug location for instructions added for the
+  ///                     workshare-loop construct itself.
+  /// \param CLI          A descriptor of the canonical loop to workshare.
+  /// \param AllocaIP     An insertion point for Alloca instructions usable in
+  ///                     the preheader of the loop.
+  /// \param NeedsBarrier Indicates whether a barrier must be inserted after the
+  ///                     loop.
+  /// \param ChunkSize    The user-specified chunk size.
+  ///
+  /// \returns Point where to insert code after the workshare construct.
+  InsertPointTy applyStaticChunkedWorkshareLoop(DebugLoc DL,
+                                                CanonicalLoopInfo *CLI,
+                                                InsertPointTy AllocaIP,
                                                 bool NeedsBarrier,
-                                         Value *Chunk = nullptr);
+                                                Value *ChunkSize);
 
   /// Modifies the canonical loop to be a dynamically-scheduled workshare loop.
   ///
   /// This takes a \p LoopInfo representing a canonical loop, such as the one
   /// created by \p createCanonicalLoop and emits additional instructions to
   /// turn it into a workshare loop. In particular, it calls to an OpenMP
   /// runtime function in the preheader to obtain, and then in each iteration
   /// to update the loop counter.
public:
   ///
   /// This takes a \p LoopInfo representing a canonical loop, such as the one
   /// created by \p createCanonicalLoop and emits additional instructions to
   /// turn it into a workshare loop. In particular, it calls to an OpenMP
   /// runtime function in the preheader to obtain the loop bounds to be used in
   /// the current thread, updates the relevant instructions in the canonical
   /// loop and calls to an OpenMP runtime finalization function after the loop.
   ///
+  /// The concrete transformation is done by applyStaticWorkshareLoop,
+  /// applyStaticChunkedWorkshareLoop, or applyDynamicWorkshareLoop, depending
+  /// on the value of \p SchedKind and \p ChunkSize.
+  ///
   /// \param DL       Debug location for instructions added for the
   ///                 workshare-loop construct itself.
   /// \param CLI      A descriptor of the canonical loop to workshare.
   /// \param AllocaIP An insertion point for Alloca instructions usable in the
   ///                 preheader of the loop.
   /// \param NeedsBarrier Indicates whether a barrier must be insterted after
   ///                     the loop.
+  /// \param SchedKind Scheduling algorithm to use.
+  /// \param ChunkSize The chunk size for the inner loop.
   ///
   /// \returns Point where to insert code after the workshare construct.
-  InsertPointTy applyWorkshareLoop(DebugLoc DL, CanonicalLoopInfo *CLI,
-                                   InsertPointTy AllocaIP, bool NeedsBarrier);
+  InsertPointTy applyWorkshareLoop(
+      DebugLoc DL, CanonicalLoopInfo *CLI, InsertPointTy AllocaIP,
+      bool NeedsBarrier,
+      llvm::omp::ScheduleKind SchedKind = llvm::omp::OMP_SCHEDULE_Default,
+      Value *ChunkSize = nullptr);
 
   /// Tile a loop nest.
   ///
   /// Tiles the loops of \p Loops by the tile sizes in \p TileSizes. Loops in
   /// \p/ Loops must be perfectly nested, from outermost to innermost loop
   /// (i.e. Loops.front() is the outermost loop). The trip count llvm::Value
   /// of every loop and every tile sizes must be usable in the outermost
   /// loop's preheader. This implies that the loop nest is rectangular.
private:
   /// This does not include any block from the body, including the one returned
   /// by getBody().
   ///
   /// FIXME: This currently includes the Preheader and After blocks even though
   /// their content is (mostly) not under CanonicalLoopInfo's control.
   /// Re-evaluated whether this makes sense.
   void collectControlBlocks(SmallVectorImpl<BasicBlock *> &BBs);
 
+  /// Sets the number of loop iterations to the given value. This value must be
+  /// valid in the condition block (i.e., defined in the preheader) and is
+  /// interpreted as an unsigned integer.
+  void setTripCount(Value *TripCount);
+
+  /// Replace all uses of the canonical induction variable in the loop body with
+  /// a new one.
+  ///
+  /// The intended use case is to update the induction variable for an updated
+  /// iteration space such that it can stay normalized in the 0...tripcount-1
+  /// range.
+  ///
+  /// The \p Updater is called with the (presumable updated) current normalized
+  /// induction variable and is expected to return the value that uses of the
+  /// pre-updated induction values should use instead, typically dependent on
+  /// the new induction variable. This is a lambda (instead of e.g. just passing
+  /// the new value) to be able to distinguish the uses of the pre-updated
+  /// induction variable and uses of the induction varible to compute the
+  /// updated induction variable value.
+  void mapIndVar(llvm::function_ref<Value *(Instruction *)> Updater);
+
 public:
   /// Returns whether this object currently represents the IR of a loop. If
   /// returning false, it may have been consumed by a loop transformation or not
   /// been intialized. Do not use in this case;
   bool isValid() const { return Header; }
 
   /// The preheader ensures that there is only a single edge entering the loop.
   /// Code that must be execute before any loop iteration can be emitted here,

llvm/lib/Frontend/OpenMP/OMPIRBuilder.cpp

OptimisticAttributes("openmp-ir-builder-optimistic-attributes", cl::Hidden,
                          cl::init(false));
 
 static cl::opt<double> UnrollThresholdFactor(
     "openmp-ir-builder-unroll-threshold-factor", cl::Hidden,
     cl::desc("Factor for the unroll threshold to account for code "
              "simplifications still taking place"),
     cl::init(1.5));
 
+/// Make \p Source branch to \p Target.
+///
+/// Handles two situations:
+/// * \p Source already has an unconditional branch.
+/// * \p Source is a degenerate block (no terminator because the BB is
+///             the current head of the IR construction).
+static void redirectTo(BasicBlock *Source, BasicBlock *Target, DebugLoc DL) {
+  if (Instruction *Term = Source->getTerminator()) {
+    auto *Br = cast<BranchInst>(Term);
+    assert(!Br->isConditional() &&
+           "BB's terminator must be an unconditional branch (or degenerate)");
+    BasicBlock *Succ = Br->getSuccessor(0);
+    Succ->removePredecessor(Source, /*KeepOneInputPHIs=*/true);
+    Br->setSuccessor(0, Target);
+    return;
+  }
+
+  auto *NewBr = BranchInst::Create(Target, Source);
+  NewBr->setDebugLoc(DL);
+}
+
+/// Move the instruction after an InsertPoint to the beginning of another
+/// BasicBlock.
+///
+/// The instructions after \p IP are moved to the beginning of \p New which must
+/// not have any PHINodes. If \p CreateBranch is true, a branch instruction to
+/// \p New will be added such that there is no semantic change. Otherwise, the
+/// \p IP insert block remains degenerate and it is up to the caller to insert a
+/// terminator.
+static void spliceBB(OpenMPIRBuilder::InsertPointTy IP, BasicBlock *New,
+                     bool CreateBranch) {
+  assert(New->getFirstInsertionPt() == New->begin() &&
+         "Target BB must not have PHI nodes");
+
+  // Move instructions to new block.
+  BasicBlock *Old = IP.getBlock();
+  New->getInstList().splice(New->begin(), Old->getInstList(), IP.getPoint(),
+                            Old->end());
+
+  if (CreateBranch)
+    BranchInst::Create(New, Old);
+}
+
+/// Splice a BasicBlock at an IRBuilder's current insertion point. Its new
+/// insert location will stick to after the instruction before the insertion
+/// point (instead of moving with the instruction the InsertPoint stores
+/// internally).
+static void spliceBB(IRBuilder<> &Builder, BasicBlock *New, bool CreateBranch) {
+  DebugLoc DebugLoc = Builder.getCurrentDebugLocation();
+  BasicBlock *Old = Builder.GetInsertBlock();
+
+  spliceBB(Builder.saveIP(), New, CreateBranch);
+  if (CreateBranch)
+    Builder.SetInsertPoint(Old->getTerminator());
+  else
+    Builder.SetInsertPoint(Old);
+
+  // SetInsertPoint also updates the Builder's debug location, but we want to
+  // keep the one the Builder was configured to use.
+  Builder.SetCurrentDebugLocation(DebugLoc);
+}
+
+/// Split a BasicBlock at an InsertPoint, even if the block is degenerate
+/// (missing the terminator).
+///
+/// llvm::SplitBasicBlock and BasicBlock::splitBasicBlock require a well-formed
+/// BasicBlock. \p Name is used for the new successor block. If \p CreateBranch
+/// is true, a branch to the new successor will new created such that
+/// semantically there is no change; otherwise the block of the insertion point
+/// remains degenerate and it is the caller's responsibility to insert a
+/// terminator. Returns the new successor block.
+static BasicBlock *splitBB(OpenMPIRBuilder::InsertPointTy IP, bool CreateBranch,
+                           llvm::Twine Name = {}) {
+  BasicBlock *Old = IP.getBlock();
+  BasicBlock *New = BasicBlock::Create(
+      Old->getContext(), Name.isTriviallyEmpty() ? Old->getName() : Name,
+      Old->getParent(), Old->getNextNode());
+  spliceBB(IP, New, CreateBranch);
+  New->replaceSuccessorsPhiUsesWith(Old, New);
+  return New;
+}
+
+/// Split a BasicBlock at \p Builder's insertion point, even if the block is
+/// degenerate (missing the terminator).  Its new insert location will stick to
+/// after the instruction before the insertion point (instead of moving with the
+/// instruction the InsertPoint stores internally).
+static BasicBlock *splitBB(IRBuilder<> &Builder, bool CreateBranch,
+                           llvm::Twine Name = {}) {
+  DebugLoc DebugLoc = Builder.getCurrentDebugLocation();
+  BasicBlock *New = splitBB(Builder.saveIP(), CreateBranch, Name);
+  if (CreateBranch)
+    Builder.SetInsertPoint(Builder.GetInsertBlock()->getTerminator());
+  else
+    Builder.SetInsertPoint(Builder.GetInsertBlock());
+  // SetInsertPoint also updates the Builder's debug location, but we want to
+  // keep the one the Builder was configured to use.
+  Builder.SetCurrentDebugLocation(DebugLoc);
+  return New;
+}
+
 void OpenMPIRBuilder::addAttributes(omp::RuntimeFunction FnID, Function &Fn) {
   LLVMContext &Ctx = Fn.getContext();
 
   // Get the function's current attributes.
   auto Attrs = Fn.getAttributes();
   auto FnAttrs = Attrs.getFnAttrs();
   auto RetAttrs = Attrs.getRetAttrs();
   SmallVector<AttributeSet, 4> ArgAttrs;
CanonicalLoopInfo *CL = createLoopSkeleton(Loc.DL, TripCount, BB->getParent(),
                                              NextBB, NextBB, Name);
   BasicBlock *After = CL->getAfter();
 
   // If location is not set, don't connect the loop.
   if (updateToLocation(Loc)) {
     // Split the loop at the insertion point: Branch to the preheader and move
     // every following instruction to after the loop (the After BB). Also, the
     // new successor is the loop's after block.
+    spliceBB(Builder, After, /*CreateBranch=*/false);
     Builder.CreateBr(CL->getPreheader());
-    After->getInstList().splice(After->begin(), BB->getInstList(),
-                                Builder.GetInsertPoint(), BB->end());
-    After->replaceSuccessorsPhiUsesWith(BB, After);
   }
 
   // Emit the body content. We do it after connecting the loop to the CFG to
   // avoid that the callback encounters degenerate BBs.
   BodyGenCB(CL->getBodyIP(), CL->getIndVar());
 
 #ifndef NDEBUG
   CL->assertOK();
if (Bitwidth == 32)
     return OMPBuilder.getOrCreateRuntimeFunction(
         M, omp::RuntimeFunction::OMPRTL___kmpc_for_static_init_4u);
   if (Bitwidth == 64)
     return OMPBuilder.getOrCreateRuntimeFunction(
         M, omp::RuntimeFunction::OMPRTL___kmpc_for_static_init_8u);
   llvm_unreachable("unknown OpenMP loop iterator bitwidth");
 }
 
-// Sets the number of loop iterations to the given value. This value must be
-// valid in the condition block (i.e., defined in the preheader) and is
-// interpreted as an unsigned integer.
-void setCanonicalLoopTripCount(CanonicalLoopInfo *CLI, Value *TripCount) {
-  Instruction *CmpI = &CLI->getCond()->front();
-  assert(isa<CmpInst>(CmpI) && "First inst must compare IV with TripCount");
-  CmpI->setOperand(1, TripCount);
-  CLI->assertOK();
-}
-
 OpenMPIRBuilder::InsertPointTy
 OpenMPIRBuilder::applyStaticWorkshareLoop(DebugLoc DL, CanonicalLoopInfo *CLI,
                                           InsertPointTy AllocaIP,
-                                          bool NeedsBarrier, Value *Chunk) {
+                                          bool NeedsBarrier) {
   assert(CLI->isValid() && "Requires a valid canonical loop");
 
   // Set up the source location value for OpenMP runtime.
   Builder.restoreIP(CLI->getPreheaderIP());
   Builder.SetCurrentDebugLocation(DL);
 
   Constant *SrcLocStr = getOrCreateSrcLocStr(DL);
   Value *SrcLoc = getOrCreateIdent(SrcLocStr);
OpenMPIRBuilder::applyStaticWorkshareLoop(DebugLoc DL, CanonicalLoopInfo *CLI,
   Builder.SetInsertPoint(CLI->getPreheader()->getTerminator());
   Constant *Zero = ConstantInt::get(IVTy, 0);
   Constant *One = ConstantInt::get(IVTy, 1);
   Builder.CreateStore(Zero, PLowerBound);
   Value *UpperBound = Builder.CreateSub(CLI->getTripCount(), One);
   Builder.CreateStore(UpperBound, PUpperBound);
   Builder.CreateStore(One, PStride);
 
-  // FIXME: schedule(static) is NOT the same as schedule(static,1)
-  if (!Chunk)
-    Chunk = One;
-
   Value *ThreadNum = getOrCreateThreadID(SrcLoc);
 
   Constant *SchedulingType =
       ConstantInt::get(I32Type, static_cast<int>(OMPScheduleType::Static));
 
   // Call the "init" function and update the trip count of the loop with the
   // value it produced.
   Builder.CreateCall(StaticInit,
                      {SrcLoc, ThreadNum, SchedulingType, PLastIter, PLowerBound,
-                      PUpperBound, PStride, One, Chunk});
+                      PUpperBound, PStride, One, Zero});
   Value *LowerBound = Builder.CreateLoad(IVTy, PLowerBound);
   Value *InclusiveUpperBound = Builder.CreateLoad(IVTy, PUpperBound);
   Value *TripCountMinusOne = Builder.CreateSub(InclusiveUpperBound, LowerBound);
   Value *TripCount = Builder.CreateAdd(TripCountMinusOne, One);
-  setCanonicalLoopTripCount(CLI, TripCount);
+  CLI->setTripCount(TripCount);
 
   // Update all uses of the induction variable except the one in the condition
   // block that compares it with the actual upper bound, and the increment in
   // the latch block.
-  // TODO: this can eventually move to CanonicalLoopInfo or to a new
-  // CanonicalLoopInfoUpdater interface.
-  Builder.SetInsertPoint(CLI->getBody(), CLI->getBody()->getFirstInsertionPt());
-  Value *UpdatedIV = Builder.CreateAdd(IV, LowerBound);
-  IV->replaceUsesWithIf(UpdatedIV, [&](Use &U) {
-    auto *Instr = dyn_cast<Instruction>(U.getUser());
-    return !Instr ||
-           (Instr->getParent() != CLI->getCond() &&
-            Instr->getParent() != CLI->getLatch() && Instr != UpdatedIV);
+
+  CLI->mapIndVar([&](Instruction *OldIV) -> Value * {
+    Builder.SetInsertPoint(CLI->getBody(),
+                           CLI->getBody()->getFirstInsertionPt());
+    Builder.SetCurrentDebugLocation(DL);
+    return Builder.CreateAdd(OldIV, LowerBound);
   });
 
   // In the "exit" block, call the "fini" function.
   Builder.SetInsertPoint(CLI->getExit(),
                          CLI->getExit()->getTerminator()->getIterator());
   Builder.CreateCall(StaticFini, {SrcLoc, ThreadNum});
 
   // Add the barrier if requested.
   if (NeedsBarrier)
     createBarrier(LocationDescription(Builder.saveIP(), DL),
                   omp::Directive::OMPD_for, /* ForceSimpleCall */ false,
                   /* CheckCancelFlag */ false);
 
   InsertPointTy AfterIP = CLI->getAfterIP();
   CLI->invalidate();
 
   return AfterIP;
 }
 
+OpenMPIRBuilder::InsertPointTy OpenMPIRBuilder::applyStaticChunkedWorkshareLoop(
+    DebugLoc DL, CanonicalLoopInfo *CLI, InsertPointTy AllocaIP,
+    bool NeedsBarrier, Value *ChunkSize) {
+  assert(CLI->isValid() && "Requires a valid canonical loop");
+  assert(ChunkSize && "Chunk size is required");
+
+  Value *IV = CLI->getIndVar();
+  Value *OrigTripCount = CLI->getTripCount();
+  Type *IVTy = IV->getType();
+  Type *I32Type = Type::getInt32Ty(M.getContext());
+  Constant *Zero = ConstantInt::get(IVTy, 0);
+  Constant *One = ConstantInt::get(IVTy, 1);
+
+  // Declare useful OpenMP runtime functions.
+  FunctionCallee StaticInit = getKmpcForStaticInitForType(IVTy, M, *this);
+  FunctionCallee StaticFini =
+      getOrCreateRuntimeFunction(M, omp::OMPRTL___kmpc_for_static_fini);
+
+  // Allocate space for computed loop bounds as expected by the "init" function.
+  Builder.restoreIP(AllocaIP);
+  Builder.SetCurrentDebugLocation(DL);
+  Value *PLastIter = Builder.CreateAlloca(I32Type, nullptr, "p.lastiter");
+  Value *PLowerBound = Builder.CreateAlloca(IVTy, nullptr, "p.lowerbound");
+  Value *PUpperBound = Builder.CreateAlloca(IVTy, nullptr, "p.upperbound");
+  Value *PStride = Builder.CreateAlloca(IVTy, nullptr, "p.stride");
+
+  // Set up the source location value for the OpenMP runtime.
+  Builder.restoreIP(CLI->getPreheaderIP());
+  Builder.SetCurrentDebugLocation(DL);
+
+  // Call the "init" function and update the trip count of the loop with the
+  // value it produced.
+  Value *SrcLoc = getOrCreateIdent(getOrCreateSrcLocStr(DL));
+  Value *ThreadNum = getOrCreateThreadID(SrcLoc);
+  Constant *SchedulingType = ConstantInt::get(
+      I32Type, static_cast<int>(OMPScheduleType::StaticChunked));
+  Builder.CreateStore(Zero, PLowerBound);
+  Value *OrigUpperBound = Builder.CreateSub(OrigTripCount, One);
+  Builder.CreateStore(OrigUpperBound, PUpperBound);
+  Builder.CreateStore(One, PStride);
+  Builder.CreateCall(StaticInit,
+                     {/*loc=*/SrcLoc, /*global_tid=*/ThreadNum,
+                      /*schedtype=*/SchedulingType, /*plastiter=*/PLastIter,
+                      /*plower=*/PLowerBound, /*pupper=*/PUpperBound,
+                      /*pstride=*/PStride, /*incr=*/One, /*chunk=*/ChunkSize});
+
+  // Load values written by the "init" function.
+  Value *FirstChunkLB =
+      Builder.CreateLoad(IVTy, PLowerBound, "omp_firstchunk.lb");
+  Value *NextChunkStride =
+      Builder.CreateLoad(IVTy, PStride, "omp_dispatch.stride");
+
+  // Create outer "dispatch" loop for enumerating the chunks.
+  BasicBlock *DispatchEnter = splitBB(Builder, true);
+  Value *DispatchCounter;
+  CanonicalLoopInfo *DispatchCLI = createCanonicalLoop(
+      {Builder.saveIP(), DL},
+      [&](InsertPointTy BodyIP, Value *Counter) { DispatchCounter = Counter; },
+      FirstChunkLB, OrigTripCount, NextChunkStride,
+      /*IsSigned=*/false, /*InclusiveStop=*/false, /*ComputeIP=*/{},
+      "dispatch");
+
+  // Remember the BasicBlocks of the dispatch loop we need, then invalidate to
+  // not have to preserve the canonical invariant.
+  BasicBlock *DispatchBody = DispatchCLI->getBody();
+  BasicBlock *DispatchLatch = DispatchCLI->getLatch();
+  BasicBlock *DispatchExit = DispatchCLI->getExit();
+  BasicBlock *DispatchAfter = DispatchCLI->getAfter();
+  DispatchCLI->invalidate();
+
+  // Rewire the original loop to become the chunk loop inside the dispatch loop.
+  redirectTo(DispatchAfter, CLI->getAfter(), DL);
+  redirectTo(CLI->getExit(), DispatchLatch, DL);
+  redirectTo(DispatchBody, DispatchEnter, DL);
+
+  // Prepare the prolog of the chunk loop.
+  Builder.restoreIP(CLI->getPreheaderIP());
+  Builder.SetCurrentDebugLocation(DL);
+
+  // Compute the number of iterations of the chunk loop.
+  Builder.SetInsertPoint(CLI->getPreheader()->getTerminator());
+  Value *ChunkEnd = Builder.CreateAdd(DispatchCounter, ChunkSize);
+  Value *IsLastChunk =
+      Builder.CreateICmpUGE(ChunkEnd, OrigTripCount, "omp_chunk.is_last");
+  Value *CountUntilOrigTripCount =
+      Builder.CreateSub(OrigTripCount, DispatchCounter);
+  Value *ChunkTripCount = Builder.CreateSelect(
+      IsLastChunk, CountUntilOrigTripCount, ChunkSize, "omp_chunk.tripcount");
+  CLI->setTripCount(ChunkTripCount);
+
+  // Update all uses of the induction variable except the one in the condition
+  // block that compares it with the actual upper bound, and the increment in
+  // the latch block.
+  CLI->mapIndVar([&](Instruction *) -> Value * {
+    Builder.restoreIP(CLI->getBodyIP());
+    return Builder.CreateAdd(IV, DispatchCounter);
+  });
+
+  // In the "exit" block, call the "fini" function.
+  Builder.SetInsertPoint(DispatchExit, DispatchExit->getFirstInsertionPt());
+  Builder.CreateCall(StaticFini, {SrcLoc, ThreadNum});
+
+  // Add the barrier if requested.
+  if (NeedsBarrier)
+    createBarrier(LocationDescription(Builder.saveIP(), DL), OMPD_for,
+                  /* ForceSimpleCall */ false, /* CheckCancelFlag */ false);
+
+#ifndef NDEBUG
+  // Even though we currently do not support applying additional methods to it,
+  // the chunk loop should remain a canonical loop.
+  CLI->assertOK();
+#endif
+
+  return {DispatchAfter, DispatchAfter->getFirstInsertionPt()};
+}
+
 OpenMPIRBuilder::InsertPointTy
 OpenMPIRBuilder::applyWorkshareLoop(DebugLoc DL, CanonicalLoopInfo *CLI,
-                                    InsertPointTy AllocaIP, bool NeedsBarrier) {
-  // Currently only supports static schedules.
+                                    InsertPointTy AllocaIP, bool NeedsBarrier,
+                                    llvm::omp::ScheduleKind SchedKind,
+                                    llvm::Value *ChunkSize) {
+  switch (SchedKind) {
+  case llvm::omp::ScheduleKind ::OMP_SCHEDULE_Default:
+    assert(!ChunkSize && "No chunk size with default schedule (which for clang "
+                         "is static non-chunked)");
+    LLVM_FALLTHROUGH
+  case llvm::omp::ScheduleKind ::OMP_SCHEDULE_Static:
+    if (ChunkSize)
+      return applyStaticChunkedWorkshareLoop(DL, CLI, AllocaIP, NeedsBarrier,
+                                             ChunkSize);
     return applyStaticWorkshareLoop(DL, CLI, AllocaIP, NeedsBarrier);
+  case llvm::omp::ScheduleKind ::OMP_SCHEDULE_Auto:
+    assert(!ChunkSize && "Chunk size with auto scheduling not user-defined");
+    return applyDynamicWorkshareLoop(DL, CLI, AllocaIP, OMPScheduleType::Auto,
+                                     NeedsBarrier, nullptr);
+  case llvm::omp::ScheduleKind ::OMP_SCHEDULE_Dynamic:
+    return applyDynamicWorkshareLoop(DL, CLI, AllocaIP,
+                                     OMPScheduleType::DynamicChunked,
+                                     NeedsBarrier, ChunkSize);
+  case llvm::omp::ScheduleKind ::OMP_SCHEDULE_Guided:
+    return applyDynamicWorkshareLoop(DL, CLI, AllocaIP,
+                                     OMPScheduleType::GuidedChunked,
+                                     NeedsBarrier, ChunkSize);
+  case llvm::omp::ScheduleKind ::OMP_SCHEDULE_Runtime:
+    assert(!ChunkSize &&
+           "Chunk size with runtime scheduling implied to be one");
+    return applyDynamicWorkshareLoop(
+        DL, CLI, AllocaIP, OMPScheduleType::Runtime, NeedsBarrier, nullptr);
+  }
+
+  llvm_unreachable("Unknown/unimplemented schedule kind");
 }
 
 /// Returns an LLVM function to call for initializing loop bounds using OpenMP
 /// dynamic scheduling depending on `type`. Only i32 and i64 are supported by
 /// the runtime. Always interpret integers as unsigned similarly to
 /// CanonicalLoopInfo.
 static FunctionCallee
 getKmpcForDynamicInitForType(Type *Ty, Module &M, OpenMPIRBuilder &OMPBuilder) {
createBarrier(LocationDescription(Builder.saveIP(), DL),
                   omp::Directive::OMPD_for, /* ForceSimpleCall */ false,
                   /* CheckCancelFlag */ false);
   }
 
   CLI->invalidate();
   return AfterIP;
 }
 
-/// Make \p Source branch to \p Target.
-///
-/// Handles two situations:
-/// * \p Source already has an unconditional branch.
-/// * \p Source is a degenerate block (no terminator because the BB is
-///             the current head of the IR construction).
-static void redirectTo(BasicBlock *Source, BasicBlock *Target, DebugLoc DL) {
-  if (Instruction *Term = Source->getTerminator()) {
-    auto *Br = cast<BranchInst>(Term);
-    assert(!Br->isConditional() &&
-           "BB's terminator must be an unconditional branch (or degenerate)");
-    BasicBlock *Succ = Br->getSuccessor(0);
-    Succ->removePredecessor(Source, /*KeepOneInputPHIs=*/true);
-    Br->setSuccessor(0, Target);
-    return;
-  }
-
-  auto *NewBr = BranchInst::Create(Target, Source);
-  NewBr->setDebugLoc(DL);
-}
-
 /// Redirect all edges that branch to \p OldTarget to \p NewTarget. That is,
 /// after this \p OldTarget will be orphaned.
 static void redirectAllPredecessorsTo(BasicBlock *OldTarget,
                                       BasicBlock *NewTarget, DebugLoc DL) {
   for (BasicBlock *Pred : make_early_inc_range(predecessors(OldTarget)))
     redirectTo(Pred, NewTarget, DL);
 }
 
void CanonicalLoopInfo::collectControlBlocks(
   // We only count those BBs as control block for which we do not need to
   // reverse the CFG, i.e. not the loop body which can contain arbitrary control
   // flow. For consistency, this also means we do not add the Body block, which
   // is just the entry to the body code.
   BBs.reserve(BBs.size() + 6);
   BBs.append({getPreheader(), Header, Cond, Latch, Exit, getAfter()});
 }
 
+void CanonicalLoopInfo::setTripCount(Value *TripCount) {
+  assert(isValid() && "Requires a valid canonical loop");
+
+  Instruction *CmpI = &getCond()->front();
+  assert(isa<CmpInst>(CmpI) && "First inst must compare IV with TripCount");
+  CmpI->setOperand(1, TripCount);
+
+#ifndef NDEBUG
+  assertOK();
+#endif
+}
+
+#if 0
+void CanonicalLoopInfo::replaceIndVarByCounter(Value *Counter){
+    assert(isValid() && "Requires a valid canonical loop");
+
+    Instruction* OldIV = getIndVar();
+    OldIV->replaceUsesWithIf(Counter, [this,Counter](Use &U) {
+        auto *Instr = dyn_cast<Instruction>(U.getUser());
+        if (!Instr)
+            return false;
+        BasicBlock* UserBB = Instr->getParent();
+       return  (UserBB != getCond()) && (UserBB != getLatch()) && (Instr != Counter);
+        });
+
+#ifndef NDEBUG
+    assertOK();
+#endif
+}
+#endif
+
+void CanonicalLoopInfo::mapIndVar(
+    llvm::function_ref<Value *(Instruction *)> Updater) {
+  assert(isValid() && "Requires a valid canonical loop");
+
+  Instruction *OldIV = getIndVar();
+
+  // Record all uses excluding those introduced by the updater. Uses by the
+  // CanonicalLoopInfo itself to keep track of the number of iterations are
+  // excluded.
+  SmallVector<Use *> ReplacableUses;
+  for (Use &U : OldIV->uses()) {
+    auto *User = dyn_cast<Instruction>(U.getUser());
+    if (!User)
+      continue;
+    if (User->getParent() == getCond())
+      continue;
+    if (User->getParent() == getLatch())
+      continue;
+    ReplacableUses.push_back(&U);
+  }
+
+  // Run the updater that may introduce new uses
+  Value *NewIV = Updater(OldIV);
+
+  // Replace the old uses with the value returned by the updater.
+  for (Use *U : ReplacableUses)
+    U->set(NewIV);
+
+#ifndef NDEBUG
+  assertOK();
+#endif
+}
+
 BasicBlock *CanonicalLoopInfo::getPreheader() const {
   assert(isValid() && "Requires a valid canonical loop");
   for (BasicBlock *Pred : predecessors(Header)) {
     if (Pred != Latch)
       return Pred;
   }
   llvm_unreachable("Missing preheader");
 }

llvm/unittests/Frontend/OpenMPIRBuilderTest.cpp

verifyListOrder(Function *F, ArrayRef<BasicBlock *> RefOrder) {
   if (!Prev)
     return testing::AssertionFailure() << "Did not find " << (*It)->getName()
                                        << " in function " << F->getName();
   return testing::AssertionFailure()
          << "Expected " << Prev->getName() << " before " << (*It)->getName()
          << " in function " << F->getName();
 }
 
+/// Populate Calls with call instructions calling the function with the given
+/// FnID from the given function F.
+static void findCalls(Function *F, omp::RuntimeFunction FnID,
+                      OpenMPIRBuilder &OMPBuilder,
+                      SmallVectorImpl<CallInst *> &Calls) {
+  Function *Fn = OMPBuilder.getOrCreateRuntimeFunctionPtr(FnID);
+  for (BasicBlock &BB : *F) {
+    for (Instruction &I : BB) {
+      auto *Call = dyn_cast<CallInst>(&I);
+      if (Call && Call->getCalledFunction() == Fn)
+        Calls.push_back(Call);
+    }
+  }
+}
+
+/// Assuming \p F contains only one call to the function with the given \p FnID,
+/// return that call.
+static CallInst *findSingleCall(Function *F, omp::RuntimeFunction FnID,
+                                OpenMPIRBuilder &OMPBuilder) {
+  SmallVector<CallInst *, 1> Calls;
+  findCalls(F, FnID, OMPBuilder, Calls);
+  EXPECT_EQ(1u, Calls.size());
+  if (Calls.size() != 1)
+    return nullptr;
+  return Calls.front();
+}
+
 class OpenMPIRBuilderTest : public testing::Test {
 protected:
   void SetUp() override {
     M.reset(new Module("MyModule", Ctx));
     FunctionType *FTy =
         FunctionType::get(Type::getVoidTy(Ctx), {Type::getInt32Ty(Ctx)},
                           /*isVarArg=*/false);
     F = Function::Create(FTy, Function::ExternalLinkage, "", M.get());
TEST_F(OpenMPIRBuilderTest, StaticWorkShareLoop) {
 
   // The exit block should contain the "fini" call and the barrier call,
   // plus the call to obtain the thread ID.
   size_t NumCallsInExitBlock =
       count_if(*ExitBlock, [](Instruction &I) { return isa<CallInst>(I); });
   EXPECT_EQ(NumCallsInExitBlock, 3u);
 }
 
+TEST_F(OpenMPIRBuilderTest, StaticChunkedWorkshareLoop) {
+  using InsertPointTy = OpenMPIRBuilder::InsertPointTy;
+  OpenMPIRBuilder OMPBuilder(*M);
+
+  BasicBlock *Body;
+  CanonicalLoopInfo *CLI =
+      buildSingleLoopFunction(DL, OMPBuilder, nullptr, &Body);
+
+  Type *LCTy = Type::getInt32Ty(Ctx);
+  Value *ChunkSize = ConstantInt::get(LCTy, 5);
+  InsertPointTy AllocaIP{&F->getEntryBlock(),
+                         F->getEntryBlock().getFirstInsertionPt()};
+  OMPBuilder.applyStaticChunkedWorkshareLoop(DL, CLI, AllocaIP,
+                                             /*NeedsBarrier=*/true, ChunkSize);
+
+  OMPBuilder.finalize();
+  EXPECT_FALSE(verifyModule(*M, &errs()));
+
+  BasicBlock *Entry = &F->getEntryBlock();
+  BasicBlock *Preheader = Entry->getSingleSuccessor();
+
+  BasicBlock *DispatchPreheader = Preheader->getSingleSuccessor();
+  BasicBlock *DispatchHeader = DispatchPreheader->getSingleSuccessor();
+  BasicBlock *DispatchCond = DispatchHeader->getSingleSuccessor();
+  BasicBlock *DispatchBody = succ_begin(DispatchCond)[0];
+  BasicBlock *DispatchExit = succ_begin(DispatchCond)[1];
+  BasicBlock *DispatchAfter = DispatchExit->getSingleSuccessor();
+  BasicBlock *Return = DispatchAfter->getSingleSuccessor();
+
+  BasicBlock *ChunkPreheader = DispatchBody->getSingleSuccessor();
+  BasicBlock *ChunkHeader = ChunkPreheader->getSingleSuccessor();
+  BasicBlock *ChunkCond = ChunkHeader->getSingleSuccessor();
+  BasicBlock *ChunkBody = succ_begin(ChunkCond)[0];
+  BasicBlock *ChunkExit = succ_begin(ChunkCond)[1];
+  BasicBlock *ChunkInc = ChunkBody->getSingleSuccessor();
+  BasicBlock *ChunkAfter = ChunkExit->getSingleSuccessor();
+
+  BasicBlock *DispatchInc = ChunkAfter;
+
+  EXPECT_EQ(ChunkBody, Body);
+  EXPECT_EQ(ChunkInc->getSingleSuccessor(), ChunkHeader);
+  EXPECT_EQ(DispatchInc->getSingleSuccessor(), DispatchHeader);
+
+  EXPECT_TRUE(isa<ReturnInst>(Return->front()));
+
+  CallInst *InitCall = findSingleCall(
+      F, omp::RuntimeFunction::OMPRTL___kmpc_for_static_init_4u, OMPBuilder);
+  EXPECT_EQ(InitCall->getParent(), Preheader);
+  EXPECT_EQ(cast<ConstantInt>(InitCall->getArgOperand(2))->getSExtValue(), 33);
+  EXPECT_EQ(cast<ConstantInt>(InitCall->getArgOperand(7))->getSExtValue(), 1);
+  EXPECT_EQ(cast<ConstantInt>(InitCall->getArgOperand(8))->getSExtValue(), 5);
+
+  CallInst *FiniCall = findSingleCall(
+      F, omp::RuntimeFunction::OMPRTL___kmpc_for_static_fini, OMPBuilder);
+  EXPECT_EQ(FiniCall->getParent(), DispatchExit);
+
+  CallInst *BarrierCall = findSingleCall(
+      F, omp::RuntimeFunction::OMPRTL___kmpc_barrier, OMPBuilder);
+  EXPECT_EQ(BarrierCall->getParent(), DispatchExit);
+}
+
 TEST_P(OpenMPIRBuilderTestWithParams, DynamicWorkShareLoop) {
   using InsertPointTy = OpenMPIRBuilder::InsertPointTy;
   OpenMPIRBuilder OMPBuilder(*M);
   OMPBuilder.initialize();
   IRBuilder<> Builder(BB);
   OpenMPIRBuilder::LocationDescription Loc({Builder.saveIP(), DL});
 
   omp::OMPScheduleType SchedType = GetParam();
xorAtomicReduction(OpenMPIRBuilder::InsertPointTy IP, Value *LHS, Value *RHS) {
   IRBuilder<> Builder(IP.getBlock(), IP.getPoint());
   Value *Partial = Builder.CreateLoad(RHS->getType()->getPointerElementType(),
                                       RHS, "red.partial");
   Builder.CreateAtomicRMW(AtomicRMWInst::Xor, LHS, Partial, None,
                           AtomicOrdering::Monotonic);
   return Builder.saveIP();
 }
 
-/// Populate Calls with call instructions calling the function with the given
-/// FnID from the given function F.
-static void findCalls(Function *F, omp::RuntimeFunction FnID,
-                      OpenMPIRBuilder &OMPBuilder,
-                      SmallVectorImpl<CallInst *> &Calls) {
-  Function *Fn = OMPBuilder.getOrCreateRuntimeFunctionPtr(FnID);
-  for (BasicBlock &BB : *F) {
-    for (Instruction &I : BB) {
-      auto *Call = dyn_cast<CallInst>(&I);
-      if (Call && Call->getCalledFunction() == Fn)
-        Calls.push_back(Call);
-    }
-  }
-}
-
 TEST_F(OpenMPIRBuilderTest, CreateReductions) {
   using InsertPointTy = OpenMPIRBuilder::InsertPointTy;
   OpenMPIRBuilder OMPBuilder(*M);
   OMPBuilder.initialize();
   F->setName("func");
   IRBuilder<> Builder(BB);
   OpenMPIRBuilder::LocationDescription Loc({Builder.saveIP(), DL});
 

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

convertOmpWsLoop(Operation &opInst, llvm::IRBuilderBase &builder,
   // Collapse loops. Store the insertion point because LoopInfos may get
   // invalidated.
   llvm::IRBuilderBase::InsertPoint afterIP = loopInfos.front()->getAfterIP();
   llvm::CanonicalLoopInfo *loopInfo =
       ompBuilder->collapseLoops(diLoc, loopInfos, {});
 
   allocaIP = findAllocaInsertPoint(builder, moduleTranslation);
   if (schedule == omp::ClauseScheduleKind::Static) {
-    ompBuilder->applyStaticWorkshareLoop(ompLoc.DL, loopInfo, allocaIP,
-                                         !loop.nowait(), chunk);
+    ompBuilder->applyWorkshareLoop(ompLoc.DL, loopInfo, allocaIP,
+                                   !loop.nowait(),
+                                   llvm::omp::OMP_SCHEDULE_Static, chunk);
   } else {
     llvm::omp::OMPScheduleType schedType;
     switch (schedule) {
     case omp::ClauseScheduleKind::Dynamic:
       schedType = llvm::omp::OMPScheduleType::DynamicChunked;
       break;
     case omp::ClauseScheduleKind::Guided:
       schedType = llvm::omp::OMPScheduleType::GuidedChunked;