1 // RUN: %clang_cc1 -fsyntax-only -verify %s
3 #include "Inputs/cuda.h"
6 // expected-error@+2 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
7 __attribute__((amdgpu_flat_work_group_size(32, 64)))
8 __global__ void flat_work_group_size_32_64() {}
10 // expected-error@+2 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
11 __attribute__((amdgpu_waves_per_eu(2)))
12 __global__ void waves_per_eu_2() {}
14 // expected-error@+2 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
15 __attribute__((amdgpu_waves_per_eu(2, 4)))
16 __global__ void waves_per_eu_2_4() {}
18 // expected-error@+2 {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
19 __attribute__((amdgpu_num_sgpr(32)))
20 __global__ void num_sgpr_32() {}
22 // expected-error@+2 {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
23 __attribute__((amdgpu_num_vgpr(64)))
24 __global__ void num_vgpr_64() {}
27 // expected-error@+3 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
28 // fixme-expected-error@+2 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
29 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_waves_per_eu(2)))
30 __global__ void flat_work_group_size_32_64_waves_per_eu_2() {}
32 // expected-error@+3 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
33 // fixme-expected-error@+2 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
34 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_waves_per_eu(2, 4)))
35 __global__ void flat_work_group_size_32_64_waves_per_eu_2_4() {}
37 // expected-error@+3 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
38 // fixme-expected-error@+2 {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
39 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_num_sgpr(32)))
40 __global__ void flat_work_group_size_32_64_num_sgpr_32() {}
42 // expected-error@+3 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
43 // fixme-expected-error@+2 {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
44 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_num_vgpr(64)))
45 __global__ void flat_work_group_size_32_64_num_vgpr_64() {}
47 // expected-error@+3 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
48 // fixme-expected-error@+2 {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
49 __attribute__((amdgpu_waves_per_eu(2), amdgpu_num_sgpr(32)))
50 __global__ void waves_per_eu_2_num_sgpr_32() {}
52 // expected-error@+3 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
53 // fixme-expected-error@+2 {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
54 __attribute__((amdgpu_waves_per_eu(2), amdgpu_num_vgpr(64)))
55 __global__ void waves_per_eu_2_num_vgpr_64() {}
57 // expected-error@+3 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
58 // fixme-expected-error@+2 {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
59 __attribute__((amdgpu_waves_per_eu(2, 4), amdgpu_num_sgpr(32)))
60 __global__ void waves_per_eu_2_4_num_sgpr_32() {}
62 // expected-error@+3 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
63 // fixme-expected-error@+2 {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
64 __attribute__((amdgpu_waves_per_eu(2, 4), amdgpu_num_vgpr(64)))
65 __global__ void waves_per_eu_2_4_num_vgpr_64() {}
67 // expected-error@+3 {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
68 // fixme-expected-error@+2 {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
69 __attribute__((amdgpu_num_sgpr(32), amdgpu_num_vgpr(64)))
70 __global__ void num_sgpr_32_num_vgpr_64() {}
73 // expected-error@+4 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
74 // fixme-expected-error@+3 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
75 // fixme-expected-error@+2 {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
76 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_waves_per_eu(2), amdgpu_num_sgpr(32)))
77 __global__ void flat_work_group_size_32_64_waves_per_eu_2_num_sgpr_32() {}
79 // expected-error@+4 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
80 // fixme-expected-error@+3 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
81 // fixme-expected-error@+2 {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
82 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_waves_per_eu(2), amdgpu_num_vgpr(64)))
83 __global__ void flat_work_group_size_32_64_waves_per_eu_2_num_vgpr_64() {}
85 // expected-error@+4 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
86 // fixme-expected-error@+3 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
87 // fixme-expected-error@+2 {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
88 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_waves_per_eu(2, 4), amdgpu_num_sgpr(32)))
89 __global__ void flat_work_group_size_32_64_waves_per_eu_2_4_num_sgpr_32() {}
91 // expected-error@+4 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
92 // fixme-expected-error@+3 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
93 // fixme-expected-error@+2 {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
94 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_waves_per_eu(2, 4), amdgpu_num_vgpr(64)))
95 __global__ void flat_work_group_size_32_64_waves_per_eu_2_4_num_vgpr_64() {}
98 // expected-error@+5 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
99 // fixme-expected-error@+4 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
100 // fixme-expected-error@+3 {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
101 // fixme-expected-error@+2 {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
102 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_waves_per_eu(2), amdgpu_num_sgpr(32), amdgpu_num_vgpr(64)))
103 __global__ void flat_work_group_size_32_64_waves_per_eu_2_num_sgpr_32_num_vgpr_64() {}
105 // expected-error@+5 {{'amdgpu_flat_work_group_size' attribute only applies to kernel functions}}
106 // fixme-expected-error@+4 {{'amdgpu_waves_per_eu' attribute only applies to kernel functions}}
107 // fixme-expected-error@+3 {{'amdgpu_num_sgpr' attribute only applies to kernel functions}}
108 // fixme-expected-error@+2 {{'amdgpu_num_vgpr' attribute only applies to kernel functions}}
109 __attribute__((amdgpu_flat_work_group_size(32, 64), amdgpu_waves_per_eu(2, 4), amdgpu_num_sgpr(32), amdgpu_num_vgpr(64)))
110 __global__ void flat_work_group_size_32_64_waves_per_eu_2_4_num_sgpr_32_num_vgpr_64() {}