439 lines · cpp
1// RUN: %clang_cc1 %s -fopenacc -verify2 3constexpr int three() { return 3; }4constexpr int one() { return 1; }5constexpr int neg() { return -1; }6constexpr int zero() { return 0; }7 8struct NotConstexpr {9 constexpr NotConstexpr(){};10 11 operator int(){ return 1; }12};13struct ConvertsNegative {14 constexpr ConvertsNegative(){};15 16 constexpr operator int(){ return -1; }17};18struct ConvertsOne{19 constexpr ConvertsOne(){};20 21 constexpr operator int(){ return 1; }22};23 24struct ConvertsThree{25 constexpr ConvertsThree(){};26 27 constexpr operator int(){ return 3; }28};29 30template<typename T, int Val>31void negative_zero_constexpr_templ() {32 // expected-error@+1 2{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to 0}}33#pragma acc loop tile(*, T{})34 for(int i = 0; i < 5; ++i)35 for(int j = 0; j < 5; ++j);36 37 // expected-error@+1{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to -1}}38#pragma acc loop tile(Val, *)39 for(int i = 0; i < 5; ++i)40 for(int j = 0; j < 5; ++j);41 42 // expected-error@+1{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to 0}}43#pragma acc loop tile(zero(), *)44 for(int i = 0; i < 5; ++i)45 for(int j = 0; j < 5; ++j);46}47 48void negative_zero_constexpr() {49 negative_zero_constexpr_templ<int, 1>(); // expected-note{{in instantiation of function template specialization}}50 negative_zero_constexpr_templ<int, -1>(); // expected-note{{in instantiation of function template specialization}}51 52 // expected-error@+1{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to 0}}53#pragma acc loop tile(0, *)54 for(int i = 0; i < 5; ++i)55 for(int j = 0; j < 5; ++j);56 57 // expected-error@+1{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to 0}}58#pragma acc loop tile(1, 0)59 for(int i = 0; i < 5; ++i)60 for(int j = 0; j < 5; ++j);61 62 // expected-error@+1{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to -1}}63#pragma acc loop tile(1, -1)64 for(int i = 0; i < 5; ++i)65 for(int j = 0; j < 5; ++j);66 67 // expected-error@+1{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to -1}}68#pragma acc loop tile(-1, 0)69 for(int i = 0; i < 5; ++i)70 for(int j = 0; j < 5; ++j);71 72 // expected-error@+1{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to 0}}73#pragma acc loop tile(zero(), 0)74 for(int i = 0; i < 5; ++i)75 for(int j = 0; j < 5; ++j);76 77 // expected-error@+1{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to -1}}78#pragma acc loop tile(1, neg())79 for(int i = 0; i < 5; ++i)80 for(int j = 0; j < 5; ++j);81 82 // expected-error@+1{{OpenACC 'tile' clause size expression must be an asterisk or a constant expression}}83#pragma acc loop tile(NotConstexpr{})84 for(int i = 0; i < 5; ++i);85 86 // expected-error@+1{{OpenACC 'tile' clause size expression must be positive integer value, evaluated to -1}}87#pragma acc loop tile(1, ConvertsNegative{})88 for(int i = 0; i < 5; ++i)89 for(int j = 0; j < 5; ++j);90 91#pragma acc loop tile(*, ConvertsOne{})92 for(int i = 0; i < 5; ++i)93 for(int j = 0; j < 5; ++j);94}95 96template<unsigned One>97void only_for_loops_templ() {98 // expected-note@+1{{'loop' construct is here}}99#pragma acc loop tile(One)100 // expected-error@+1{{OpenACC 'loop' construct can only be applied to a 'for' loop}}101 while(true);102 103 // expected-note@+1{{'loop' construct is here}}104#pragma acc loop tile(One)105 // expected-error@+1{{OpenACC 'loop' construct can only be applied to a 'for' loop}}106 do {} while(true);107 108 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}109#pragma acc loop tile(One, 2) // expected-note 2{{active 'tile' clause defined here}}110 for(int i = 0; i < 5; ++i)111 // expected-error@+1{{while loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}112 while(true);113 114 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}115#pragma acc loop tile(One, 2) // expected-note 2{{active 'tile' clause defined here}}116 for(int i = 0; i < 5; ++i)117 // expected-error@+1{{do loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}118 do{}while(true);119}120 121 122void only_for_loops() {123 // expected-note@+1{{'loop' construct is here}}124#pragma acc loop tile(1)125 // expected-error@+1{{OpenACC 'loop' construct can only be applied to a 'for' loop}}126 while(true);127 128 // expected-note@+1{{'loop' construct is here}}129#pragma acc loop tile(1)130 // expected-error@+1{{OpenACC 'loop' construct can only be applied to a 'for' loop}}131 do {} while(true);132 133 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}134#pragma acc loop tile(1, 2) // expected-note 2{{active 'tile' clause defined here}}135 for(int i = 0; i < 5; ++i)136 // expected-error@+1{{while loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}137 while(true);138 139 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}140#pragma acc loop tile(1, 2) // expected-note 2{{active 'tile' clause defined here}}141 for(int i = 0; i < 5; ++i)142 // expected-error@+1{{do loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}143 do{}while(true);144}145 146void only_one_on_loop() {147 // expected-error@+2{{OpenACC 'tile' clause cannot appear more than once on a 'loop' directive}}148 // expected-note@+1{{previous 'tile' clause is here}}149#pragma acc loop tile(1) tile(1)150 for(int i = 0; i < 5; ++i);151}152 153template<unsigned Val>154void depth_too_high_templ() {155 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}156#pragma acc loop tile (Val, *, Val) // expected-note{{active 'tile' clause defined here}}157 for(int i = 0; i < 5; ++i)158 for(int j = 0; j < 5; ++j);159 160 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}161#pragma acc loop tile (Val, *, Val) // expected-note 2{{active 'tile' clause defined here}}162 for(int i = 0; i < 5; ++i)163 for(int j = 0; j < 5; ++j)164 // expected-error@+1{{while loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}165 while(true);166 167 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}168#pragma acc loop tile (Val, *, Val) // expected-note 2{{active 'tile' clause defined here}}169 for(int i = 0; i < 5; ++i)170 for(int j = 0; j < 5; ++j)171 // expected-error@+1{{do loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}172 do{}while(true);173 174 int Arr[Val+5];175 176 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}177#pragma acc loop tile (Val, *, Val) // expected-note 2{{active 'tile' clause defined here}}178 for(int i = 0; i < 5; ++i)179 for(auto x : Arr)180 // expected-error@+1{{while loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}181 while(true)182 for(int j = 0; j < 5; ++j);183 184#pragma acc loop tile (Val, *, Val)185 for(int i = 0; i < 5; ++i)186 for(auto x : Arr)187 for(int j = 0; j < 5; ++j)188 while(true);189}190 191void depth_too_high() {192 depth_too_high_templ<3>();193 194int Arr[5];195 196 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}197#pragma acc loop tile (1, *, 3) // expected-note{{active 'tile' clause defined here}}198 for(int i = 0; i < 5; ++i)199 for(int j = 0; j < 5; ++j);200 201 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}202#pragma acc loop tile (1, *, 3) // expected-note 2{{active 'tile' clause defined here}}203 for(int i = 0; i < 5; ++i)204 for(int j = 0; j < 5; ++j)205 // expected-error@+1{{while loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}206 while(true);207 208 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}209#pragma acc loop tile (1, *, 3) // expected-note 2{{active 'tile' clause defined here}}210 for(int i = 0; i < 5; ++i)211 for(int j = 0; j < 5; ++j)212 // expected-error@+1{{do loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}213 do{}while(true);214 215 // expected-error@+1{{'tile' clause specifies a loop count greater than the number of available loops}}216#pragma acc loop tile (1, *, 3) // expected-note 2{{active 'tile' clause defined here}}217 for(int i = 0; i < 5; ++i)218 for(int j = 0; j < 5; ++j)219 // expected-error@+1{{while loop cannot appear in intervening code of a 'loop' with a 'tile' clause}}220 while(true)221 for(int j = 0; j < 5; ++j);222 223#pragma acc loop tile (1, *, 3)224 for(int i = 0; i < 5; ++i)225 for(auto x : Arr)226 for(int j = 0; j < 5; ++j)227 while(true);228}229 230template<unsigned Val>231void not_single_loop_templ() {232 233 int Arr[Val];234 235#pragma acc loop tile (Val, *, 3) // expected-note{{active 'tile' clause defined here}}236 for(int i = 0; i < 5; ++i) {237 for (auto x : Arr)238 for(int k = 0; k < 5; ++k);239 // expected-error@+1{{more than one for-loop in a loop associated with OpenACC 'loop' construct with a 'tile' clause}}240 for(int j = 0; j < 5; ++j)241 for(int k = 0; k < 5; ++k);242 }243}244 245void not_single_loop() {246 not_single_loop_templ<3>(); // no diagnostic, was diagnosed in phase 1.247 248 int Arr[5];249 250#pragma acc loop tile (1, *, 3)// expected-note{{active 'tile' clause defined here}}251 for(int i = 0; i < 5; ++i) {252 for (auto x : Arr)253 for(int k = 0; k < 5; ++k);254 // expected-error@+1{{more than one for-loop in a loop associated with OpenACC 'loop' construct with a 'tile' clause}}255 for(int j = 0; j < 5; ++j)256 for(int k = 0; k < 5; ++k);257 }258}259 260template<unsigned Val>261void no_other_directives_templ() {262 263 int Arr[Val];264 265#pragma acc loop tile (Val, *, 3) // expected-note{{active 'tile' clause defined here}}266 for(int i = 0; i < 5; ++i) {267 for (auto x : Arr) {268 // expected-error@+1{{OpenACC 'serial' construct cannot appear in intervening code of a 'loop' with a 'tile' clause}}269#pragma acc serial270 ;271 for(int j = 0; j < 5; ++j);272 }273 }274 275 // OK, in innermost276#pragma acc loop tile (Val, *, 3)277 for(int i = 0; i < 5; ++i) {278 for(int j = 0; j < 5; ++j) {279 for (auto x : Arr) {280#pragma acc serial281 ;282 }283 }284 }285}286 287void no_other_directives() {288 no_other_directives_templ<3>();289 int Arr[5];290 291#pragma acc loop tile (1, *, 3) // expected-note{{active 'tile' clause defined here}}292 for(int i = 0; i < 5; ++i) {293 for (auto x : Arr) {294 // expected-error@+1{{OpenACC 'serial' construct cannot appear in intervening code of a 'loop' with a 'tile' clause}}295#pragma acc serial296 ;297 for(int j = 0; j < 5; ++j);298 }299 }300 301 // OK, in innermost302#pragma acc loop tile (3, *, 3)303 for(int i = 0; i < 5; ++i) {304 for(int j = 0; j < 5; ++j) {305 for (auto x : Arr) {306#pragma acc serial307 ;308 }309 }310 }311}312 313void call();314template<unsigned Val>315void intervening_templ() {316#pragma acc loop tile(1, Val, *) // expected-note{{active 'tile' clause defined here}}317 for(int i = 0; i < 5; ++i) {318 //expected-error@+1{{inner loops must be tightly nested inside a 'tile' clause on a 'loop' construct}}319 call();320 for(int j = 0; j < 5; ++j)321 for(int k = 0; k < 5; ++k);322 }323 324#pragma acc loop tile(1, Val, *) // expected-note{{active 'tile' clause defined here}}325 for(int i = 0; i < 5; ++i) {326 //expected-error@+1{{inner loops must be tightly nested inside a 'tile' clause on a 'loop' construct}}327 unsigned I;328 for(int j = 0; j < 5; ++j)329 for(int k = 0; k < 5; ++k);330 }331 332#pragma acc loop tile(1, Val, *)333 // expected-error@+2{{OpenACC 'loop' construct must have a terminating condition}}334 // expected-note@-2{{'loop' construct is here}}335 for(int i = 0;;++i) {336 // expected-error@+2{{OpenACC 'loop' construct must have a terminating condition}}337 // expected-note@-5{{'loop' construct is here}}338 for(int j = 0;;++j)339 // expected-error@+2{{OpenACC 'loop' construct must have a terminating condition}}340 // expected-note@-8{{'loop' construct is here}}341 for(int k = 0;;++k)342 call();343 }344}345 346void intervening() {347 intervening_templ<3>();348 349#pragma acc loop tile(1, 2, *) // expected-note{{active 'tile' clause defined here}}350 for(int i = 0; i < 5; ++i) {351 //expected-error@+1{{inner loops must be tightly nested inside a 'tile' clause on a 'loop' construct}}352 call();353 for(int j = 0; j < 5; ++j)354 for(int k = 0; k < 5; ++k);355 }356 357#pragma acc loop tile(1, 2, *) // expected-note{{active 'tile' clause defined here}}358 for(int i = 0; i < 5; ++i) {359 //expected-error@+1{{inner loops must be tightly nested inside a 'tile' clause on a 'loop' construct}}360 unsigned I;361 for(int j = 0; j < 5; ++j)362 for(int k = 0; k < 5; ++k);363 }364 365#pragma acc loop tile(1, 2, *)366 for(int i = 0; i < 5; ++i) {367 for(int j = 0; j < 5; ++j)368 for(int k = 0; k < 5; ++k)369 call();370 }371 372#pragma acc loop tile(1, 2, *)373 // expected-error@+2{{OpenACC 'loop' construct must have a terminating condition}}374 // expected-note@-2{{'loop' construct is here}}375 for(int i = 0;;++i) {376 // expected-error@+2{{OpenACC 'loop' construct must have a terminating condition}}377 // expected-note@-5{{'loop' construct is here}}378 for(int j = 0;;++j)379 // expected-error@+2{{OpenACC 'loop' construct must have a terminating condition}}380 // expected-note@-8{{'loop' construct is here}}381 for(int k = 0;;++k)382 for(;;)383 call();384 }385}386 387void use_largest_tile() {388// expected-error@+2{{'tile' clause specifies a loop count greater than the number of available loops}}389// expected-note@+1{{active 'tile' clause defined here}}390#pragma acc loop tile(1,2) device_type(*) tile (3,4,5)391 for(int i = 0; i < 5; ++i)392 for (int j = 0; j < 5; ++j);393 394// expected-error@+2{{'tile' clause specifies a loop count greater than the number of available loops}}395// expected-note@+1{{active 'tile' clause defined here}}396#pragma acc loop tile (3,4,5) device_type(*) tile(1,2)397 for(int i = 0; i < 5; ++i)398 for (int j = 0; j < 5; ++j);399}400 401void collapse_tile_depth() {402 // expected-error@+4{{'collapse' clause specifies a loop count greater than the number of available loops}}403 // expected-note@+3{{active 'collapse' clause defined here}}404 // expected-error@+2{{'tile' clause specifies a loop count greater than the number of available loops}}405 // expected-note@+1{{active 'tile' clause defined here}}406#pragma acc loop tile(1, 2, 3) collapse (3)407 for(int i = 0; i < 5;++i) {408 for(int j = 0; j < 5; ++j);409 }410}411void no_dupes_since_last_device_type() {412 // expected-error@+3{{OpenACC 'tile' clause cannot appear more than once in a 'device_type' region on a 'loop' directive}}413 // expected-note@+2{{previous 'tile' clause is here}}414 // expected-note@+1{{active 'device_type' clause here}}415#pragma acc loop tile(1) device_type(*) tile(1) tile(2)416 for(unsigned i = 0; i < 5; ++i)417 for(unsigned j = 0; j < 5; ++j);418 419#pragma acc loop tile(1) device_type(*) tile(1) device_type(nvidia) tile(2)420 for(unsigned i = 0; i < 5; ++i)421 for(unsigned j = 0; j < 5; ++j);422 423 // expected-error@+4{{OpenACC 'tile' clause applies to 'device_type' 'nvidiA', which conflicts with previous 'tile' clause}}424 // expected-note@+3{{active 'device_type' clause here}}425 // expected-note@+2{{previous 'tile' clause is here}}426 // expected-note@+1{{which applies to 'device_type' clause here}}427#pragma acc loop device_type(nvidia, radeon) tile(1) device_type(nvidiA) tile(2)428 for(unsigned i = 0; i < 5; ++i)429 for(unsigned j = 0; j < 5; ++j);430 431 // expected-error@+4{{OpenACC 'tile' clause applies to 'device_type' 'radeon', which conflicts with previous 'tile' clause}}432 // expected-note@+3{{active 'device_type' clause here}}433 // expected-note@+2{{previous 'tile' clause is here}}434 // expected-note@+1{{which applies to 'device_type' clause here}}435#pragma acc loop device_type(radeon) tile(1) device_type(nvidia, radeon) tile(2)436 for(unsigned i = 0; i < 5; ++i)437 for(unsigned j = 0; j < 5; ++j);438}439