Revert "[Remarks] Emit optimization remarks for atomics generating CAS loop"

This reverts commit 435785214f73ff0c92e97f2ade6356e3ba3bf661. Still same compile time issues for -O0 -g, eg. +1.3% for sqlite3.

diff --git a/clang/test/CodeGenCUDA/atomics-remarks-gfx90a.cu b/clang/test/CodeGenCUDA/atomics-remarks-gfx90a.cu
deleted file mode 100644 (file)
index 9689228..0000000
--- a/clang/test/CodeGenCUDA/atomics-remarks-gfx90a.cu
+++ /dev/null
@@ -1,16 +0,0 @@
-// RUN: %clang_cc1 %s -triple=amdgcn-amd-amdhsa -fcuda-is-device \
-// RUN:   -target-cpu gfx90a -Rpass=atomic-expand -S -o - 2>&1 | \
-// RUN:   FileCheck %s --check-prefix=GFX90A-CAS
-
-// REQUIRES: amdgpu-registered-target
-
-#include "Inputs/cuda.h"
-#include <stdatomic.h>
-
-// GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at system memory scope
-// GFX90A-CAS-LABEL: _Z14atomic_add_casPf
-// GFX90A-CAS:  flat_atomic_cmpswap v0, v[2:3], v[4:5] glc
-// GFX90A-CAS:  s_cbranch_execnz
-__device__ float atomic_add_cas(float *p) {
-  return __atomic_fetch_add(p, 1.0f, memory_order_relaxed);
-}

diff --git a/clang/test/CodeGenOpenCL/atomics-remarks-gfx90a.cl b/clang/test/CodeGenOpenCL/atomics-remarks-gfx90a.cl
deleted file mode 100644 (file)
index 2d8b68f..0000000
--- a/clang/test/CodeGenOpenCL/atomics-remarks-gfx90a.cl
+++ /dev/null
@@ -1,46 +0,0 @@
-// RUN: %clang_cc1 %s -cl-std=CL2.0 -O0 -triple=amdgcn-amd-amdhsa -target-cpu gfx90a \
-// RUN:     -Rpass=atomic-expand -S -o - 2>&1 | \
-// RUN:     FileCheck %s --check-prefix=REMARK
-
-// RUN: %clang_cc1 %s -cl-std=CL2.0 -O0 -triple=amdgcn-amd-amdhsa -target-cpu gfx90a \
-// RUN:     -Rpass=atomic-expand -S -emit-llvm -o - 2>&1 | \
-// RUN:     FileCheck %s --check-prefix=GFX90A-CAS
-
-// REQUIRES: amdgpu-registered-target
-
-typedef enum memory_order {
-  memory_order_relaxed = __ATOMIC_RELAXED,
-  memory_order_acquire = __ATOMIC_ACQUIRE,
-  memory_order_release = __ATOMIC_RELEASE,
-  memory_order_acq_rel = __ATOMIC_ACQ_REL,
-  memory_order_seq_cst = __ATOMIC_SEQ_CST
-} memory_order;
-
-typedef enum memory_scope {
-  memory_scope_work_item = __OPENCL_MEMORY_SCOPE_WORK_ITEM,
-  memory_scope_work_group = __OPENCL_MEMORY_SCOPE_WORK_GROUP,
-  memory_scope_device = __OPENCL_MEMORY_SCOPE_DEVICE,
-  memory_scope_all_svm_devices = __OPENCL_MEMORY_SCOPE_ALL_SVM_DEVICES,
-#if defined(cl_intel_subgroups) || defined(cl_khr_subgroups)
-  memory_scope_sub_group = __OPENCL_MEMORY_SCOPE_SUB_GROUP
-#endif
-} memory_scope;
-
-// REMARK: remark: A compare and swap loop was generated for an atomic fadd operation at workgroup-one-as memory scope [-Rpass=atomic-expand]
-// REMARK: remark: A compare and swap loop was generated for an atomic fadd operation at agent-one-as memory scope [-Rpass=atomic-expand]
-// REMARK: remark: A compare and swap loop was generated for an atomic fadd operation at one-as memory scope [-Rpass=atomic-expand]
-// REMARK: remark: A compare and swap loop was generated for an atomic fadd operation at wavefront-one-as memory scope [-Rpass=atomic-expand]
-// GFX90A-CAS-LABEL: @atomic_cas
-// GFX90A-CAS: atomicrmw fadd float addrspace(1)* {{.*}} syncscope("workgroup-one-as") monotonic
-// GFX90A-CAS: atomicrmw fadd float addrspace(1)* {{.*}} syncscope("agent-one-as") monotonic
-// GFX90A-CAS: atomicrmw fadd float addrspace(1)* {{.*}} syncscope("one-as") monotonic
-// GFX90A-CAS: atomicrmw fadd float addrspace(1)* {{.*}} syncscope("wavefront-one-as") monotonic
-float atomic_cas(__global atomic_float *d, float a) {
-  float ret1 = __opencl_atomic_fetch_add(d, a, memory_order_relaxed, memory_scope_work_group);
-  float ret2 = __opencl_atomic_fetch_add(d, a, memory_order_relaxed, memory_scope_device);
-  float ret3 = __opencl_atomic_fetch_add(d, a, memory_order_relaxed, memory_scope_all_svm_devices);
-  float ret4 = __opencl_atomic_fetch_add(d, a, memory_order_relaxed, memory_scope_sub_group);
-}
-
-
-

