Mercurial > hg > CbC > CbC_llvm

annotate clang/test/CodeGenCUDA/struct-mangling-number.cu @ 266:00f31e85ec16 default tip

Find changesets by keywords (author, files, the commit message), revision number or hash, or revset expression.
Added tag current for changeset 31d058e83c98
author Shinji KONO <kono@ie.u-ryukyu.ac.jp>
date Sat, 14 Oct 2023 10:13:55 +0900
parents c4bab56944e8
children
Ignore whitespace changes - Everywhere: Within whitespace: At end of lines:
rev   line source
236
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
1 // RUN: %clang_cc1 -emit-llvm -o - -aux-triple x86_64-pc-windows-msvc \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
2 // RUN: -fms-extensions -triple amdgcn-amd-amdhsa \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
3 // RUN: -target-cpu gfx1030 -fcuda-is-device -x hip %s \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
4 // RUN: | FileCheck -check-prefix=DEV %s
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
5
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
6 // RUN: %clang_cc1 -emit-llvm -o - -triple x86_64-pc-windows-msvc \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
7 // RUN: -fms-extensions -aux-triple amdgcn-amd-amdhsa \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
8 // RUN: -aux-target-cpu gfx1030 -x hip %s \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
9 // RUN: | FileCheck -check-prefix=HOST %s
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
10
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
11 // RUN: %clang_cc1 -emit-llvm -o - -triple x86_64-pc-windows-msvc \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
12 // RUN: -fms-extensions -aux-triple amdgcn-amd-amdhsa \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
13 // RUN: -aux-target-cpu gfx1030 -x hip %s \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
14 // RUN: | FileCheck -check-prefix=HOST-NEG %s
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
15
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
16 // RUN: %clang_cc1 -emit-llvm -o - -triple x86_64-pc-windows-msvc \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
17 // RUN: -fms-extensions -x c++ %s \
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
18 // RUN: | FileCheck -check-prefix=CPP %s
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
19
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
20 #if __HIP__
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
21 #include "Inputs/cuda.h"
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
22 #endif
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
23
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
24 // Check local struct 'Op' uses Itanium mangling number instead of MSVC mangling
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
25 // number in device side name mangling. It is the same in device and host
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
26 // compilation.
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
27
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
28 // DEV: define amdgpu_kernel void @_Z6kernelIZN4TestIiE3runEvE2OpEvv(
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
29
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
30 // HOST-DAG: @{{.*}} = {{.*}}c"_Z6kernelIZN4TestIiE3runEvE2OpEvv\00"
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
31
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
32 // HOST-NEG-NOT: @{{.*}} = {{.*}}c"_Z6kernelIZN4TestIiE3runEvE2Op_1Evv\00"
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
33 #if __HIP__
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
34 template<typename T>
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
35 __attribute__((global)) void kernel()
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
36 {
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
37 }
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
38 #endif
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
39
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
40 // Check local struct 'Op' uses MSVC mangling number in host function name mangling.
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
41 // It is the same when compiled as HIP or C++ program.
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
42
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
43 // HOST-DAG: call void @"??$fun@UOp@?2??run@?$Test@H@@QEAAXXZ@@@YAXXZ"()
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
44 // CPP: call void @"??$fun@UOp@?2??run@?$Test@H@@QEAAXXZ@@@YAXXZ"()
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
45 template<typename T>
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
46 void fun()
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
47 {
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
48 }
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
49
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
50 template <typename T>
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
51 class Test {
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
52 public:
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
53 void run()
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
54 {
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
55 struct Op
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
56 {
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
57 };
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
58 #if __HIP__
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
59 kernel<Op><<<1, 1>>>();
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
60 #endif
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
61 fun<Op>();
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
62 }
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
63 };
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
64
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
65 int main() {
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
66 Test<int> A;
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
67 A.run();
c4bab56944e8 LLVM 16
kono
parents:
diff changeset
68 }