[SYCL-MLIR] Run buildO0DefaultPipeline for OptimizationLevel::O0 (#7452)

whitneywhtsang · web-flow · commit 29a4a892bd97 · 2022-11-19T19:04:22.000-05:00
The result of lit are changed because functions with `alwaysinline`
attribute are now inlined in the `O0` default pipeline.

Signed-off-by: Tsang, Whitney &lt;whitney.tsang@intel.com&gt;
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/constructors.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/constructors.cpp
@@ -1,5 +1,5 @@
-// RUN: clang++ -fsycl -fsycl-device-only -emit-mlir %s -o - 2> /dev/null | FileCheck %s --check-prefixes=CHECK
-// RUN: clang++ -fsycl -fsycl-device-only -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefixes=CHECK-LLVM
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-mlir %s -o - 2> /dev/null | FileCheck %s --check-prefixes=CHECK
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefixes=CHECK-LLVM
 
 #include <sycl/sycl.hpp>
 
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/functions.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/functions.cpp
@@ -1,8 +1,8 @@
-// RUN: clang++ -fsycl -fsycl-device-only -emit-mlir %s -o - | FileCheck %s --check-prefix=CHECK-MLIR
-// RUN: clang++ -fsycl -fsycl-device-only -emit-mlir %s -o - -Xcgeist -no-mangled-function-name | FileCheck %s --check-prefix=CHECK-MLIR-NO-MANGLED-FUNCTION-NAME
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-mlir %s -o - | FileCheck %s --check-prefix=CHECK-MLIR
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-mlir %s -o - -Xcgeist -no-mangled-function-name | FileCheck %s --check-prefix=CHECK-MLIR-NO-MANGLED-FUNCTION-NAME
 // COM: These two should obtain the same results.
-// RUN: clang++ -fsycl -fsycl-device-only -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
-// RUN: clang++ -fsycl -fsycl-device-only -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - -Xcgeist -no-mangled-function-name | FileCheck %s --check-prefix=CHECK-LLVM
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - -Xcgeist -no-mangled-function-name | FileCheck %s --check-prefix=CHECK-LLVM
 
 #include <sycl/sycl.hpp>
 
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/kernels.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/kernels.cpp
@@ -1,5 +1,5 @@
-// RUN: clang++ -fsycl -fsycl-device-only -emit-mlir %s -o - 2> /dev/null| FileCheck %s --check-prefix=CHECK-MLIR
-// RUN: clang++ -fsycl -fsycl-device-only -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-mlir %s -o - 2> /dev/null| FileCheck %s --check-prefix=CHECK-MLIR
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
 
 #include <sycl/sycl.hpp>
 
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/kernels_funcs.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/kernels_funcs.cpp
@@ -1,5 +1,5 @@
-// RUN: clang++ -fsycl -fsycl-device-only -emit-mlir %s -o - 2> /dev/null | FileCheck %s --check-prefix=CHECK-MLIR
-// RUN: clang++ -fsycl -fsycl-device-only -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-mlir %s -o - 2> /dev/null | FileCheck %s --check-prefix=CHECK-MLIR
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
 
 #include <sycl/sycl.hpp>
 
@@ -17,12 +17,10 @@
 // CHECK-MLIR-SAME: attributes {llvm.cconv = #llvm.cconv<spir_funccc>, llvm.linkage = #llvm.linkage<linkonce_odr>, {{.*}}
 
 // COM: StoreWrapper constructor:
-// CHECK-LLVM-DAG:      define linkonce_odr spir_func void @_ZN12StoreWrapperIiLi1ELN4sycl3_V16access4modeE1026EEclEv({{.*}}) #[[FUNCATTRS2:[0-9]+]]
 // CHECK-LLVM-DAG:      define weak_odr spir_kernel void @_ZTS8kernel_1({{.*}}) #[[FUNCATTRS1:[0-9]+]]
 // CHECK-LLVM-DAG:      define weak_odr spir_kernel void @_ZTSZZ6host_2vENKUlRN4sycl3_V17handlerEE_clES2_E8kernel_2({{.*}}) #[[FUNCATTRS1]]
 
 // CHECK-LLVM-DAG: attributes #[[FUNCATTRS1]] = { convergent mustprogress norecurse nounwind "frame-pointer"="all" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="{{.*}}/polygeist/tools/cgeist/Test/Verification/sycl/kernels_funcs.cpp" }
-// CHECK-LLVM-DAG: attributes #[[FUNCATTRS2]] = { alwaysinline convergent mustprogress norecurse nounwind "frame-pointer"="all" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="{{.*}}/polygeist/tools/cgeist/Test/Verification/sycl/kernels_funcs.cpp" }
 
 template <typename DataT,
           int Dimensions = 1,
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/parallel_for.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/parallel_for.cpp
@@ -1,6 +1,6 @@
-// RUN: clang++ -fsycl -fsycl-targets=spir64-unknown-unknown-syclmlir -Xcgeist -gen-all-sycl-funcs %s -o %t.out 2>&1 | FileCheck %s --implicit-check-not="{{error|Error}}:"
+// RUN: clang++ -fsycl -fsycl-targets=spir64-unknown-unknown-syclmlir -O0 -Xcgeist -gen-all-sycl-funcs %s -o %t.out 2>&1 | FileCheck %s --implicit-check-not="{{error|Error}}:"
 
-// RUN: clang++ -fsycl -fsycl-device-only -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir -Xcgeist -gen-all-sycl-funcs %s -o %t.bc 2>/dev/null
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir -Xcgeist -gen-all-sycl-funcs %s -o %t.bc 2>/dev/null
 
 // Test that the LLVMIR generated is verifiable.
 // RUN: opt -verify -disable-output < %t.bc
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/single_task.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/single_task.cpp
@@ -1,6 +1,6 @@
-// RUN: clang++ -fsycl -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o %t.out 2>&1 | FileCheck %s --implicit-check-not="{{error|Error}}:"
+// RUN: clang++ -fsycl -fsycl-targets=spir64-unknown-unknown-syclmlir -O0 %s -o %t.out 2>&1 | FileCheck %s --implicit-check-not="{{error|Error}}:"
 
-// RUN: clang++ -fsycl -fsycl-device-only -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o %t.bc 2>/dev/null
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o %t.bc 2>/dev/null
 
 // Test that the LLVMIR generated is verifiable.
 // RUN: opt -verify -disable-output < %t.bc
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/stream-copy.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/stream-copy.cpp
@@ -1,6 +1,6 @@
-// RUN: clang++ -fsycl -fsycl-targets=spir64-unknown-unknown-syclmlir -Xcgeist -gen-all-sycl-funcs %s -o %t.out 2>&1 | FileCheck %s --implicit-check-not="{{error|Error}}:"
+// RUN: clang++ -fsycl -fsycl-targets=spir64-unknown-unknown-syclmlir -O0 -Xcgeist -gen-all-sycl-funcs %s -o %t.out 2>&1 | FileCheck %s --implicit-check-not="{{error|Error}}:"
 
-// RUN: clang++ -fsycl -fsycl-device-only -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir -Xcgeist -gen-all-sycl-funcs %s -o %t.bc 2>/dev/null
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir -Xcgeist -gen-all-sycl-funcs %s -o %t.bc 2>/dev/null
 
 // Test that the LLVMIR generated is verifiable.
 // RUN: opt -verify -disable-output < %t.bc
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/stream-triad.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/stream-triad.cpp
@@ -1,6 +1,6 @@
-// RUN: clang++ -fsycl -fsycl-targets=spir64-unknown-unknown-syclmlir -Xcgeist -gen-all-sycl-funcs %s -o %t.out 2>&1 | FileCheck %s --implicit-check-not="{{error|Error}}:"
+// RUN: clang++ -fsycl -fsycl-targets=spir64-unknown-unknown-syclmlir -O0 -Xcgeist -gen-all-sycl-funcs %s -o %t.out 2>&1 | FileCheck %s --implicit-check-not="{{error|Error}}:"
 
-// RUN: clang++ -fsycl -fsycl-device-only -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir -Xcgeist -gen-all-sycl-funcs %s -o %t.bc 2>/dev/null
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir -Xcgeist -gen-all-sycl-funcs %s -o %t.bc 2>/dev/null
 
 // Test that the LLVMIR generated is verifiable.
 // RUN: opt -verify -disable-output < %t.bc
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/sycl_extern_func.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/sycl_extern_func.cpp
@@ -1,5 +1,5 @@
-// RUN: clang++ -fsycl -fsycl-device-only -emit-mlir %s -o - 2> /dev/null | FileCheck %s --check-prefix=CHECK-MLIR
-// RUN: clang++ -fsycl -fsycl-device-only -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-mlir %s -o - 2> /dev/null | FileCheck %s --check-prefix=CHECK-MLIR
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
 
 #include <sycl/sycl.hpp>
 
@@ -15,10 +15,6 @@
 // CHECK-MLIR:           sycl.constructor(%{{.*}})
 // CHECK-MLIR-NEXT:      return
 
-// CHECK-LLVM-LABEL: define internal spir_func void @_ZZZ16host_single_taskvENKUlRN4sycl3_V17handlerEE_clES2_ENKUlvE_clEv
-// CHECK-LLVM-SAME: #[[FUNCATTRS2:[0-9]+]]
-// CHECK: call void @cons_5()
-
 // CHECK-LLVM-LABEL: define spir_func void @cons_5()
 // CHECK-LLVM-SAME:  #[[FUNCATTRS1:[0-9]+]] {
 // CHECK-LLVM-NEXT:  [[ACCESSOR:%.*]] = alloca %"class.sycl::_V1::accessor.1", align 8
@@ -27,6 +23,7 @@
 
 // CHECK-LLVM-LABEL: define weak_odr spir_kernel void @_ZTSZZ16host_single_taskvENKUlRN4sycl3_V17handlerEE_clES2_E18kernel_single_task
 // CHECK-LLVM-SAME: #[[FUNCATTRS1]]
+// CHECK: call void @cons_5()
 
 extern "C" SYCL_EXTERNAL void cons_5() {
   sycl::accessor<sycl::cl_int, 1, sycl::access::mode::write> accessor;
@@ -47,4 +44,3 @@ void host_single_task() {
 
 // Keep at the end of the file.
 // CHECK-LLVM: attributes #[[FUNCATTRS1]] = { convergent mustprogress norecurse nounwind "frame-pointer"="all" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="{{.*}}/polygeist/tools/cgeist/Test/Verification/sycl/sycl_extern_func.cpp" }
-// CHECK-LLVM-NEXT: attributes #[[FUNCATTRS2]] = { alwaysinline convergent mustprogress norecurse nounwind "frame-pointer"="all" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="{{.*}}/polygeist/tools/cgeist/Test/Verification/sycl/sycl_extern_func.cpp" }
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/types.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/types.cpp
@@ -1,4 +1,4 @@
-// RUN: clang++ -fsycl -fsycl-device-only -emit-mlir -o - %s 2> /dev/null | FileCheck %s
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-mlir -o - %s 2> /dev/null | FileCheck %s
 
 #include <sycl/sycl.hpp>
 
diff --git a/polygeist/tools/cgeist/Test/Verification/sycl/vec_add.cpp b/polygeist/tools/cgeist/Test/Verification/sycl/vec_add.cpp
@@ -1,5 +1,5 @@
-// RUN: clang++ -fsycl -fsycl-device-only -S -emit-mlir %s -o - | FileCheck %s --check-prefix=CHECK-MLIR
-// RUN: clang++ -fsycl -fsycl-device-only -fsycl-targets=spir64-unknown-unknown-syclmlir -S -emit-llvm %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-mlir %s -o - | FileCheck %s --check-prefix=CHECK-MLIR
+// RUN: clang++ -fsycl -fsycl-device-only -O0 -S -emit-llvm -fsycl-targets=spir64-unknown-unknown-syclmlir %s -o - | FileCheck %s --check-prefix=CHECK-LLVM
 
 #include <sycl/sycl.hpp>
 #define N 32
@@ -16,15 +16,12 @@
 // CHECK-MLIR-NEXT: [[RESULT:%.*]] = arith.addf [[V1]], [[V2]] : f32
 // CHECK-MLIR:      affine.store [[RESULT]], {{.*}}[0] : memref<?xf32, 4>
 
-// CHECK-LLVM:       define internal spir_func void [[FUNC:@.*vec_add_device_simple.*_]]({{.*}}) #[[FUNCATTRS2:[0-9]+]]
+// CHECK-LLVM:       define weak_odr spir_kernel void @{{.*}}vec_add_simple({{.*}}) #[[FUNCATTRS1:[0-9]+]]
 // CHECK-LLVM-DAG:   [[V1:%.*]] = load float, float addrspace(4)* {{.*}}, align 4
 // CHECK-LLVM-DAG:   [[V2:%.*]] = load float, float addrspace(4)* {{.*}}, align 4
 // CHECK-LLVM:       [[RESULT:%.*]] = fadd float [[V1]], [[V2]]
 // CHECK-LLVM:       store float [[RESULT]], float addrspace(4)* {{.*}}, align 4
 
-// CHECK-LLVM:       define weak_odr spir_kernel void @{{.*}}vec_add_simple({{.*}}) #[[FUNCATTRS1:[0-9]+]]
-// CHECK-LLVM:       call void [[FUNC]]
-
 void vec_add_device_simple(std::array<float, N> &VA, std::array<float, N> &VB,
                            std::array<float, N> &VC) {
   auto q = sycl::queue{};
@@ -96,4 +93,3 @@ int main() {
 
 // Keep at the end of the file.
 // CHECK-LLVM-DAG: attributes #[[FUNCATTRS1]] = { convergent mustprogress norecurse nounwind "frame-pointer"="all" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="{{.*}}/polygeist/tools/cgeist/Test/Verification/sycl/vec_add.cpp" }
-// CHECK-LLVM-DAG: attributes #[[FUNCATTRS2]] = { alwaysinline convergent mustprogress norecurse nounwind "frame-pointer"="all" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="{{.*}}/polygeist/tools/cgeist/Test/Verification/sycl/vec_add.cpp" }
diff --git a/polygeist/tools/cgeist/driver.cc b/polygeist/tools/cgeist/driver.cc
@@ -700,11 +700,6 @@ static int createAndExecutePassPipeline(
 static void
 runOptimizationPipeline(llvm::Module &module,
                         const llvm::OptimizationLevel &OptimizationLevel) {
-  if (OptimizationLevel == llvm::OptimizationLevel::O0) {
-    // No optimizations should be run in this case.
-    return;
-  }
-
   llvm::LoopAnalysisManager LAM;
   llvm::FunctionAnalysisManager FAM;
   llvm::CGSCCAnalysisManager CGAM;
@@ -719,7 +714,9 @@ runOptimizationPipeline(llvm::Module &module,
   PB.crossRegisterProxies(LAM, FAM, CGAM, MAM);
 
   llvm::ModulePassManager MPM =
-      PB.buildPerModuleDefaultPipeline(OptimizationLevel);
+      (OptimizationLevel == llvm::OptimizationLevel::O0)
+          ? PB.buildO0DefaultPipeline(OptimizationLevel)
+          : PB.buildPerModuleDefaultPipeline(OptimizationLevel);
 
   MPM.run(module, MAM);
 }

Original file line number	Diff line number	Diff line change
`@@ -1,4 +1,4 @@`
`1`		`-// RUN: clang++ -fsycl -fsycl-device-only -emit-mlir -o - %s 2> /dev/null \| FileCheck %s`
	`1`	`+// RUN: clang++ -fsycl -fsycl-device-only -O0 -emit-mlir -o - %s 2> /dev/null \| FileCheck %s`
`2`	`2`
`3`	`3`	`#include <sycl/sycl.hpp>`
`4`	`4`