diff --git a/llvm/lib/CodeGen/AtomicExpandPass.cpp b/llvm/lib/CodeGen/AtomicExpandPass.cpp
index 5b5458e..125a3be 100644 (file)
--- a/llvm/lib/CodeGen/AtomicExpandPass.cpp
+++ b/llvm/lib/CodeGen/AtomicExpandPass.cpp
@@ -17,7 +17,6 @@
 #include "llvm/ADT/ArrayRef.h"
 #include "llvm/ADT/STLExtras.h"
 #include "llvm/ADT/SmallVector.h"
-#include "llvm/Analysis/OptimizationRemarkEmitter.h"
 #include "llvm/CodeGen/AtomicExpandUtils.h"
 #include "llvm/CodeGen/RuntimeLibcalls.h"
 #include "llvm/CodeGen/TargetLowering.h"
@@ -59,7 +58,6 @@ namespace {
 
   class AtomicExpand: public FunctionPass {
     const TargetLowering *TLI = nullptr;
-    OptimizationRemarkEmitter *ORE;
 
   public:
     static char ID; // Pass identification, replacement for typeid
@@ -71,7 +69,6 @@ namespace {
     bool runOnFunction(Function &F) override;
 
   private:
-    void getAnalysisUsage(AnalysisUsage &AU) const override;
     bool bracketInstWithFences(Instruction *I, AtomicOrdering Order);
     IntegerType *getCorrespondingIntegerType(Type *T, const DataLayout &DL);
     LoadInst *convertAtomicLoadToIntegerType(LoadInst *LI);
@@ -168,16 +165,11 @@ static bool atomicSizeSupported(const TargetLowering *TLI, Inst *I) {
          Size <= TLI->getMaxAtomicSizeInBitsSupported() / 8;
 }
 
-void AtomicExpand::getAnalysisUsage(AnalysisUsage &AU) const {
-  AU.addRequired<OptimizationRemarkEmitterWrapperPass>();
-}
-
 bool AtomicExpand::runOnFunction(Function &F) {
   auto *TPC = getAnalysisIfAvailable<TargetPassConfig>();
   if (!TPC)
     return false;
 
-  ORE = &getAnalysis<OptimizationRemarkEmitterWrapperPass>().getORE();
   auto &TM = TPC->getTM<TargetMachine>();
   if (!TM.getSubtargetImpl(F)->enableAtomicExpand())
     return false;
@@ -578,9 +570,7 @@ static Value *performAtomicOp(AtomicRMWInst::BinOp Op, IRBuilder<> &Builder,
 }
 
 bool AtomicExpand::tryExpandAtomicRMW(AtomicRMWInst *AI) {
-  LLVMContext &Ctx = AI->getModule()->getContext();
-  TargetLowering::AtomicExpansionKind Kind = TLI->shouldExpandAtomicRMWInIR(AI);
-  switch (Kind) {
+  switch (TLI->shouldExpandAtomicRMWInIR(AI)) {
   case TargetLoweringBase::AtomicExpansionKind::None:
     return false;
   case TargetLoweringBase::AtomicExpansionKind::LLSC: {
@@ -610,17 +600,6 @@ bool AtomicExpand::tryExpandAtomicRMW(AtomicRMWInst *AI) {
       expandPartwordAtomicRMW(AI,
                               TargetLoweringBase::AtomicExpansionKind::CmpXChg);
     } else {
-      SmallVector<StringRef> SSNs;
-      Ctx.getSyncScopeNames(SSNs);
-      auto MemScope = SSNs[AI->getSyncScopeID()].empty()
-                          ? "system"
-                          : SSNs[AI->getSyncScopeID()];
-      ORE->emit([&]() {
-        return OptimizationRemark(DEBUG_TYPE, "Passed", AI->getFunction())
-               << "A compare and swap loop was generated for an atomic "
-               << AI->getOperationName(AI->getOperation()) << " operation at "
-               << MemScope << " memory scope";
-      });
       expandAtomicRMWToCmpXchg(AI, createCmpXchgInstFun);
     }
     return true;

diff --git a/llvm/test/CodeGen/AArch64/O0-pipeline.ll b/llvm/test/CodeGen/AArch64/O0-pipeline.ll
index c724389..c382ad0 100644 (file)
--- a/llvm/test/CodeGen/AArch64/O0-pipeline.ll
+++ b/llvm/test/CodeGen/AArch64/O0-pipeline.ll
@@ -8,18 +8,13 @@
 ; CHECK-NEXT: Target Pass Configuration
 ; CHECK-NEXT: Machine Module Information
 ; CHECK-NEXT: Target Transform Information
-; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Create Garbage Collector Module Metadata
 ; CHECK-NEXT: Assumption Cache Tracker
+; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Machine Branch Probability Analysis
 ; CHECK-NEXT:   ModulePass Manager
 ; CHECK-NEXT:     Pre-ISel Intrinsic Lowering
 ; CHECK-NEXT:     FunctionPass Manager
-; CHECK-NEXT:       Dominator Tree Construction
-; CHECK-NEXT:       Natural Loop Information
-; CHECK-NEXT:       Lazy Branch Probability Analysis
-; CHECK-NEXT:       Lazy Block Frequency Analysis
-; CHECK-NEXT:       Optimization Remark Emitter
 ; CHECK-NEXT:       Expand Atomic instructions
 ; CHECK-NEXT:       Module Verifier
 ; CHECK-NEXT:       Lower Garbage Collection Instructions

diff --git a/llvm/test/CodeGen/AArch64/O3-pipeline.ll b/llvm/test/CodeGen/AArch64/O3-pipeline.ll
index 50ec6f6..8d03f7d 100644 (file)
--- a/llvm/test/CodeGen/AArch64/O3-pipeline.ll
+++ b/llvm/test/CodeGen/AArch64/O3-pipeline.ll
@@ -8,8 +8,8 @@
 ; CHECK-NEXT: Target Pass Configuration
 ; CHECK-NEXT: Machine Module Information
 ; CHECK-NEXT: Target Transform Information
-; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Assumption Cache Tracker
+; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Type-Based Alias Analysis
 ; CHECK-NEXT: Scoped NoAlias Alias Analysis
 ; CHECK-NEXT: Create Garbage Collector Module Metadata
@@ -17,11 +17,6 @@
 ; CHECK-NEXT:   ModulePass Manager
 ; CHECK-NEXT:     Pre-ISel Intrinsic Lowering
 ; CHECK-NEXT:     FunctionPass Manager
-; CHECK-NEXT:       Dominator Tree Construction
-; CHECK-NEXT:       Natural Loop Information
-; CHECK-NEXT:       Lazy Branch Probability Analysis
-; CHECK-NEXT:       Lazy Block Frequency Analysis
-; CHECK-NEXT:       Optimization Remark Emitter
 ; CHECK-NEXT:       Expand Atomic instructions
 ; CHECK-NEXT:     SVE intrinsics optimizations
 ; CHECK-NEXT:       FunctionPass Manager

diff --git a/llvm/test/CodeGen/AMDGPU/atomics-remarks-gfx90a.ll b/llvm/test/CodeGen/AMDGPU/atomics-remarks-gfx90a.ll
deleted file mode 100644 (file)
index 240963c..0000000
--- a/llvm/test/CodeGen/AMDGPU/atomics-remarks-gfx90a.ll
+++ /dev/null
@@ -1,103 +0,0 @@
-; RUN: llc -march=amdgcn -mcpu=gfx90a -verify-machineinstrs --pass-remarks=atomic-expand \
-; RUN:      %s -o - 2>&1 | FileCheck %s --check-prefix=GFX90A-CAS
-
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at system memory scope
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at agent memory scope
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at workgroup memory scope
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at wavefront memory scope
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at singlethread memory scope
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at one-as memory scope
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at agent-one-as memory scope
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at workgroup-one-as memory scope
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at wavefront-one-as memory scope
-; GFX90A-CAS: A compare and swap loop was generated for an atomic fadd operation at singlethread-one-as memory scope
-
-; GFX90A-CAS-LABEL: atomic_add_cas:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q monotonic, align 4
-  ret void
-}
-
-; GFX90A-CAS-LABEL: atomic_add_cas_agent:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas_agent(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q syncscope("agent") monotonic, align 4
-  ret void
-}
-
-; GFX90A-CAS-LABEL: atomic_add_cas_workgroup:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas_workgroup(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q syncscope("workgroup") monotonic, align 4
-  ret void
-}
-
-; GFX90A-CAS-LABEL: atomic_add_cas_wavefront:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas_wavefront(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q syncscope("wavefront") monotonic, align 4
-  ret void
-}
-
-; GFX90A-CAS-LABEL: atomic_add_cas_singlethread:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas_singlethread(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q syncscope("singlethread") monotonic, align 4
-  ret void
-}
-
-; GFX90A-CAS-LABEL: atomic_add_cas_one_as:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas_one_as(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q syncscope("one-as") monotonic, align 4
-  ret void
-}
-
-; GFX90A-CAS-LABEL: atomic_add_cas_agent_one_as:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas_agent_one_as(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q syncscope("agent-one-as") monotonic, align 4
-  ret void
-}
-
-; GFX90A-CAS-LABEL: atomic_add_cas_workgroup_one_as:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas_workgroup_one_as(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q syncscope("workgroup-one-as") monotonic, align 4
-  ret void
-}
-
-; GFX90A-CAS-LABEL: atomic_add_cas_wavefront_one_as:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas_wavefront_one_as(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q syncscope("wavefront-one-as") monotonic, align 4
-  ret void
-}
-
-; GFX90A-CAS-LABEL: atomic_add_cas_singlethread_one_as:
-; GFX90A-CAS: flat_atomic_cmpswap v3, v[0:1], v[4:5] glc
-; GFX90A-CAS: s_cbranch_execnz
-define dso_local void @atomic_add_cas_singlethread_one_as(float* %p, float %q) {
-entry:
-  %ret = atomicrmw fadd float* %p, float %q syncscope("singlethread-one-as") monotonic, align 4
-  ret void
-}

diff --git a/llvm/test/CodeGen/AMDGPU/llc-pipeline.ll b/llvm/test/CodeGen/AMDGPU/llc-pipeline.ll
index dba871e..73909dc 100644 (file)
--- a/llvm/test/CodeGen/AMDGPU/llc-pipeline.ll
+++ b/llvm/test/CodeGen/AMDGPU/llc-pipeline.ll
@@ -44,11 +44,6 @@
 ; GCN-O0-NEXT:    Lower OpenCL enqueued blocks
 ; GCN-O0-NEXT:    Lower uses of LDS variables from non-kernel functions
 ; GCN-O0-NEXT:    FunctionPass Manager
-; GCN-O0-NEXT:      Dominator Tree Construction
-; GCN-O0-NEXT:      Natural Loop Information
-; GCN-O0-NEXT:      Lazy Branch Probability Analysis
-; GCN-O0-NEXT:      Lazy Block Frequency Analysis
-; GCN-O0-NEXT:      Optimization Remark Emitter
 ; GCN-O0-NEXT:      Expand Atomic instructions
 ; GCN-O0-NEXT:      Lower constant intrinsics
 ; GCN-O0-NEXT:      Remove unreachable blocks from the CFG
@@ -185,11 +180,6 @@
 ; GCN-O1-NEXT:    Lower uses of LDS variables from non-kernel functions
 ; GCN-O1-NEXT:    FunctionPass Manager
 ; GCN-O1-NEXT:      Infer address spaces
-; GCN-O1-NEXT:      Dominator Tree Construction
-; GCN-O1-NEXT:      Natural Loop Information
-; GCN-O1-NEXT:      Lazy Branch Probability Analysis
-; GCN-O1-NEXT:      Lazy Block Frequency Analysis
-; GCN-O1-NEXT:      Optimization Remark Emitter
 ; GCN-O1-NEXT:      Expand Atomic instructions
 ; GCN-O1-NEXT:      AMDGPU Promote Alloca
 ; GCN-O1-NEXT:      Dominator Tree Construction
@@ -441,11 +431,6 @@
 ; GCN-O1-OPTS-NEXT:    Lower uses of LDS variables from non-kernel functions
 ; GCN-O1-OPTS-NEXT:    FunctionPass Manager
 ; GCN-O1-OPTS-NEXT:      Infer address spaces
-; GCN-O1-OPTS-NEXT:      Dominator Tree Construction
-; GCN-O1-OPTS-NEXT:      Natural Loop Information
-; GCN-O1-OPTS-NEXT:      Lazy Branch Probability Analysis
-; GCN-O1-OPTS-NEXT:      Lazy Block Frequency Analysis
-; GCN-O1-OPTS-NEXT:      Optimization Remark Emitter
 ; GCN-O1-OPTS-NEXT:      Expand Atomic instructions
 ; GCN-O1-OPTS-NEXT:      AMDGPU Promote Alloca
 ; GCN-O1-OPTS-NEXT:      Dominator Tree Construction
@@ -730,11 +715,6 @@
 ; GCN-O2-NEXT:    Lower uses of LDS variables from non-kernel functions
 ; GCN-O2-NEXT:    FunctionPass Manager
 ; GCN-O2-NEXT:      Infer address spaces
-; GCN-O2-NEXT:      Dominator Tree Construction
-; GCN-O2-NEXT:      Natural Loop Information
-; GCN-O2-NEXT:      Lazy Branch Probability Analysis
-; GCN-O2-NEXT:      Lazy Block Frequency Analysis
-; GCN-O2-NEXT:      Optimization Remark Emitter
 ; GCN-O2-NEXT:      Expand Atomic instructions
 ; GCN-O2-NEXT:      AMDGPU Promote Alloca
 ; GCN-O2-NEXT:      Dominator Tree Construction
@@ -1021,11 +1001,6 @@
 ; GCN-O3-NEXT:    Lower uses of LDS variables from non-kernel functions
 ; GCN-O3-NEXT:    FunctionPass Manager
 ; GCN-O3-NEXT:      Infer address spaces
-; GCN-O3-NEXT:      Dominator Tree Construction
-; GCN-O3-NEXT:      Natural Loop Information
-; GCN-O3-NEXT:      Lazy Branch Probability Analysis
-; GCN-O3-NEXT:      Lazy Block Frequency Analysis
-; GCN-O3-NEXT:      Optimization Remark Emitter
 ; GCN-O3-NEXT:      Expand Atomic instructions
 ; GCN-O3-NEXT:      AMDGPU Promote Alloca
 ; GCN-O3-NEXT:      Dominator Tree Construction

diff --git a/llvm/test/CodeGen/ARM/O3-pipeline.ll b/llvm/test/CodeGen/ARM/O3-pipeline.ll
index e25f7b3..2a5ba76 100644 (file)
--- a/llvm/test/CodeGen/ARM/O3-pipeline.ll
+++ b/llvm/test/CodeGen/ARM/O3-pipeline.ll
@@ -5,11 +5,6 @@
 ; CHECK:       ModulePass Manager
 ; CHECK-NEXT:    Pre-ISel Intrinsic Lowering
 ; CHECK-NEXT:    FunctionPass Manager
-; CHECK-NEXT:       Dominator Tree Construction
-; CHECK-NEXT:       Natural Loop Information
-; CHECK-NEXT:       Lazy Branch Probability Analysis
-; CHECK-NEXT:       Lazy Block Frequency Analysis
-; CHECK-NEXT:       Optimization Remark Emitter
 ; CHECK-NEXT:      Expand Atomic instructions
 ; CHECK-NEXT:      Simplify the CFG
 ; CHECK-NEXT:      Dominator Tree Construction

diff --git a/llvm/test/CodeGen/PowerPC/O3-pipeline.ll b/llvm/test/CodeGen/PowerPC/O3-pipeline.ll
index c37646c..aee62db 100644 (file)
--- a/llvm/test/CodeGen/PowerPC/O3-pipeline.ll
+++ b/llvm/test/CodeGen/PowerPC/O3-pipeline.ll
@@ -8,21 +8,16 @@
 ; CHECK-NEXT: Target Pass Configuration
 ; CHECK-NEXT: Machine Module Information
 ; CHECK-NEXT: Target Transform Information
-; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Assumption Cache Tracker
 ; CHECK-NEXT: Type-Based Alias Analysis
 ; CHECK-NEXT: Scoped NoAlias Alias Analysis
+; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Create Garbage Collector Module Metadata
 ; CHECK-NEXT: Machine Branch Probability Analysis
 ; CHECK-NEXT:   ModulePass Manager
 ; CHECK-NEXT:     Pre-ISel Intrinsic Lowering
 ; CHECK-NEXT:     FunctionPass Manager
 ; CHECK-NEXT:       Convert i1 constants to i32/i64 if they are returned
-; CHECK-NEXT:       Dominator Tree Construction
-; CHECK-NEXT:       Natural Loop Information
-; CHECK-NEXT:       Lazy Branch Probability Analysis
-; CHECK-NEXT:       Lazy Block Frequency Analysis
-; CHECK-NEXT:       Optimization Remark Emitter
 ; CHECK-NEXT:       Expand Atomic instructions
 ; CHECK-NEXT:     PPC Lower MASS Entries
 ; CHECK-NEXT:     FunctionPass Manager
@@ -211,5 +206,4 @@
 
 define void @f() {
   ret void
-}
-
+}
\ No newline at end of file

diff --git a/llvm/test/CodeGen/X86/O0-pipeline.ll b/llvm/test/CodeGen/X86/O0-pipeline.ll
index 8f02757..bf3ae61 100644 (file)
--- a/llvm/test/CodeGen/X86/O0-pipeline.ll
+++ b/llvm/test/CodeGen/X86/O0-pipeline.ll
@@ -10,18 +10,13 @@
 ; CHECK-NEXT: Target Pass Configuration
 ; CHECK-NEXT: Machine Module Information
 ; CHECK-NEXT: Target Transform Information
-; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Create Garbage Collector Module Metadata
 ; CHECK-NEXT: Assumption Cache Tracker
+; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Machine Branch Probability Analysis
 ; CHECK-NEXT:   ModulePass Manager
 ; CHECK-NEXT:     Pre-ISel Intrinsic Lowering
 ; CHECK-NEXT:     FunctionPass Manager
-; CHECK-NEXT:       Dominator Tree Construction
-; CHECK-NEXT:       Natural Loop Information
-; CHECK-NEXT:       Lazy Branch Probability Analysis
-; CHECK-NEXT:       Lazy Block Frequency Analysis
-; CHECK-NEXT:       Optimization Remark Emitter
 ; CHECK-NEXT:       Expand Atomic instructions
 ; CHECK-NEXT:       Lower AMX intrinsics
 ; CHECK-NEXT:       Lower AMX type for load/store

diff --git a/llvm/test/CodeGen/X86/opt-pipeline.ll b/llvm/test/CodeGen/X86/opt-pipeline.ll
index a480d90..c809433 100644 (file)
--- a/llvm/test/CodeGen/X86/opt-pipeline.ll
+++ b/llvm/test/CodeGen/X86/opt-pipeline.ll
@@ -16,20 +16,15 @@
 ; CHECK-NEXT: Target Pass Configuration
 ; CHECK-NEXT: Machine Module Information
 ; CHECK-NEXT: Target Transform Information
-; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Type-Based Alias Analysis
 ; CHECK-NEXT: Scoped NoAlias Alias Analysis
 ; CHECK-NEXT: Assumption Cache Tracker
+; CHECK-NEXT: Profile summary info
 ; CHECK-NEXT: Create Garbage Collector Module Metadata
 ; CHECK-NEXT: Machine Branch Probability Analysis
 ; CHECK-NEXT:   ModulePass Manager
 ; CHECK-NEXT:     Pre-ISel Intrinsic Lowering
 ; CHECK-NEXT:     FunctionPass Manager
-; CHECK-NEXT:       Dominator Tree Construction
-; CHECK-NEXT:       Natural Loop Information
-; CHECK-NEXT:       Lazy Branch Probability Analysis
-; CHECK-NEXT:       Lazy Block Frequency Analysis
-; CHECK-NEXT:       Optimization Remark Emitter
 ; CHECK-NEXT:       Expand Atomic instructions
 ; CHECK-NEXT:       Lower AMX intrinsics
 ; CHECK-NEXT:       Lower AMX type for load/store