[Openmp-commits] [openmp] 5eb7a42 - [Attributor][NFC] Precommit tests

[Johannes Doerfert via Openmp-commits]

Author: Johannes Doerfert
Date: 2023-08-17T22:42:38-07:00
New Revision: 5eb7a427b057531b03df6f96d65e013a88c536c2

URL: https://github.com/llvm/llvm-project/commit/5eb7a427b057531b03df6f96d65e013a88c536c2
DIFF: https://github.com/llvm/llvm-project/commit/5eb7a427b057531b03df6f96d65e013a88c536c2.diff

LOG: [Attributor][NFC] Precommit tests

Added: 
    openmp/libomptarget/test/jit/type_punning.c

Modified: 
    llvm/test/Transforms/Attributor/callgraph.ll

Removed: 
    


################################################################################
diff  --git a/llvm/test/Transforms/Attributor/callgraph.ll b/llvm/test/Transforms/Attributor/callgraph.ll
index 61c6ce92491fab..6dddd84452dabd 100644
--- a/llvm/test/Transforms/Attributor/callgraph.ll
+++ b/llvm/test/Transforms/Attributor/callgraph.ll
@@ -49,6 +49,65 @@ define void @func5(i32 %0) {
   ret void
 }
 
+define i32 @musttailCall(i32 %0) {
+; CHECK-LABEL: @musttailCall(
+; CHECK-NEXT:    [[TMP2:%.*]] = icmp ne i32 [[TMP0:%.*]], 0
+; CHECK-NEXT:    [[TMP3:%.*]] = select i1 [[TMP2]], ptr @func4, ptr @func3
+; CHECK-NEXT:    [[C:%.*]] = musttail call i32 [[TMP3]](i32 0)
+; CHECK-NEXT:    ret i32 [[C]]
+;
+  %2 = icmp ne i32 %0, 0
+  %3 = select i1 %2, ptr @func4, ptr @func3
+  %c = musttail call i32 (i32) %3(i32 0)
+  ret i32 %c
+}
+
+declare i32 @retI32()
+declare void @takeI32(i32)
+declare float @retFloatTakeFloat(float)
+declare void @void()
+
+define i32 @non_matching_fp1(i1 %c1, i1 %c2, i1 %c) {
+; CHECK-LABEL: @non_matching_fp1(
+; CHECK-NEXT:    [[FP1:%.*]] = select i1 [[C1:%.*]], ptr @retI32, ptr @takeI32
+; CHECK-NEXT:    [[FP2:%.*]] = select i1 [[C2:%.*]], ptr @retFloatTakeFloat, ptr @void
+; CHECK-NEXT:    [[FP:%.*]] = select i1 [[C:%.*]], ptr [[FP1]], ptr [[FP2]]
+; CHECK-NEXT:    [[CALL:%.*]] = call i32 [[FP]](i32 42)
+; CHECK-NEXT:    ret i32 [[CALL]]
+;
+  %fp1 = select i1 %c1, ptr @retI32, ptr @takeI32
+  %fp2 = select i1 %c2, ptr @retFloatTakeFloat, ptr @void
+  %fp = select i1 %c, ptr %fp1, ptr %fp2
+  %call = call i32 %fp(i32 42)
+  ret i32 %call
+}
+
+define void @non_matching_fp2(i1 %c1, i1 %c2, i1 %c, ptr %unknown) {
+; CHECK-LABEL: @non_matching_fp2(
+; CHECK-NEXT:    [[FP1:%.*]] = select i1 [[C1:%.*]], ptr @retI32, ptr @takeI32
+; CHECK-NEXT:    [[FP2:%.*]] = select i1 [[C2:%.*]], ptr @retFloatTakeFloat, ptr [[UNKNOWN:%.*]]
+; CHECK-NEXT:    [[FP:%.*]] = select i1 [[C:%.*]], ptr [[FP1]], ptr [[FP2]]
+; CHECK-NEXT:    call void [[FP]]()
+; CHECK-NEXT:    ret void
+;
+  %fp1 = select i1 %c1, ptr @retI32, ptr @takeI32
+  %fp2 = select i1 %c2, ptr @retFloatTakeFloat, ptr %unknown
+  %fp = select i1 %c, ptr %fp1, ptr %fp2
+  call void %fp()
+  ret void
+}
+
+define i32 @non_matching_unknown(i1 %c, ptr %fn) {
+; CHECK-LABEL: @non_matching_unknown(
+; CHECK-NEXT:    [[FP:%.*]] = select i1 [[C:%.*]], ptr @retI32, ptr [[FN:%.*]]
+; CHECK-NEXT:    [[CALL:%.*]] = call i32 [[FP]](i32 42)
+; CHECK-NEXT:    ret i32 [[CALL]]
+;
+  %fp = select i1 %c, ptr @retI32, ptr %fn
+  %call = call i32 %fp(i32 42)
+  ret i32 %call
+}
+
 define void @broker(ptr %unknown) !callback !0 {
 ; CHECK-LABEL: @broker(
 ; CHECK-NEXT:    call void [[UNKNOWN:%.*]]()

diff  --git a/openmp/libomptarget/test/jit/type_punning.c b/openmp/libomptarget/test/jit/type_punning.c
new file mode 100644
index 00000000000000..8531a84f74ec1a
--- /dev/null
+++ b/openmp/libomptarget/test/jit/type_punning.c
@@ -0,0 +1,39 @@
+// clang-format off
+//
+// RUN: %libomptarget-compileopt-generic -fopenmp-target-jit
+// RUN: env LIBOMPTARGET_JIT_PRE_OPT_IR_MODULE=%t.pre.ll     \
+// RUN:     LIBOMPTARGET_JIT_SKIP_OPT=true                   \
+// RUN:     %libomptarget-run-generic
+// RUN: %fcheck-plain-generic --input-file %t.pre.ll %s
+//
+// clang-format on
+
+// UNSUPPORTED: aarch64-unknown-linux-gnu
+// UNSUPPORTED: aarch64-unknown-linux-gnu-LTO
+// UNSUPPORTED: x86_64-pc-linux-gnu
+// UNSUPPORTED: x86_64-pc-linux-gnu-LTO
+
+// FIXME: We want that there is only the kernel function left, not two outlined
+// parallel regions.
+//
+// CHECK: define
+// CHECK: define
+// CHECK: define
+
+#include <omp.h>
+void f(long *A, int N) {
+  long i = 0;
+#pragma omp target map(A[ : N])
+  {
+#pragma omp parallel firstprivate(i)
+    A[omp_get_thread_num()] = i;
+#pragma omp parallel firstprivate(i, N)
+    A[omp_get_thread_num()] += i + N;
+  }
+}
+
+int main() {
+  long A[1];
+  f(&A[0], 1);
+  return A[0];
+}