1 /* General types and functions that are uselful for processing of OpenMP,
2 OpenACC and similar directivers at various stages of compilation.
3
4 Copyright (C) 2005-2022 Free Software Foundation, Inc.
5
6 This file is part of GCC.
7
8 GCC is free software; you can redistribute it and/or modify it under
9 the terms of the GNU General Public License as published by the Free
10 Software Foundation; either version 3, or (at your option) any later
11 version.
12
13 GCC is distributed in the hope that it will be useful, but WITHOUT ANY
14 WARRANTY; without even the implied warranty of MERCHANTABILITY or
15 FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License
16 for more details.
17
18 You should have received a copy of the GNU General Public License
19 along with GCC; see the file COPYING3. If not see
20 <http://www.gnu.org/licenses/>. */
21
22 /* Find an OMP clause of type KIND within CLAUSES. */
23
24 #include "config.h"
25 #include "system.h"
26 #include "coretypes.h"
27 #include "backend.h"
28 #include "target.h"
29 #include "tree.h"
30 #include "gimple.h"
31 #include "ssa.h"
32 #include "diagnostic-core.h"
33 #include "fold-const.h"
34 #include "langhooks.h"
35 #include "omp-general.h"
36 #include "stringpool.h"
37 #include "attribs.h"
38 #include "gimplify.h"
39 #include "cgraph.h"
40 #include "alloc-pool.h"
41 #include "symbol-summary.h"
42 #include "tree-pass.h"
43 #include "omp-device-properties.h"
44 #include "tree-iterator.h"
45 #include "data-streamer.h"
46 #include "streamer-hooks.h"
47 #include "opts.h"
48
49 enum omp_requires omp_requires_mask;
50
51 tree
omp_find_clause(tree clauses,enum omp_clause_code kind)52 omp_find_clause (tree clauses, enum omp_clause_code kind)
53 {
54 for (; clauses ; clauses = OMP_CLAUSE_CHAIN (clauses))
55 if (OMP_CLAUSE_CODE (clauses) == kind)
56 return clauses;
57
58 return NULL_TREE;
59 }
60
61 /* True if OpenMP should regard this DECL as being a scalar which has Fortran's
62 allocatable or pointer attribute. */
63 bool
omp_is_allocatable_or_ptr(tree decl)64 omp_is_allocatable_or_ptr (tree decl)
65 {
66 return lang_hooks.decls.omp_is_allocatable_or_ptr (decl);
67 }
68
69 /* Check whether this DECL belongs to a Fortran optional argument.
70 With 'for_present_check' set to false, decls which are optional parameters
71 themselve are returned as tree - or a NULL_TREE otherwise. Those decls are
72 always pointers. With 'for_present_check' set to true, the decl for checking
73 whether an argument is present is returned; for arguments with value
74 attribute this is the hidden argument and of BOOLEAN_TYPE. If the decl is
75 unrelated to optional arguments, NULL_TREE is returned. */
76
77 tree
omp_check_optional_argument(tree decl,bool for_present_check)78 omp_check_optional_argument (tree decl, bool for_present_check)
79 {
80 return lang_hooks.decls.omp_check_optional_argument (decl, for_present_check);
81 }
82
83 /* True if OpenMP should privatize what this DECL points to rather
84 than the DECL itself. */
85
86 bool
omp_privatize_by_reference(tree decl)87 omp_privatize_by_reference (tree decl)
88 {
89 return lang_hooks.decls.omp_privatize_by_reference (decl);
90 }
91
92 /* Adjust *COND_CODE and *N2 so that the former is either LT_EXPR or GT_EXPR,
93 given that V is the loop index variable and STEP is loop step. */
94
95 void
omp_adjust_for_condition(location_t loc,enum tree_code * cond_code,tree * n2,tree v,tree step)96 omp_adjust_for_condition (location_t loc, enum tree_code *cond_code, tree *n2,
97 tree v, tree step)
98 {
99 switch (*cond_code)
100 {
101 case LT_EXPR:
102 case GT_EXPR:
103 break;
104
105 case NE_EXPR:
106 gcc_assert (TREE_CODE (step) == INTEGER_CST);
107 if (TREE_CODE (TREE_TYPE (v)) == INTEGER_TYPE)
108 {
109 if (integer_onep (step))
110 *cond_code = LT_EXPR;
111 else
112 {
113 gcc_assert (integer_minus_onep (step));
114 *cond_code = GT_EXPR;
115 }
116 }
117 else
118 {
119 tree unit = TYPE_SIZE_UNIT (TREE_TYPE (TREE_TYPE (v)));
120 gcc_assert (TREE_CODE (unit) == INTEGER_CST);
121 if (tree_int_cst_equal (unit, step))
122 *cond_code = LT_EXPR;
123 else
124 {
125 gcc_assert (wi::neg (wi::to_widest (unit))
126 == wi::to_widest (step));
127 *cond_code = GT_EXPR;
128 }
129 }
130
131 break;
132
133 case LE_EXPR:
134 if (POINTER_TYPE_P (TREE_TYPE (*n2)))
135 *n2 = fold_build_pointer_plus_hwi_loc (loc, *n2, 1);
136 else
137 *n2 = fold_build2_loc (loc, PLUS_EXPR, TREE_TYPE (*n2), *n2,
138 build_int_cst (TREE_TYPE (*n2), 1));
139 *cond_code = LT_EXPR;
140 break;
141 case GE_EXPR:
142 if (POINTER_TYPE_P (TREE_TYPE (*n2)))
143 *n2 = fold_build_pointer_plus_hwi_loc (loc, *n2, -1);
144 else
145 *n2 = fold_build2_loc (loc, MINUS_EXPR, TREE_TYPE (*n2), *n2,
146 build_int_cst (TREE_TYPE (*n2), 1));
147 *cond_code = GT_EXPR;
148 break;
149 default:
150 gcc_unreachable ();
151 }
152 }
153
154 /* Return the looping step from INCR, extracted from the step of a gimple omp
155 for statement. */
156
157 tree
omp_get_for_step_from_incr(location_t loc,tree incr)158 omp_get_for_step_from_incr (location_t loc, tree incr)
159 {
160 tree step;
161 switch (TREE_CODE (incr))
162 {
163 case PLUS_EXPR:
164 step = TREE_OPERAND (incr, 1);
165 break;
166 case POINTER_PLUS_EXPR:
167 step = fold_convert (ssizetype, TREE_OPERAND (incr, 1));
168 break;
169 case MINUS_EXPR:
170 step = TREE_OPERAND (incr, 1);
171 step = fold_build1_loc (loc, NEGATE_EXPR, TREE_TYPE (step), step);
172 break;
173 default:
174 gcc_unreachable ();
175 }
176 return step;
177 }
178
179 /* Extract the header elements of parallel loop FOR_STMT and store
180 them into *FD. */
181
182 void
omp_extract_for_data(gomp_for * for_stmt,struct omp_for_data * fd,struct omp_for_data_loop * loops)183 omp_extract_for_data (gomp_for *for_stmt, struct omp_for_data *fd,
184 struct omp_for_data_loop *loops)
185 {
186 tree t, var, *collapse_iter, *collapse_count;
187 tree count = NULL_TREE, iter_type = long_integer_type_node;
188 struct omp_for_data_loop *loop;
189 int i;
190 struct omp_for_data_loop dummy_loop;
191 location_t loc = gimple_location (for_stmt);
192 bool simd = gimple_omp_for_kind (for_stmt) == GF_OMP_FOR_KIND_SIMD;
193 bool distribute = gimple_omp_for_kind (for_stmt)
194 == GF_OMP_FOR_KIND_DISTRIBUTE;
195 bool taskloop = gimple_omp_for_kind (for_stmt)
196 == GF_OMP_FOR_KIND_TASKLOOP;
197 bool order_reproducible = false;
198 tree iterv, countv;
199
200 fd->for_stmt = for_stmt;
201 fd->pre = NULL;
202 fd->have_nowait = distribute || simd;
203 fd->have_ordered = false;
204 fd->have_reductemp = false;
205 fd->have_pointer_condtemp = false;
206 fd->have_scantemp = false;
207 fd->have_nonctrl_scantemp = false;
208 fd->non_rect = false;
209 fd->lastprivate_conditional = 0;
210 fd->tiling = NULL_TREE;
211 fd->collapse = 1;
212 fd->ordered = 0;
213 fd->first_nonrect = -1;
214 fd->last_nonrect = -1;
215 fd->sched_kind = OMP_CLAUSE_SCHEDULE_STATIC;
216 fd->sched_modifiers = 0;
217 fd->chunk_size = NULL_TREE;
218 fd->simd_schedule = false;
219 fd->first_inner_iterations = NULL_TREE;
220 fd->factor = NULL_TREE;
221 fd->adjn1 = NULL_TREE;
222 collapse_iter = NULL;
223 collapse_count = NULL;
224
225 for (t = gimple_omp_for_clauses (for_stmt); t ; t = OMP_CLAUSE_CHAIN (t))
226 switch (OMP_CLAUSE_CODE (t))
227 {
228 case OMP_CLAUSE_NOWAIT:
229 fd->have_nowait = true;
230 break;
231 case OMP_CLAUSE_ORDERED:
232 fd->have_ordered = true;
233 if (OMP_CLAUSE_ORDERED_EXPR (t))
234 fd->ordered = tree_to_shwi (OMP_CLAUSE_ORDERED_EXPR (t));
235 break;
236 case OMP_CLAUSE_SCHEDULE:
237 gcc_assert (!distribute && !taskloop);
238 fd->sched_kind
239 = (enum omp_clause_schedule_kind)
240 (OMP_CLAUSE_SCHEDULE_KIND (t) & OMP_CLAUSE_SCHEDULE_MASK);
241 fd->sched_modifiers = (OMP_CLAUSE_SCHEDULE_KIND (t)
242 & ~OMP_CLAUSE_SCHEDULE_MASK);
243 fd->chunk_size = OMP_CLAUSE_SCHEDULE_CHUNK_EXPR (t);
244 fd->simd_schedule = OMP_CLAUSE_SCHEDULE_SIMD (t);
245 break;
246 case OMP_CLAUSE_DIST_SCHEDULE:
247 gcc_assert (distribute);
248 fd->chunk_size = OMP_CLAUSE_DIST_SCHEDULE_CHUNK_EXPR (t);
249 break;
250 case OMP_CLAUSE_COLLAPSE:
251 fd->collapse = tree_to_shwi (OMP_CLAUSE_COLLAPSE_EXPR (t));
252 if (fd->collapse > 1)
253 {
254 collapse_iter = &OMP_CLAUSE_COLLAPSE_ITERVAR (t);
255 collapse_count = &OMP_CLAUSE_COLLAPSE_COUNT (t);
256 }
257 break;
258 case OMP_CLAUSE_TILE:
259 fd->tiling = OMP_CLAUSE_TILE_LIST (t);
260 fd->collapse = list_length (fd->tiling);
261 gcc_assert (fd->collapse);
262 collapse_iter = &OMP_CLAUSE_TILE_ITERVAR (t);
263 collapse_count = &OMP_CLAUSE_TILE_COUNT (t);
264 break;
265 case OMP_CLAUSE__REDUCTEMP_:
266 fd->have_reductemp = true;
267 break;
268 case OMP_CLAUSE_LASTPRIVATE:
269 if (OMP_CLAUSE_LASTPRIVATE_CONDITIONAL (t))
270 fd->lastprivate_conditional++;
271 break;
272 case OMP_CLAUSE__CONDTEMP_:
273 if (POINTER_TYPE_P (TREE_TYPE (OMP_CLAUSE_DECL (t))))
274 fd->have_pointer_condtemp = true;
275 break;
276 case OMP_CLAUSE__SCANTEMP_:
277 fd->have_scantemp = true;
278 if (!OMP_CLAUSE__SCANTEMP__ALLOC (t)
279 && !OMP_CLAUSE__SCANTEMP__CONTROL (t))
280 fd->have_nonctrl_scantemp = true;
281 break;
282 case OMP_CLAUSE_ORDER:
283 /* FIXME: For OpenMP 5.2 this should change to
284 if (OMP_CLAUSE_ORDER_REPRODUCIBLE (t))
285 (with the exception of loop construct but that lowers to
286 no schedule/dist_schedule clauses currently). */
287 if (!OMP_CLAUSE_ORDER_UNCONSTRAINED (t))
288 order_reproducible = true;
289 default:
290 break;
291 }
292
293 /* For order(reproducible:concurrent) schedule ({dynamic,guided,runtime})
294 we have either the option to expensively remember at runtime how we've
295 distributed work from first loop and reuse that in following loops with
296 the same number of iterations and schedule, or just force static schedule.
297 OpenMP API calls etc. aren't allowed in order(concurrent) bodies so
298 users can't observe it easily anyway. */
299 if (order_reproducible)
300 fd->sched_kind = OMP_CLAUSE_SCHEDULE_STATIC;
301 if (fd->collapse > 1 || fd->tiling)
302 fd->loops = loops;
303 else
304 fd->loops = &fd->loop;
305
306 if (fd->ordered && fd->collapse == 1 && loops != NULL)
307 {
308 fd->loops = loops;
309 iterv = NULL_TREE;
310 countv = NULL_TREE;
311 collapse_iter = &iterv;
312 collapse_count = &countv;
313 }
314
315 /* FIXME: for now map schedule(auto) to schedule(static).
316 There should be analysis to determine whether all iterations
317 are approximately the same amount of work (then schedule(static)
318 is best) or if it varies (then schedule(dynamic,N) is better). */
319 if (fd->sched_kind == OMP_CLAUSE_SCHEDULE_AUTO)
320 {
321 fd->sched_kind = OMP_CLAUSE_SCHEDULE_STATIC;
322 gcc_assert (fd->chunk_size == NULL);
323 }
324 gcc_assert ((fd->collapse == 1 && !fd->tiling) || collapse_iter != NULL);
325 if (taskloop)
326 fd->sched_kind = OMP_CLAUSE_SCHEDULE_RUNTIME;
327 if (fd->sched_kind == OMP_CLAUSE_SCHEDULE_RUNTIME)
328 gcc_assert (fd->chunk_size == NULL);
329 else if (fd->chunk_size == NULL)
330 {
331 /* We only need to compute a default chunk size for ordered
332 static loops and dynamic loops. */
333 if (fd->sched_kind != OMP_CLAUSE_SCHEDULE_STATIC
334 || fd->have_ordered)
335 fd->chunk_size = (fd->sched_kind == OMP_CLAUSE_SCHEDULE_STATIC)
336 ? integer_zero_node : integer_one_node;
337 }
338
339 int cnt = fd->ordered ? fd->ordered : fd->collapse;
340 int single_nonrect = -1;
341 tree single_nonrect_count = NULL_TREE;
342 enum tree_code single_nonrect_cond_code = ERROR_MARK;
343 for (i = 1; i < cnt; i++)
344 {
345 tree n1 = gimple_omp_for_initial (for_stmt, i);
346 tree n2 = gimple_omp_for_final (for_stmt, i);
347 if (TREE_CODE (n1) == TREE_VEC)
348 {
349 if (fd->non_rect)
350 {
351 single_nonrect = -1;
352 break;
353 }
354 for (int j = i - 1; j >= 0; j--)
355 if (TREE_VEC_ELT (n1, 0) == gimple_omp_for_index (for_stmt, j))
356 {
357 single_nonrect = j;
358 break;
359 }
360 fd->non_rect = true;
361 }
362 else if (TREE_CODE (n2) == TREE_VEC)
363 {
364 if (fd->non_rect)
365 {
366 single_nonrect = -1;
367 break;
368 }
369 for (int j = i - 1; j >= 0; j--)
370 if (TREE_VEC_ELT (n2, 0) == gimple_omp_for_index (for_stmt, j))
371 {
372 single_nonrect = j;
373 break;
374 }
375 fd->non_rect = true;
376 }
377 }
378 for (i = 0; i < cnt; i++)
379 {
380 if (i == 0
381 && fd->collapse == 1
382 && !fd->tiling
383 && (fd->ordered == 0 || loops == NULL))
384 loop = &fd->loop;
385 else if (loops != NULL)
386 loop = loops + i;
387 else
388 loop = &dummy_loop;
389
390 loop->v = gimple_omp_for_index (for_stmt, i);
391 gcc_assert (SSA_VAR_P (loop->v));
392 gcc_assert (TREE_CODE (TREE_TYPE (loop->v)) == INTEGER_TYPE
393 || TREE_CODE (TREE_TYPE (loop->v)) == POINTER_TYPE);
394 var = TREE_CODE (loop->v) == SSA_NAME ? SSA_NAME_VAR (loop->v) : loop->v;
395 loop->n1 = gimple_omp_for_initial (for_stmt, i);
396 loop->m1 = NULL_TREE;
397 loop->m2 = NULL_TREE;
398 loop->outer = 0;
399 loop->non_rect_referenced = false;
400 if (TREE_CODE (loop->n1) == TREE_VEC)
401 {
402 for (int j = i - 1; j >= 0; j--)
403 if (TREE_VEC_ELT (loop->n1, 0) == gimple_omp_for_index (for_stmt, j))
404 {
405 loop->outer = i - j;
406 if (loops != NULL)
407 loops[j].non_rect_referenced = true;
408 if (fd->first_nonrect == -1 || fd->first_nonrect > j)
409 fd->first_nonrect = j;
410 break;
411 }
412 gcc_assert (loop->outer);
413 loop->m1 = TREE_VEC_ELT (loop->n1, 1);
414 loop->n1 = TREE_VEC_ELT (loop->n1, 2);
415 fd->non_rect = true;
416 fd->last_nonrect = i;
417 }
418
419 loop->cond_code = gimple_omp_for_cond (for_stmt, i);
420 loop->n2 = gimple_omp_for_final (for_stmt, i);
421 gcc_assert (loop->cond_code != NE_EXPR
422 || (gimple_omp_for_kind (for_stmt)
423 != GF_OMP_FOR_KIND_OACC_LOOP));
424 if (TREE_CODE (loop->n2) == TREE_VEC)
425 {
426 if (loop->outer)
427 gcc_assert (TREE_VEC_ELT (loop->n2, 0)
428 == gimple_omp_for_index (for_stmt, i - loop->outer));
429 else
430 for (int j = i - 1; j >= 0; j--)
431 if (TREE_VEC_ELT (loop->n2, 0) == gimple_omp_for_index (for_stmt, j))
432 {
433 loop->outer = i - j;
434 if (loops != NULL)
435 loops[j].non_rect_referenced = true;
436 if (fd->first_nonrect == -1 || fd->first_nonrect > j)
437 fd->first_nonrect = j;
438 break;
439 }
440 gcc_assert (loop->outer);
441 loop->m2 = TREE_VEC_ELT (loop->n2, 1);
442 loop->n2 = TREE_VEC_ELT (loop->n2, 2);
443 fd->non_rect = true;
444 fd->last_nonrect = i;
445 }
446
447 t = gimple_omp_for_incr (for_stmt, i);
448 gcc_assert (TREE_OPERAND (t, 0) == var);
449 loop->step = omp_get_for_step_from_incr (loc, t);
450
451 omp_adjust_for_condition (loc, &loop->cond_code, &loop->n2, loop->v,
452 loop->step);
453
454 if (simd
455 || (fd->sched_kind == OMP_CLAUSE_SCHEDULE_STATIC
456 && !fd->have_ordered))
457 {
458 if (fd->collapse == 1 && !fd->tiling)
459 iter_type = TREE_TYPE (loop->v);
460 else if (i == 0
461 || TYPE_PRECISION (iter_type)
462 < TYPE_PRECISION (TREE_TYPE (loop->v)))
463 iter_type
464 = build_nonstandard_integer_type
465 (TYPE_PRECISION (TREE_TYPE (loop->v)), 1);
466 }
467 else if (iter_type != long_long_unsigned_type_node)
468 {
469 if (POINTER_TYPE_P (TREE_TYPE (loop->v)))
470 iter_type = long_long_unsigned_type_node;
471 else if (TYPE_UNSIGNED (TREE_TYPE (loop->v))
472 && TYPE_PRECISION (TREE_TYPE (loop->v))
473 >= TYPE_PRECISION (iter_type))
474 {
475 tree n;
476
477 if (loop->cond_code == LT_EXPR)
478 n = fold_build2_loc (loc, PLUS_EXPR, TREE_TYPE (loop->v),
479 loop->n2, loop->step);
480 else
481 n = loop->n1;
482 if (loop->m1
483 || loop->m2
484 || TREE_CODE (n) != INTEGER_CST
485 || tree_int_cst_lt (TYPE_MAX_VALUE (iter_type), n))
486 iter_type = long_long_unsigned_type_node;
487 }
488 else if (TYPE_PRECISION (TREE_TYPE (loop->v))
489 > TYPE_PRECISION (iter_type))
490 {
491 tree n1, n2;
492
493 if (loop->cond_code == LT_EXPR)
494 {
495 n1 = loop->n1;
496 n2 = fold_build2_loc (loc, PLUS_EXPR, TREE_TYPE (loop->v),
497 loop->n2, loop->step);
498 }
499 else
500 {
501 n1 = fold_build2_loc (loc, MINUS_EXPR, TREE_TYPE (loop->v),
502 loop->n2, loop->step);
503 n2 = loop->n1;
504 }
505 if (loop->m1
506 || loop->m2
507 || TREE_CODE (n1) != INTEGER_CST
508 || TREE_CODE (n2) != INTEGER_CST
509 || !tree_int_cst_lt (TYPE_MIN_VALUE (iter_type), n1)
510 || !tree_int_cst_lt (n2, TYPE_MAX_VALUE (iter_type)))
511 iter_type = long_long_unsigned_type_node;
512 }
513 }
514
515 if (i >= fd->collapse)
516 continue;
517
518 if (collapse_count && *collapse_count == NULL)
519 {
520 if (count && integer_zerop (count))
521 continue;
522 tree n1first = NULL_TREE, n2first = NULL_TREE;
523 tree n1last = NULL_TREE, n2last = NULL_TREE;
524 tree ostep = NULL_TREE;
525 if (loop->m1 || loop->m2)
526 {
527 if (count == NULL_TREE)
528 continue;
529 if (single_nonrect == -1
530 || (loop->m1 && TREE_CODE (loop->m1) != INTEGER_CST)
531 || (loop->m2 && TREE_CODE (loop->m2) != INTEGER_CST)
532 || TREE_CODE (loop->n1) != INTEGER_CST
533 || TREE_CODE (loop->n2) != INTEGER_CST
534 || TREE_CODE (loop->step) != INTEGER_CST)
535 {
536 count = NULL_TREE;
537 continue;
538 }
539 tree var = gimple_omp_for_initial (for_stmt, single_nonrect);
540 tree itype = TREE_TYPE (var);
541 tree first = gimple_omp_for_initial (for_stmt, single_nonrect);
542 t = gimple_omp_for_incr (for_stmt, single_nonrect);
543 ostep = omp_get_for_step_from_incr (loc, t);
544 t = fold_binary (MINUS_EXPR, long_long_unsigned_type_node,
545 single_nonrect_count,
546 build_one_cst (long_long_unsigned_type_node));
547 t = fold_convert (itype, t);
548 first = fold_convert (itype, first);
549 ostep = fold_convert (itype, ostep);
550 tree last = fold_binary (PLUS_EXPR, itype, first,
551 fold_binary (MULT_EXPR, itype, t,
552 ostep));
553 if (TREE_CODE (first) != INTEGER_CST
554 || TREE_CODE (last) != INTEGER_CST)
555 {
556 count = NULL_TREE;
557 continue;
558 }
559 if (loop->m1)
560 {
561 tree m1 = fold_convert (itype, loop->m1);
562 tree n1 = fold_convert (itype, loop->n1);
563 n1first = fold_binary (PLUS_EXPR, itype,
564 fold_binary (MULT_EXPR, itype,
565 first, m1), n1);
566 n1last = fold_binary (PLUS_EXPR, itype,
567 fold_binary (MULT_EXPR, itype,
568 last, m1), n1);
569 }
570 else
571 n1first = n1last = loop->n1;
572 if (loop->m2)
573 {
574 tree n2 = fold_convert (itype, loop->n2);
575 tree m2 = fold_convert (itype, loop->m2);
576 n2first = fold_binary (PLUS_EXPR, itype,
577 fold_binary (MULT_EXPR, itype,
578 first, m2), n2);
579 n2last = fold_binary (PLUS_EXPR, itype,
580 fold_binary (MULT_EXPR, itype,
581 last, m2), n2);
582 }
583 else
584 n2first = n2last = loop->n2;
585 n1first = fold_convert (TREE_TYPE (loop->v), n1first);
586 n2first = fold_convert (TREE_TYPE (loop->v), n2first);
587 n1last = fold_convert (TREE_TYPE (loop->v), n1last);
588 n2last = fold_convert (TREE_TYPE (loop->v), n2last);
589 t = fold_binary (loop->cond_code, boolean_type_node,
590 n1first, n2first);
591 tree t2 = fold_binary (loop->cond_code, boolean_type_node,
592 n1last, n2last);
593 if (t && t2 && integer_nonzerop (t) && integer_nonzerop (t2))
594 /* All outer loop iterators have at least one inner loop
595 iteration. Try to compute the count at compile time. */
596 t = NULL_TREE;
597 else if (t && t2 && integer_zerop (t) && integer_zerop (t2))
598 /* No iterations of the inner loop. count will be set to
599 zero cst below. */;
600 else if (TYPE_UNSIGNED (itype)
601 || t == NULL_TREE
602 || t2 == NULL_TREE
603 || TREE_CODE (t) != INTEGER_CST
604 || TREE_CODE (t2) != INTEGER_CST)
605 {
606 /* Punt (for now). */
607 count = NULL_TREE;
608 continue;
609 }
610 else
611 {
612 /* Some iterations of the outer loop have zero iterations
613 of the inner loop, while others have at least one.
614 In this case, we need to adjust one of those outer
615 loop bounds. If ADJ_FIRST, we need to adjust outer n1
616 (first), otherwise outer n2 (last). */
617 bool adj_first = integer_zerop (t);
618 tree n1 = fold_convert (itype, loop->n1);
619 tree n2 = fold_convert (itype, loop->n2);
620 tree m1 = loop->m1 ? fold_convert (itype, loop->m1)
621 : build_zero_cst (itype);
622 tree m2 = loop->m2 ? fold_convert (itype, loop->m2)
623 : build_zero_cst (itype);
624 t = fold_binary (MINUS_EXPR, itype, n1, n2);
625 t2 = fold_binary (MINUS_EXPR, itype, m2, m1);
626 t = fold_binary (TRUNC_DIV_EXPR, itype, t, t2);
627 t2 = fold_binary (MINUS_EXPR, itype, t, first);
628 t2 = fold_binary (TRUNC_MOD_EXPR, itype, t2, ostep);
629 t = fold_binary (MINUS_EXPR, itype, t, t2);
630 tree n1cur
631 = fold_binary (PLUS_EXPR, itype, n1,
632 fold_binary (MULT_EXPR, itype, m1, t));
633 tree n2cur
634 = fold_binary (PLUS_EXPR, itype, n2,
635 fold_binary (MULT_EXPR, itype, m2, t));
636 t2 = fold_binary (loop->cond_code, boolean_type_node,
637 n1cur, n2cur);
638 tree t3 = fold_binary (MULT_EXPR, itype, m1, ostep);
639 tree t4 = fold_binary (MULT_EXPR, itype, m2, ostep);
640 tree diff;
641 if (adj_first)
642 {
643 tree new_first;
644 if (integer_nonzerop (t2))
645 {
646 new_first = t;
647 n1first = n1cur;
648 n2first = n2cur;
649 if (flag_checking)
650 {
651 t3 = fold_binary (MINUS_EXPR, itype, n1cur, t3);
652 t4 = fold_binary (MINUS_EXPR, itype, n2cur, t4);
653 t3 = fold_binary (loop->cond_code,
654 boolean_type_node, t3, t4);
655 gcc_assert (integer_zerop (t3));
656 }
657 }
658 else
659 {
660 t3 = fold_binary (PLUS_EXPR, itype, n1cur, t3);
661 t4 = fold_binary (PLUS_EXPR, itype, n2cur, t4);
662 new_first = fold_binary (PLUS_EXPR, itype, t, ostep);
663 n1first = t3;
664 n2first = t4;
665 if (flag_checking)
666 {
667 t3 = fold_binary (loop->cond_code,
668 boolean_type_node, t3, t4);
669 gcc_assert (integer_nonzerop (t3));
670 }
671 }
672 diff = fold_binary (MINUS_EXPR, itype, new_first, first);
673 first = new_first;
674 fd->adjn1 = first;
675 }
676 else
677 {
678 tree new_last;
679 if (integer_zerop (t2))
680 {
681 t3 = fold_binary (MINUS_EXPR, itype, n1cur, t3);
682 t4 = fold_binary (MINUS_EXPR, itype, n2cur, t4);
683 new_last = fold_binary (MINUS_EXPR, itype, t, ostep);
684 n1last = t3;
685 n2last = t4;
686 if (flag_checking)
687 {
688 t3 = fold_binary (loop->cond_code,
689 boolean_type_node, t3, t4);
690 gcc_assert (integer_nonzerop (t3));
691 }
692 }
693 else
694 {
695 new_last = t;
696 n1last = n1cur;
697 n2last = n2cur;
698 if (flag_checking)
699 {
700 t3 = fold_binary (PLUS_EXPR, itype, n1cur, t3);
701 t4 = fold_binary (PLUS_EXPR, itype, n2cur, t4);
702 t3 = fold_binary (loop->cond_code,
703 boolean_type_node, t3, t4);
704 gcc_assert (integer_zerop (t3));
705 }
706 }
707 diff = fold_binary (MINUS_EXPR, itype, last, new_last);
708 }
709 if (TYPE_UNSIGNED (itype)
710 && single_nonrect_cond_code == GT_EXPR)
711 diff = fold_binary (TRUNC_DIV_EXPR, itype,
712 fold_unary (NEGATE_EXPR, itype, diff),
713 fold_unary (NEGATE_EXPR, itype,
714 ostep));
715 else
716 diff = fold_binary (TRUNC_DIV_EXPR, itype, diff, ostep);
717 diff = fold_convert (long_long_unsigned_type_node, diff);
718 single_nonrect_count
719 = fold_binary (MINUS_EXPR, long_long_unsigned_type_node,
720 single_nonrect_count, diff);
721 t = NULL_TREE;
722 }
723 }
724 else
725 t = fold_binary (loop->cond_code, boolean_type_node,
726 fold_convert (TREE_TYPE (loop->v), loop->n1),
727 fold_convert (TREE_TYPE (loop->v), loop->n2));
728 if (t && integer_zerop (t))
729 count = build_zero_cst (long_long_unsigned_type_node);
730 else if ((i == 0 || count != NULL_TREE)
731 && TREE_CODE (TREE_TYPE (loop->v)) == INTEGER_TYPE
732 && TREE_CONSTANT (loop->n1)
733 && TREE_CONSTANT (loop->n2)
734 && TREE_CODE (loop->step) == INTEGER_CST)
735 {
736 tree itype = TREE_TYPE (loop->v);
737
738 if (POINTER_TYPE_P (itype))
739 itype = signed_type_for (itype);
740 t = build_int_cst (itype, (loop->cond_code == LT_EXPR ? -1 : 1));
741 t = fold_build2 (PLUS_EXPR, itype,
742 fold_convert (itype, loop->step), t);
743 tree n1 = loop->n1;
744 tree n2 = loop->n2;
745 if (loop->m1 || loop->m2)
746 {
747 gcc_assert (single_nonrect != -1);
748 n1 = n1first;
749 n2 = n2first;
750 }
751 t = fold_build2 (PLUS_EXPR, itype, t, fold_convert (itype, n2));
752 t = fold_build2 (MINUS_EXPR, itype, t, fold_convert (itype, n1));
753 tree step = fold_convert_loc (loc, itype, loop->step);
754 if (TYPE_UNSIGNED (itype) && loop->cond_code == GT_EXPR)
755 t = fold_build2 (TRUNC_DIV_EXPR, itype,
756 fold_build1 (NEGATE_EXPR, itype, t),
757 fold_build1 (NEGATE_EXPR, itype, step));
758 else
759 t = fold_build2 (TRUNC_DIV_EXPR, itype, t, step);
760 tree llutype = long_long_unsigned_type_node;
761 t = fold_convert (llutype, t);
762 if (loop->m1 || loop->m2)
763 {
764 /* t is number of iterations of inner loop at either first
765 or last value of the outer iterator (the one with fewer
766 iterations).
767 Compute t2 = ((m2 - m1) * ostep) / step
768 and niters = outer_count * t
769 + t2 * ((outer_count - 1) * outer_count / 2)
770 */
771 tree m1 = loop->m1 ? loop->m1 : integer_zero_node;
772 tree m2 = loop->m2 ? loop->m2 : integer_zero_node;
773 m1 = fold_convert (itype, m1);
774 m2 = fold_convert (itype, m2);
775 tree t2 = fold_build2 (MINUS_EXPR, itype, m2, m1);
776 t2 = fold_build2 (MULT_EXPR, itype, t2, ostep);
777 if (TYPE_UNSIGNED (itype) && loop->cond_code == GT_EXPR)
778 t2 = fold_build2 (TRUNC_DIV_EXPR, itype,
779 fold_build1 (NEGATE_EXPR, itype, t2),
780 fold_build1 (NEGATE_EXPR, itype, step));
781 else
782 t2 = fold_build2 (TRUNC_DIV_EXPR, itype, t2, step);
783 t2 = fold_convert (llutype, t2);
784 fd->first_inner_iterations = t;
785 fd->factor = t2;
786 t = fold_build2 (MULT_EXPR, llutype, t,
787 single_nonrect_count);
788 tree t3 = fold_build2 (MINUS_EXPR, llutype,
789 single_nonrect_count,
790 build_one_cst (llutype));
791 t3 = fold_build2 (MULT_EXPR, llutype, t3,
792 single_nonrect_count);
793 t3 = fold_build2 (TRUNC_DIV_EXPR, llutype, t3,
794 build_int_cst (llutype, 2));
795 t2 = fold_build2 (MULT_EXPR, llutype, t2, t3);
796 t = fold_build2 (PLUS_EXPR, llutype, t, t2);
797 }
798 if (i == single_nonrect)
799 {
800 if (integer_zerop (t) || TREE_CODE (t) != INTEGER_CST)
801 count = t;
802 else
803 {
804 single_nonrect_count = t;
805 single_nonrect_cond_code = loop->cond_code;
806 if (count == NULL_TREE)
807 count = build_one_cst (llutype);
808 }
809 }
810 else if (count != NULL_TREE)
811 count = fold_build2 (MULT_EXPR, llutype, count, t);
812 else
813 count = t;
814 if (TREE_CODE (count) != INTEGER_CST)
815 count = NULL_TREE;
816 }
817 else if (count && !integer_zerop (count))
818 count = NULL_TREE;
819 }
820 }
821
822 if (count
823 && !simd
824 && (fd->sched_kind != OMP_CLAUSE_SCHEDULE_STATIC
825 || fd->have_ordered))
826 {
827 if (!tree_int_cst_lt (count, TYPE_MAX_VALUE (long_integer_type_node)))
828 iter_type = long_long_unsigned_type_node;
829 else
830 iter_type = long_integer_type_node;
831 }
832 else if (collapse_iter && *collapse_iter != NULL)
833 iter_type = TREE_TYPE (*collapse_iter);
834 fd->iter_type = iter_type;
835 if (collapse_iter && *collapse_iter == NULL)
836 *collapse_iter = create_tmp_var (iter_type, ".iter");
837 if (collapse_count && *collapse_count == NULL)
838 {
839 if (count)
840 {
841 *collapse_count = fold_convert_loc (loc, iter_type, count);
842 if (fd->first_inner_iterations && fd->factor)
843 {
844 t = make_tree_vec (4);
845 TREE_VEC_ELT (t, 0) = *collapse_count;
846 TREE_VEC_ELT (t, 1) = fd->first_inner_iterations;
847 TREE_VEC_ELT (t, 2) = fd->factor;
848 TREE_VEC_ELT (t, 3) = fd->adjn1;
849 *collapse_count = t;
850 }
851 }
852 else
853 *collapse_count = create_tmp_var (iter_type, ".count");
854 }
855
856 if (fd->collapse > 1 || fd->tiling || (fd->ordered && loops))
857 {
858 fd->loop.v = *collapse_iter;
859 fd->loop.n1 = build_int_cst (TREE_TYPE (fd->loop.v), 0);
860 fd->loop.n2 = *collapse_count;
861 if (TREE_CODE (fd->loop.n2) == TREE_VEC)
862 {
863 gcc_assert (fd->non_rect);
864 fd->first_inner_iterations = TREE_VEC_ELT (fd->loop.n2, 1);
865 fd->factor = TREE_VEC_ELT (fd->loop.n2, 2);
866 fd->adjn1 = TREE_VEC_ELT (fd->loop.n2, 3);
867 fd->loop.n2 = TREE_VEC_ELT (fd->loop.n2, 0);
868 }
869 fd->loop.step = build_int_cst (TREE_TYPE (fd->loop.v), 1);
870 fd->loop.m1 = NULL_TREE;
871 fd->loop.m2 = NULL_TREE;
872 fd->loop.outer = 0;
873 fd->loop.cond_code = LT_EXPR;
874 }
875 else if (loops)
876 loops[0] = fd->loop;
877 }
878
879 /* Build a call to GOMP_barrier. */
880
881 gimple *
omp_build_barrier(tree lhs)882 omp_build_barrier (tree lhs)
883 {
884 tree fndecl = builtin_decl_explicit (lhs ? BUILT_IN_GOMP_BARRIER_CANCEL
885 : BUILT_IN_GOMP_BARRIER);
886 gcall *g = gimple_build_call (fndecl, 0);
887 if (lhs)
888 gimple_call_set_lhs (g, lhs);
889 return g;
890 }
891
892 /* Find OMP_FOR resp. OMP_SIMD with non-NULL OMP_FOR_INIT. Also, fill in pdata
893 array, pdata[0] non-NULL if there is anything non-trivial in between,
894 pdata[1] is address of OMP_PARALLEL in between if any, pdata[2] is address
895 of OMP_FOR in between if any and pdata[3] is address of the inner
896 OMP_FOR/OMP_SIMD. */
897
898 tree
find_combined_omp_for(tree * tp,int * walk_subtrees,void * data)899 find_combined_omp_for (tree *tp, int *walk_subtrees, void *data)
900 {
901 tree **pdata = (tree **) data;
902 *walk_subtrees = 0;
903 switch (TREE_CODE (*tp))
904 {
905 case OMP_FOR:
906 if (OMP_FOR_INIT (*tp) != NULL_TREE)
907 {
908 pdata[3] = tp;
909 return *tp;
910 }
911 pdata[2] = tp;
912 *walk_subtrees = 1;
913 break;
914 case OMP_SIMD:
915 if (OMP_FOR_INIT (*tp) != NULL_TREE)
916 {
917 pdata[3] = tp;
918 return *tp;
919 }
920 break;
921 case BIND_EXPR:
922 if (BIND_EXPR_VARS (*tp)
923 || (BIND_EXPR_BLOCK (*tp)
924 && BLOCK_VARS (BIND_EXPR_BLOCK (*tp))))
925 pdata[0] = tp;
926 *walk_subtrees = 1;
927 break;
928 case STATEMENT_LIST:
929 if (!tsi_one_before_end_p (tsi_start (*tp)))
930 pdata[0] = tp;
931 *walk_subtrees = 1;
932 break;
933 case TRY_FINALLY_EXPR:
934 pdata[0] = tp;
935 *walk_subtrees = 1;
936 break;
937 case OMP_PARALLEL:
938 pdata[1] = tp;
939 *walk_subtrees = 1;
940 break;
941 default:
942 break;
943 }
944 return NULL_TREE;
945 }
946
947 /* Return maximum possible vectorization factor for the target. */
948
949 poly_uint64
omp_max_vf(void)950 omp_max_vf (void)
951 {
952 if (!optimize
953 || optimize_debug
954 || !flag_tree_loop_optimize
955 || (!flag_tree_loop_vectorize
956 && OPTION_SET_P (flag_tree_loop_vectorize)))
957 return 1;
958
959 auto_vector_modes modes;
960 targetm.vectorize.autovectorize_vector_modes (&modes, true);
961 if (!modes.is_empty ())
962 {
963 poly_uint64 vf = 0;
964 for (unsigned int i = 0; i < modes.length (); ++i)
965 /* The returned modes use the smallest element size (and thus
966 the largest nunits) for the vectorization approach that they
967 represent. */
968 vf = ordered_max (vf, GET_MODE_NUNITS (modes[i]));
969 return vf;
970 }
971
972 machine_mode vqimode = targetm.vectorize.preferred_simd_mode (QImode);
973 if (GET_MODE_CLASS (vqimode) == MODE_VECTOR_INT)
974 return GET_MODE_NUNITS (vqimode);
975
976 return 1;
977 }
978
979 /* Return maximum SIMT width if offloading may target SIMT hardware. */
980
981 int
omp_max_simt_vf(void)982 omp_max_simt_vf (void)
983 {
984 if (!optimize)
985 return 0;
986 if (ENABLE_OFFLOADING)
987 for (const char *c = getenv ("OFFLOAD_TARGET_NAMES"); c;)
988 {
989 if (startswith (c, "nvptx"))
990 return 32;
991 else if ((c = strchr (c, ':')))
992 c++;
993 }
994 return 0;
995 }
996
997 /* Store the construct selectors as tree codes from last to first,
998 return their number. */
999
1000 int
omp_constructor_traits_to_codes(tree ctx,enum tree_code * constructs)1001 omp_constructor_traits_to_codes (tree ctx, enum tree_code *constructs)
1002 {
1003 int nconstructs = list_length (ctx);
1004 int i = nconstructs - 1;
1005 for (tree t2 = ctx; t2; t2 = TREE_CHAIN (t2), i--)
1006 {
1007 const char *sel = IDENTIFIER_POINTER (TREE_PURPOSE (t2));
1008 if (!strcmp (sel, "target"))
1009 constructs[i] = OMP_TARGET;
1010 else if (!strcmp (sel, "teams"))
1011 constructs[i] = OMP_TEAMS;
1012 else if (!strcmp (sel, "parallel"))
1013 constructs[i] = OMP_PARALLEL;
1014 else if (!strcmp (sel, "for") || !strcmp (sel, "do"))
1015 constructs[i] = OMP_FOR;
1016 else if (!strcmp (sel, "simd"))
1017 constructs[i] = OMP_SIMD;
1018 else
1019 gcc_unreachable ();
1020 }
1021 gcc_assert (i == -1);
1022 return nconstructs;
1023 }
1024
1025 /* Return true if PROP is possibly present in one of the offloading target's
1026 OpenMP contexts. The format of PROPS string is always offloading target's
1027 name terminated by '\0', followed by properties for that offloading
1028 target separated by '\0' and terminated by another '\0'. The strings
1029 are created from omp-device-properties installed files of all configured
1030 offloading targets. */
1031
1032 static bool
omp_offload_device_kind_arch_isa(const char * props,const char * prop)1033 omp_offload_device_kind_arch_isa (const char *props, const char *prop)
1034 {
1035 const char *names = getenv ("OFFLOAD_TARGET_NAMES");
1036 if (names == NULL || *names == '\0')
1037 return false;
1038 while (*props != '\0')
1039 {
1040 size_t name_len = strlen (props);
1041 bool matches = false;
1042 for (const char *c = names; c; )
1043 {
1044 if (strncmp (props, c, name_len) == 0
1045 && (c[name_len] == '\0'
1046 || c[name_len] == ':'
1047 || c[name_len] == '='))
1048 {
1049 matches = true;
1050 break;
1051 }
1052 else if ((c = strchr (c, ':')))
1053 c++;
1054 }
1055 props = props + name_len + 1;
1056 while (*props != '\0')
1057 {
1058 if (matches && strcmp (props, prop) == 0)
1059 return true;
1060 props = strchr (props, '\0') + 1;
1061 }
1062 props++;
1063 }
1064 return false;
1065 }
1066
1067 /* Return true if the current code location is or might be offloaded.
1068 Return true in declare target functions, or when nested in a target
1069 region or when unsure, return false otherwise. */
1070
1071 static bool
omp_maybe_offloaded(void)1072 omp_maybe_offloaded (void)
1073 {
1074 if (!ENABLE_OFFLOADING)
1075 return false;
1076 const char *names = getenv ("OFFLOAD_TARGET_NAMES");
1077 if (names == NULL || *names == '\0')
1078 return false;
1079
1080 if (symtab->state == PARSING)
1081 /* Maybe. */
1082 return true;
1083 if (cfun && cfun->after_inlining)
1084 return false;
1085 if (current_function_decl
1086 && lookup_attribute ("omp declare target",
1087 DECL_ATTRIBUTES (current_function_decl)))
1088 return true;
1089 if (cfun && (cfun->curr_properties & PROP_gimple_any) == 0)
1090 {
1091 enum tree_code construct = OMP_TARGET;
1092 if (omp_construct_selector_matches (&construct, 1, NULL))
1093 return true;
1094 }
1095 return false;
1096 }
1097
1098
1099 /* Diagnose errors in an OpenMP context selector, return CTX if
1100 it is correct or error_mark_node otherwise. */
1101
1102 tree
omp_check_context_selector(location_t loc,tree ctx)1103 omp_check_context_selector (location_t loc, tree ctx)
1104 {
1105 /* Each trait-set-selector-name can only be specified once.
1106 There are just 4 set names. */
1107 for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1))
1108 for (tree t2 = TREE_CHAIN (t1); t2; t2 = TREE_CHAIN (t2))
1109 if (TREE_PURPOSE (t1) == TREE_PURPOSE (t2))
1110 {
1111 error_at (loc, "selector set %qs specified more than once",
1112 IDENTIFIER_POINTER (TREE_PURPOSE (t1)));
1113 return error_mark_node;
1114 }
1115 for (tree t = ctx; t; t = TREE_CHAIN (t))
1116 {
1117 /* Each trait-selector-name can only be specified once. */
1118 if (list_length (TREE_VALUE (t)) < 5)
1119 {
1120 for (tree t1 = TREE_VALUE (t); t1; t1 = TREE_CHAIN (t1))
1121 for (tree t2 = TREE_CHAIN (t1); t2; t2 = TREE_CHAIN (t2))
1122 if (TREE_PURPOSE (t1) == TREE_PURPOSE (t2))
1123 {
1124 error_at (loc,
1125 "selector %qs specified more than once in set %qs",
1126 IDENTIFIER_POINTER (TREE_PURPOSE (t1)),
1127 IDENTIFIER_POINTER (TREE_PURPOSE (t)));
1128 return error_mark_node;
1129 }
1130 }
1131 else
1132 {
1133 hash_set<tree> pset;
1134 for (tree t1 = TREE_VALUE (t); t1; t1 = TREE_CHAIN (t1))
1135 if (pset.add (TREE_PURPOSE (t1)))
1136 {
1137 error_at (loc,
1138 "selector %qs specified more than once in set %qs",
1139 IDENTIFIER_POINTER (TREE_PURPOSE (t1)),
1140 IDENTIFIER_POINTER (TREE_PURPOSE (t)));
1141 return error_mark_node;
1142 }
1143 }
1144
1145 static const char *const kind[] = {
1146 "host", "nohost", "cpu", "gpu", "fpga", "any", NULL };
1147 static const char *const vendor[] = {
1148 "amd", "arm", "bsc", "cray", "fujitsu", "gnu", "ibm", "intel",
1149 "llvm", "nvidia", "pgi", "ti", "unknown", NULL };
1150 static const char *const extension[] = { NULL };
1151 static const char *const atomic_default_mem_order[] = {
1152 "seq_cst", "relaxed", "acq_rel", NULL };
1153 struct known_properties { const char *set; const char *selector;
1154 const char *const *props; };
1155 known_properties props[] = {
1156 { "device", "kind", kind },
1157 { "implementation", "vendor", vendor },
1158 { "implementation", "extension", extension },
1159 { "implementation", "atomic_default_mem_order",
1160 atomic_default_mem_order } };
1161 for (tree t1 = TREE_VALUE (t); t1; t1 = TREE_CHAIN (t1))
1162 for (unsigned i = 0; i < ARRAY_SIZE (props); i++)
1163 if (!strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t1)),
1164 props[i].selector)
1165 && !strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t)),
1166 props[i].set))
1167 for (tree t2 = TREE_VALUE (t1); t2; t2 = TREE_CHAIN (t2))
1168 for (unsigned j = 0; ; j++)
1169 {
1170 if (props[i].props[j] == NULL)
1171 {
1172 if (TREE_PURPOSE (t2)
1173 && !strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t2)),
1174 " score"))
1175 break;
1176 if (props[i].props == atomic_default_mem_order)
1177 {
1178 error_at (loc,
1179 "incorrect property %qs of %qs selector",
1180 IDENTIFIER_POINTER (TREE_PURPOSE (t2)),
1181 "atomic_default_mem_order");
1182 return error_mark_node;
1183 }
1184 else if (TREE_PURPOSE (t2))
1185 warning_at (loc, 0,
1186 "unknown property %qs of %qs selector",
1187 IDENTIFIER_POINTER (TREE_PURPOSE (t2)),
1188 props[i].selector);
1189 else
1190 warning_at (loc, 0,
1191 "unknown property %qE of %qs selector",
1192 TREE_VALUE (t2), props[i].selector);
1193 break;
1194 }
1195 else if (TREE_PURPOSE (t2) == NULL_TREE)
1196 {
1197 const char *str = TREE_STRING_POINTER (TREE_VALUE (t2));
1198 if (!strcmp (str, props[i].props[j])
1199 && ((size_t) TREE_STRING_LENGTH (TREE_VALUE (t2))
1200 == strlen (str) + (lang_GNU_Fortran () ? 0 : 1)))
1201 break;
1202 }
1203 else if (!strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t2)),
1204 props[i].props[j]))
1205 break;
1206 }
1207 }
1208 return ctx;
1209 }
1210
1211
1212 /* Register VARIANT as variant of some base function marked with
1213 #pragma omp declare variant. CONSTRUCT is corresponding construct
1214 selector set. */
1215
1216 void
omp_mark_declare_variant(location_t loc,tree variant,tree construct)1217 omp_mark_declare_variant (location_t loc, tree variant, tree construct)
1218 {
1219 tree attr = lookup_attribute ("omp declare variant variant",
1220 DECL_ATTRIBUTES (variant));
1221 if (attr == NULL_TREE)
1222 {
1223 attr = tree_cons (get_identifier ("omp declare variant variant"),
1224 unshare_expr (construct),
1225 DECL_ATTRIBUTES (variant));
1226 DECL_ATTRIBUTES (variant) = attr;
1227 return;
1228 }
1229 if ((TREE_VALUE (attr) != NULL_TREE) != (construct != NULL_TREE)
1230 || (construct != NULL_TREE
1231 && omp_context_selector_set_compare ("construct", TREE_VALUE (attr),
1232 construct)))
1233 error_at (loc, "%qD used as a variant with incompatible %<construct%> "
1234 "selector sets", variant);
1235 }
1236
1237
1238 /* Return a name from PROP, a property in selectors accepting
1239 name lists. */
1240
1241 static const char *
omp_context_name_list_prop(tree prop)1242 omp_context_name_list_prop (tree prop)
1243 {
1244 if (TREE_PURPOSE (prop))
1245 return IDENTIFIER_POINTER (TREE_PURPOSE (prop));
1246 else
1247 {
1248 const char *ret = TREE_STRING_POINTER (TREE_VALUE (prop));
1249 if ((size_t) TREE_STRING_LENGTH (TREE_VALUE (prop))
1250 == strlen (ret) + (lang_GNU_Fortran () ? 0 : 1))
1251 return ret;
1252 return NULL;
1253 }
1254 }
1255
1256 /* Return 1 if context selector matches the current OpenMP context, 0
1257 if it does not and -1 if it is unknown and need to be determined later.
1258 Some properties can be checked right away during parsing (this routine),
1259 others need to wait until the whole TU is parsed, others need to wait until
1260 IPA, others until vectorization. */
1261
1262 int
omp_context_selector_matches(tree ctx)1263 omp_context_selector_matches (tree ctx)
1264 {
1265 int ret = 1;
1266 for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1))
1267 {
1268 char set = IDENTIFIER_POINTER (TREE_PURPOSE (t1))[0];
1269 if (set == 'c')
1270 {
1271 /* For now, ignore the construct set. While something can be
1272 determined already during parsing, we don't know until end of TU
1273 whether additional constructs aren't added through declare variant
1274 unless "omp declare variant variant" attribute exists already
1275 (so in most of the cases), and we'd need to maintain set of
1276 surrounding OpenMP constructs, which is better handled during
1277 gimplification. */
1278 if (symtab->state == PARSING)
1279 {
1280 ret = -1;
1281 continue;
1282 }
1283
1284 enum tree_code constructs[5];
1285 int nconstructs
1286 = omp_constructor_traits_to_codes (TREE_VALUE (t1), constructs);
1287
1288 if (cfun && (cfun->curr_properties & PROP_gimple_any) != 0)
1289 {
1290 if (!cfun->after_inlining)
1291 {
1292 ret = -1;
1293 continue;
1294 }
1295 int i;
1296 for (i = 0; i < nconstructs; ++i)
1297 if (constructs[i] == OMP_SIMD)
1298 break;
1299 if (i < nconstructs)
1300 {
1301 ret = -1;
1302 continue;
1303 }
1304 /* If there is no simd, assume it is ok after IPA,
1305 constructs should have been checked before. */
1306 continue;
1307 }
1308
1309 int r = omp_construct_selector_matches (constructs, nconstructs,
1310 NULL);
1311 if (r == 0)
1312 return 0;
1313 if (r == -1)
1314 ret = -1;
1315 continue;
1316 }
1317 for (tree t2 = TREE_VALUE (t1); t2; t2 = TREE_CHAIN (t2))
1318 {
1319 const char *sel = IDENTIFIER_POINTER (TREE_PURPOSE (t2));
1320 switch (*sel)
1321 {
1322 case 'v':
1323 if (set == 'i' && !strcmp (sel, "vendor"))
1324 for (tree t3 = TREE_VALUE (t2); t3; t3 = TREE_CHAIN (t3))
1325 {
1326 const char *prop = omp_context_name_list_prop (t3);
1327 if (prop == NULL)
1328 return 0;
1329 if ((!strcmp (prop, " score") && TREE_PURPOSE (t3))
1330 || !strcmp (prop, "gnu"))
1331 continue;
1332 return 0;
1333 }
1334 break;
1335 case 'e':
1336 if (set == 'i' && !strcmp (sel, "extension"))
1337 /* We don't support any extensions right now. */
1338 return 0;
1339 break;
1340 case 'a':
1341 if (set == 'i' && !strcmp (sel, "atomic_default_mem_order"))
1342 {
1343 if (cfun && (cfun->curr_properties & PROP_gimple_any) != 0)
1344 break;
1345
1346 enum omp_memory_order omo
1347 = ((enum omp_memory_order)
1348 (omp_requires_mask
1349 & OMP_REQUIRES_ATOMIC_DEFAULT_MEM_ORDER));
1350 if (omo == OMP_MEMORY_ORDER_UNSPECIFIED)
1351 {
1352 /* We don't know yet, until end of TU. */
1353 if (symtab->state == PARSING)
1354 {
1355 ret = -1;
1356 break;
1357 }
1358 else
1359 omo = OMP_MEMORY_ORDER_RELAXED;
1360 }
1361 tree t3 = TREE_VALUE (t2);
1362 const char *prop = IDENTIFIER_POINTER (TREE_PURPOSE (t3));
1363 if (!strcmp (prop, " score"))
1364 {
1365 t3 = TREE_CHAIN (t3);
1366 prop = IDENTIFIER_POINTER (TREE_PURPOSE (t3));
1367 }
1368 if (!strcmp (prop, "relaxed")
1369 && omo != OMP_MEMORY_ORDER_RELAXED)
1370 return 0;
1371 else if (!strcmp (prop, "seq_cst")
1372 && omo != OMP_MEMORY_ORDER_SEQ_CST)
1373 return 0;
1374 else if (!strcmp (prop, "acq_rel")
1375 && omo != OMP_MEMORY_ORDER_ACQ_REL)
1376 return 0;
1377 }
1378 if (set == 'd' && !strcmp (sel, "arch"))
1379 for (tree t3 = TREE_VALUE (t2); t3; t3 = TREE_CHAIN (t3))
1380 {
1381 const char *arch = omp_context_name_list_prop (t3);
1382 if (arch == NULL)
1383 return 0;
1384 int r = 0;
1385 if (targetm.omp.device_kind_arch_isa != NULL)
1386 r = targetm.omp.device_kind_arch_isa (omp_device_arch,
1387 arch);
1388 if (r == 0 || (r == -1 && symtab->state != PARSING))
1389 {
1390 /* If we are or might be in a target region or
1391 declare target function, need to take into account
1392 also offloading values. */
1393 if (!omp_maybe_offloaded ())
1394 return 0;
1395 if (ENABLE_OFFLOADING)
1396 {
1397 const char *arches = omp_offload_device_arch;
1398 if (omp_offload_device_kind_arch_isa (arches,
1399 arch))
1400 {
1401 ret = -1;
1402 continue;
1403 }
1404 }
1405 return 0;
1406 }
1407 else if (r == -1)
1408 ret = -1;
1409 /* If arch matches on the host, it still might not match
1410 in the offloading region. */
1411 else if (omp_maybe_offloaded ())
1412 ret = -1;
1413 }
1414 break;
1415 case 'u':
1416 if (set == 'i' && !strcmp (sel, "unified_address"))
1417 {
1418 if (cfun && (cfun->curr_properties & PROP_gimple_any) != 0)
1419 break;
1420
1421 if ((omp_requires_mask & OMP_REQUIRES_UNIFIED_ADDRESS) == 0)
1422 {
1423 if (symtab->state == PARSING)
1424 ret = -1;
1425 else
1426 return 0;
1427 }
1428 break;
1429 }
1430 if (set == 'i' && !strcmp (sel, "unified_shared_memory"))
1431 {
1432 if (cfun && (cfun->curr_properties & PROP_gimple_any) != 0)
1433 break;
1434
1435 if ((omp_requires_mask
1436 & OMP_REQUIRES_UNIFIED_SHARED_MEMORY) == 0)
1437 {
1438 if (symtab->state == PARSING)
1439 ret = -1;
1440 else
1441 return 0;
1442 }
1443 break;
1444 }
1445 break;
1446 case 'd':
1447 if (set == 'i' && !strcmp (sel, "dynamic_allocators"))
1448 {
1449 if (cfun && (cfun->curr_properties & PROP_gimple_any) != 0)
1450 break;
1451
1452 if ((omp_requires_mask
1453 & OMP_REQUIRES_DYNAMIC_ALLOCATORS) == 0)
1454 {
1455 if (symtab->state == PARSING)
1456 ret = -1;
1457 else
1458 return 0;
1459 }
1460 break;
1461 }
1462 break;
1463 case 'r':
1464 if (set == 'i' && !strcmp (sel, "reverse_offload"))
1465 {
1466 if (cfun && (cfun->curr_properties & PROP_gimple_any) != 0)
1467 break;
1468
1469 if ((omp_requires_mask & OMP_REQUIRES_REVERSE_OFFLOAD) == 0)
1470 {
1471 if (symtab->state == PARSING)
1472 ret = -1;
1473 else
1474 return 0;
1475 }
1476 break;
1477 }
1478 break;
1479 case 'k':
1480 if (set == 'd' && !strcmp (sel, "kind"))
1481 for (tree t3 = TREE_VALUE (t2); t3; t3 = TREE_CHAIN (t3))
1482 {
1483 const char *prop = omp_context_name_list_prop (t3);
1484 if (prop == NULL)
1485 return 0;
1486 if (!strcmp (prop, "any"))
1487 continue;
1488 if (!strcmp (prop, "host"))
1489 {
1490 #ifdef ACCEL_COMPILER
1491 return 0;
1492 #else
1493 if (omp_maybe_offloaded ())
1494 ret = -1;
1495 continue;
1496 #endif
1497 }
1498 if (!strcmp (prop, "nohost"))
1499 {
1500 #ifndef ACCEL_COMPILER
1501 if (omp_maybe_offloaded ())
1502 ret = -1;
1503 else
1504 return 0;
1505 #endif
1506 continue;
1507 }
1508 int r = 0;
1509 if (targetm.omp.device_kind_arch_isa != NULL)
1510 r = targetm.omp.device_kind_arch_isa (omp_device_kind,
1511 prop);
1512 else
1513 r = strcmp (prop, "cpu") == 0;
1514 if (r == 0 || (r == -1 && symtab->state != PARSING))
1515 {
1516 /* If we are or might be in a target region or
1517 declare target function, need to take into account
1518 also offloading values. */
1519 if (!omp_maybe_offloaded ())
1520 return 0;
1521 if (ENABLE_OFFLOADING)
1522 {
1523 const char *kinds = omp_offload_device_kind;
1524 if (omp_offload_device_kind_arch_isa (kinds, prop))
1525 {
1526 ret = -1;
1527 continue;
1528 }
1529 }
1530 return 0;
1531 }
1532 else if (r == -1)
1533 ret = -1;
1534 /* If kind matches on the host, it still might not match
1535 in the offloading region. */
1536 else if (omp_maybe_offloaded ())
1537 ret = -1;
1538 }
1539 break;
1540 case 'i':
1541 if (set == 'd' && !strcmp (sel, "isa"))
1542 for (tree t3 = TREE_VALUE (t2); t3; t3 = TREE_CHAIN (t3))
1543 {
1544 const char *isa = omp_context_name_list_prop (t3);
1545 if (isa == NULL)
1546 return 0;
1547 int r = 0;
1548 if (targetm.omp.device_kind_arch_isa != NULL)
1549 r = targetm.omp.device_kind_arch_isa (omp_device_isa,
1550 isa);
1551 if (r == 0 || (r == -1 && symtab->state != PARSING))
1552 {
1553 /* If isa is valid on the target, but not in the
1554 current function and current function has
1555 #pragma omp declare simd on it, some simd clones
1556 might have the isa added later on. */
1557 if (r == -1
1558 && targetm.simd_clone.compute_vecsize_and_simdlen
1559 && (cfun == NULL || !cfun->after_inlining))
1560 {
1561 tree attrs
1562 = DECL_ATTRIBUTES (current_function_decl);
1563 if (lookup_attribute ("omp declare simd", attrs))
1564 {
1565 ret = -1;
1566 continue;
1567 }
1568 }
1569 /* If we are or might be in a target region or
1570 declare target function, need to take into account
1571 also offloading values. */
1572 if (!omp_maybe_offloaded ())
1573 return 0;
1574 if (ENABLE_OFFLOADING)
1575 {
1576 const char *isas = omp_offload_device_isa;
1577 if (omp_offload_device_kind_arch_isa (isas, isa))
1578 {
1579 ret = -1;
1580 continue;
1581 }
1582 }
1583 return 0;
1584 }
1585 else if (r == -1)
1586 ret = -1;
1587 /* If isa matches on the host, it still might not match
1588 in the offloading region. */
1589 else if (omp_maybe_offloaded ())
1590 ret = -1;
1591 }
1592 break;
1593 case 'c':
1594 if (set == 'u' && !strcmp (sel, "condition"))
1595 for (tree t3 = TREE_VALUE (t2); t3; t3 = TREE_CHAIN (t3))
1596 if (TREE_PURPOSE (t3) == NULL_TREE)
1597 {
1598 if (integer_zerop (TREE_VALUE (t3)))
1599 return 0;
1600 if (integer_nonzerop (TREE_VALUE (t3)))
1601 break;
1602 ret = -1;
1603 }
1604 break;
1605 default:
1606 break;
1607 }
1608 }
1609 }
1610 return ret;
1611 }
1612
1613 /* Compare construct={simd} CLAUSES1 with CLAUSES2, return 0/-1/1/2 as
1614 in omp_context_selector_set_compare. */
1615
1616 static int
omp_construct_simd_compare(tree clauses1,tree clauses2)1617 omp_construct_simd_compare (tree clauses1, tree clauses2)
1618 {
1619 if (clauses1 == NULL_TREE)
1620 return clauses2 == NULL_TREE ? 0 : -1;
1621 if (clauses2 == NULL_TREE)
1622 return 1;
1623
1624 int r = 0;
1625 struct declare_variant_simd_data {
1626 bool inbranch, notinbranch;
1627 tree simdlen;
1628 auto_vec<tree,16> data_sharing;
1629 auto_vec<tree,16> aligned;
1630 declare_variant_simd_data ()
1631 : inbranch(false), notinbranch(false), simdlen(NULL_TREE) {}
1632 } data[2];
1633 unsigned int i;
1634 for (i = 0; i < 2; i++)
1635 for (tree c = i ? clauses2 : clauses1; c; c = OMP_CLAUSE_CHAIN (c))
1636 {
1637 vec<tree> *v;
1638 switch (OMP_CLAUSE_CODE (c))
1639 {
1640 case OMP_CLAUSE_INBRANCH:
1641 data[i].inbranch = true;
1642 continue;
1643 case OMP_CLAUSE_NOTINBRANCH:
1644 data[i].notinbranch = true;
1645 continue;
1646 case OMP_CLAUSE_SIMDLEN:
1647 data[i].simdlen = OMP_CLAUSE_SIMDLEN_EXPR (c);
1648 continue;
1649 case OMP_CLAUSE_UNIFORM:
1650 case OMP_CLAUSE_LINEAR:
1651 v = &data[i].data_sharing;
1652 break;
1653 case OMP_CLAUSE_ALIGNED:
1654 v = &data[i].aligned;
1655 break;
1656 default:
1657 gcc_unreachable ();
1658 }
1659 unsigned HOST_WIDE_INT argno = tree_to_uhwi (OMP_CLAUSE_DECL (c));
1660 if (argno >= v->length ())
1661 v->safe_grow_cleared (argno + 1, true);
1662 (*v)[argno] = c;
1663 }
1664 /* Here, r is used as a bitmask, 2 is set if CLAUSES1 has something
1665 CLAUSES2 doesn't, 1 is set if CLAUSES2 has something CLAUSES1
1666 doesn't. Thus, r == 3 implies return value 2, r == 1 implies
1667 -1, r == 2 implies 1 and r == 0 implies 0. */
1668 if (data[0].inbranch != data[1].inbranch)
1669 r |= data[0].inbranch ? 2 : 1;
1670 if (data[0].notinbranch != data[1].notinbranch)
1671 r |= data[0].notinbranch ? 2 : 1;
1672 if (!simple_cst_equal (data[0].simdlen, data[1].simdlen))
1673 {
1674 if (data[0].simdlen && data[1].simdlen)
1675 return 2;
1676 r |= data[0].simdlen ? 2 : 1;
1677 }
1678 if (data[0].data_sharing.length () < data[1].data_sharing.length ()
1679 || data[0].aligned.length () < data[1].aligned.length ())
1680 r |= 1;
1681 tree c1, c2;
1682 FOR_EACH_VEC_ELT (data[0].data_sharing, i, c1)
1683 {
1684 c2 = (i < data[1].data_sharing.length ()
1685 ? data[1].data_sharing[i] : NULL_TREE);
1686 if ((c1 == NULL_TREE) != (c2 == NULL_TREE))
1687 {
1688 r |= c1 != NULL_TREE ? 2 : 1;
1689 continue;
1690 }
1691 if (c1 == NULL_TREE)
1692 continue;
1693 if (OMP_CLAUSE_CODE (c1) != OMP_CLAUSE_CODE (c2))
1694 return 2;
1695 if (OMP_CLAUSE_CODE (c1) != OMP_CLAUSE_LINEAR)
1696 continue;
1697 if (OMP_CLAUSE_LINEAR_VARIABLE_STRIDE (c1)
1698 != OMP_CLAUSE_LINEAR_VARIABLE_STRIDE (c2))
1699 return 2;
1700 if (OMP_CLAUSE_LINEAR_KIND (c1) != OMP_CLAUSE_LINEAR_KIND (c2))
1701 return 2;
1702 if (!simple_cst_equal (OMP_CLAUSE_LINEAR_STEP (c1),
1703 OMP_CLAUSE_LINEAR_STEP (c2)))
1704 return 2;
1705 }
1706 FOR_EACH_VEC_ELT (data[0].aligned, i, c1)
1707 {
1708 c2 = i < data[1].aligned.length () ? data[1].aligned[i] : NULL_TREE;
1709 if ((c1 == NULL_TREE) != (c2 == NULL_TREE))
1710 {
1711 r |= c1 != NULL_TREE ? 2 : 1;
1712 continue;
1713 }
1714 if (c1 == NULL_TREE)
1715 continue;
1716 if (!simple_cst_equal (OMP_CLAUSE_ALIGNED_ALIGNMENT (c1),
1717 OMP_CLAUSE_ALIGNED_ALIGNMENT (c2)))
1718 return 2;
1719 }
1720 switch (r)
1721 {
1722 case 0: return 0;
1723 case 1: return -1;
1724 case 2: return 1;
1725 case 3: return 2;
1726 default: gcc_unreachable ();
1727 }
1728 }
1729
1730 /* Compare properties of selectors SEL from SET other than construct.
1731 Return 0/-1/1/2 as in omp_context_selector_set_compare.
1732 Unlike set names or selector names, properties can have duplicates. */
1733
1734 static int
omp_context_selector_props_compare(const char * set,const char * sel,tree ctx1,tree ctx2)1735 omp_context_selector_props_compare (const char *set, const char *sel,
1736 tree ctx1, tree ctx2)
1737 {
1738 int ret = 0;
1739 for (int pass = 0; pass < 2; pass++)
1740 for (tree t1 = pass ? ctx2 : ctx1; t1; t1 = TREE_CHAIN (t1))
1741 {
1742 tree t2;
1743 for (t2 = pass ? ctx1 : ctx2; t2; t2 = TREE_CHAIN (t2))
1744 if (TREE_PURPOSE (t1) == TREE_PURPOSE (t2))
1745 {
1746 if (TREE_PURPOSE (t1) == NULL_TREE)
1747 {
1748 if (set[0] == 'u' && strcmp (sel, "condition") == 0)
1749 {
1750 if (integer_zerop (TREE_VALUE (t1))
1751 != integer_zerop (TREE_VALUE (t2)))
1752 return 2;
1753 break;
1754 }
1755 if (simple_cst_equal (TREE_VALUE (t1), TREE_VALUE (t2)))
1756 break;
1757 }
1758 else if (strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t1)),
1759 " score") == 0)
1760 {
1761 if (!simple_cst_equal (TREE_VALUE (t1), TREE_VALUE (t2)))
1762 return 2;
1763 break;
1764 }
1765 else
1766 break;
1767 }
1768 else if (TREE_PURPOSE (t1)
1769 && TREE_PURPOSE (t2) == NULL_TREE
1770 && TREE_CODE (TREE_VALUE (t2)) == STRING_CST)
1771 {
1772 const char *p1 = omp_context_name_list_prop (t1);
1773 const char *p2 = omp_context_name_list_prop (t2);
1774 if (p2
1775 && strcmp (p1, p2) == 0
1776 && strcmp (p1, " score"))
1777 break;
1778 }
1779 else if (TREE_PURPOSE (t1) == NULL_TREE
1780 && TREE_PURPOSE (t2)
1781 && TREE_CODE (TREE_VALUE (t1)) == STRING_CST)
1782 {
1783 const char *p1 = omp_context_name_list_prop (t1);
1784 const char *p2 = omp_context_name_list_prop (t2);
1785 if (p1
1786 && strcmp (p1, p2) == 0
1787 && strcmp (p1, " score"))
1788 break;
1789 }
1790 if (t2 == NULL_TREE)
1791 {
1792 int r = pass ? -1 : 1;
1793 if (ret && ret != r)
1794 return 2;
1795 else if (pass)
1796 return r;
1797 else
1798 {
1799 ret = r;
1800 break;
1801 }
1802 }
1803 }
1804 return ret;
1805 }
1806
1807 /* Compare single context selector sets CTX1 and CTX2 with SET name.
1808 Return 0 if CTX1 is equal to CTX2,
1809 -1 if CTX1 is a strict subset of CTX2,
1810 1 if CTX2 is a strict subset of CTX1, or
1811 2 if neither context is a subset of another one. */
1812
1813 int
omp_context_selector_set_compare(const char * set,tree ctx1,tree ctx2)1814 omp_context_selector_set_compare (const char *set, tree ctx1, tree ctx2)
1815 {
1816 bool swapped = false;
1817 int ret = 0;
1818 int len1 = list_length (ctx1);
1819 int len2 = list_length (ctx2);
1820 int cnt = 0;
1821 if (len1 < len2)
1822 {
1823 swapped = true;
1824 std::swap (ctx1, ctx2);
1825 std::swap (len1, len2);
1826 }
1827 if (set[0] == 'c')
1828 {
1829 tree t1;
1830 tree t2 = ctx2;
1831 tree simd = get_identifier ("simd");
1832 /* Handle construct set specially. In this case the order
1833 of the selector matters too. */
1834 for (t1 = ctx1; t1; t1 = TREE_CHAIN (t1))
1835 if (TREE_PURPOSE (t1) == TREE_PURPOSE (t2))
1836 {
1837 int r = 0;
1838 if (TREE_PURPOSE (t1) == simd)
1839 r = omp_construct_simd_compare (TREE_VALUE (t1),
1840 TREE_VALUE (t2));
1841 if (r == 2 || (ret && r && (ret < 0) != (r < 0)))
1842 return 2;
1843 if (ret == 0)
1844 ret = r;
1845 t2 = TREE_CHAIN (t2);
1846 if (t2 == NULL_TREE)
1847 {
1848 t1 = TREE_CHAIN (t1);
1849 break;
1850 }
1851 }
1852 else if (ret < 0)
1853 return 2;
1854 else
1855 ret = 1;
1856 if (t2 != NULL_TREE)
1857 return 2;
1858 if (t1 != NULL_TREE)
1859 {
1860 if (ret < 0)
1861 return 2;
1862 ret = 1;
1863 }
1864 if (ret == 0)
1865 return 0;
1866 return swapped ? -ret : ret;
1867 }
1868 for (tree t1 = ctx1; t1; t1 = TREE_CHAIN (t1))
1869 {
1870 tree t2;
1871 for (t2 = ctx2; t2; t2 = TREE_CHAIN (t2))
1872 if (TREE_PURPOSE (t1) == TREE_PURPOSE (t2))
1873 {
1874 const char *sel = IDENTIFIER_POINTER (TREE_PURPOSE (t1));
1875 int r = omp_context_selector_props_compare (set, sel,
1876 TREE_VALUE (t1),
1877 TREE_VALUE (t2));
1878 if (r == 2 || (ret && r && (ret < 0) != (r < 0)))
1879 return 2;
1880 if (ret == 0)
1881 ret = r;
1882 cnt++;
1883 break;
1884 }
1885 if (t2 == NULL_TREE)
1886 {
1887 if (ret == -1)
1888 return 2;
1889 ret = 1;
1890 }
1891 }
1892 if (cnt < len2)
1893 return 2;
1894 if (ret == 0)
1895 return 0;
1896 return swapped ? -ret : ret;
1897 }
1898
1899 /* Compare whole context selector specification CTX1 and CTX2.
1900 Return 0 if CTX1 is equal to CTX2,
1901 -1 if CTX1 is a strict subset of CTX2,
1902 1 if CTX2 is a strict subset of CTX1, or
1903 2 if neither context is a subset of another one. */
1904
1905 static int
omp_context_selector_compare(tree ctx1,tree ctx2)1906 omp_context_selector_compare (tree ctx1, tree ctx2)
1907 {
1908 bool swapped = false;
1909 int ret = 0;
1910 int len1 = list_length (ctx1);
1911 int len2 = list_length (ctx2);
1912 int cnt = 0;
1913 if (len1 < len2)
1914 {
1915 swapped = true;
1916 std::swap (ctx1, ctx2);
1917 std::swap (len1, len2);
1918 }
1919 for (tree t1 = ctx1; t1; t1 = TREE_CHAIN (t1))
1920 {
1921 tree t2;
1922 for (t2 = ctx2; t2; t2 = TREE_CHAIN (t2))
1923 if (TREE_PURPOSE (t1) == TREE_PURPOSE (t2))
1924 {
1925 const char *set = IDENTIFIER_POINTER (TREE_PURPOSE (t1));
1926 int r = omp_context_selector_set_compare (set, TREE_VALUE (t1),
1927 TREE_VALUE (t2));
1928 if (r == 2 || (ret && r && (ret < 0) != (r < 0)))
1929 return 2;
1930 if (ret == 0)
1931 ret = r;
1932 cnt++;
1933 break;
1934 }
1935 if (t2 == NULL_TREE)
1936 {
1937 if (ret == -1)
1938 return 2;
1939 ret = 1;
1940 }
1941 }
1942 if (cnt < len2)
1943 return 2;
1944 if (ret == 0)
1945 return 0;
1946 return swapped ? -ret : ret;
1947 }
1948
1949 /* From context selector CTX, return trait-selector with name SEL in
1950 trait-selector-set with name SET if any, or NULL_TREE if not found.
1951 If SEL is NULL, return the list of trait-selectors in SET. */
1952
1953 tree
omp_get_context_selector(tree ctx,const char * set,const char * sel)1954 omp_get_context_selector (tree ctx, const char *set, const char *sel)
1955 {
1956 tree setid = get_identifier (set);
1957 tree selid = sel ? get_identifier (sel) : NULL_TREE;
1958 for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1))
1959 if (TREE_PURPOSE (t1) == setid)
1960 {
1961 if (sel == NULL)
1962 return TREE_VALUE (t1);
1963 for (tree t2 = TREE_VALUE (t1); t2; t2 = TREE_CHAIN (t2))
1964 if (TREE_PURPOSE (t2) == selid)
1965 return t2;
1966 }
1967 return NULL_TREE;
1968 }
1969
1970 /* Compute *SCORE for context selector CTX. Return true if the score
1971 would be different depending on whether it is a declare simd clone or
1972 not. DECLARE_SIMD should be true for the case when it would be
1973 a declare simd clone. */
1974
1975 static bool
omp_context_compute_score(tree ctx,widest_int * score,bool declare_simd)1976 omp_context_compute_score (tree ctx, widest_int *score, bool declare_simd)
1977 {
1978 tree construct = omp_get_context_selector (ctx, "construct", NULL);
1979 bool has_kind = omp_get_context_selector (ctx, "device", "kind");
1980 bool has_arch = omp_get_context_selector (ctx, "device", "arch");
1981 bool has_isa = omp_get_context_selector (ctx, "device", "isa");
1982 bool ret = false;
1983 *score = 1;
1984 for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1))
1985 if (TREE_VALUE (t1) != construct)
1986 for (tree t2 = TREE_VALUE (t1); t2; t2 = TREE_CHAIN (t2))
1987 if (tree t3 = TREE_VALUE (t2))
1988 if (TREE_PURPOSE (t3)
1989 && strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t3)), " score") == 0
1990 && TREE_CODE (TREE_VALUE (t3)) == INTEGER_CST)
1991 *score += wi::to_widest (TREE_VALUE (t3));
1992 if (construct || has_kind || has_arch || has_isa)
1993 {
1994 int scores[12];
1995 enum tree_code constructs[5];
1996 int nconstructs = 0;
1997 if (construct)
1998 nconstructs = omp_constructor_traits_to_codes (construct, constructs);
1999 if (omp_construct_selector_matches (constructs, nconstructs, scores)
2000 == 2)
2001 ret = true;
2002 int b = declare_simd ? nconstructs + 1 : 0;
2003 if (scores[b + nconstructs] + 4U < score->get_precision ())
2004 {
2005 for (int n = 0; n < nconstructs; ++n)
2006 {
2007 if (scores[b + n] < 0)
2008 {
2009 *score = -1;
2010 return ret;
2011 }
2012 *score += wi::shifted_mask <widest_int> (scores[b + n], 1, false);
2013 }
2014 if (has_kind)
2015 *score += wi::shifted_mask <widest_int> (scores[b + nconstructs],
2016 1, false);
2017 if (has_arch)
2018 *score += wi::shifted_mask <widest_int> (scores[b + nconstructs] + 1,
2019 1, false);
2020 if (has_isa)
2021 *score += wi::shifted_mask <widest_int> (scores[b + nconstructs] + 2,
2022 1, false);
2023 }
2024 else /* FIXME: Implement this. */
2025 gcc_unreachable ();
2026 }
2027 return ret;
2028 }
2029
2030 /* Class describing a single variant. */
2031 struct GTY(()) omp_declare_variant_entry {
2032 /* NODE of the variant. */
2033 cgraph_node *variant;
2034 /* Score if not in declare simd clone. */
2035 widest_int score;
2036 /* Score if in declare simd clone. */
2037 widest_int score_in_declare_simd_clone;
2038 /* Context selector for the variant. */
2039 tree ctx;
2040 /* True if the context selector is known to match already. */
2041 bool matches;
2042 };
2043
2044 /* Class describing a function with variants. */
2045 struct GTY((for_user)) omp_declare_variant_base_entry {
2046 /* NODE of the base function. */
2047 cgraph_node *base;
2048 /* NODE of the artificial function created for the deferred variant
2049 resolution. */
2050 cgraph_node *node;
2051 /* Vector of the variants. */
2052 vec<omp_declare_variant_entry, va_gc> *variants;
2053 };
2054
2055 struct omp_declare_variant_hasher
2056 : ggc_ptr_hash<omp_declare_variant_base_entry> {
2057 static hashval_t hash (omp_declare_variant_base_entry *);
2058 static bool equal (omp_declare_variant_base_entry *,
2059 omp_declare_variant_base_entry *);
2060 };
2061
2062 hashval_t
hash(omp_declare_variant_base_entry * x)2063 omp_declare_variant_hasher::hash (omp_declare_variant_base_entry *x)
2064 {
2065 inchash::hash hstate;
2066 hstate.add_int (DECL_UID (x->base->decl));
2067 hstate.add_int (x->variants->length ());
2068 omp_declare_variant_entry *variant;
2069 unsigned int i;
2070 FOR_EACH_VEC_SAFE_ELT (x->variants, i, variant)
2071 {
2072 hstate.add_int (DECL_UID (variant->variant->decl));
2073 hstate.add_wide_int (variant->score);
2074 hstate.add_wide_int (variant->score_in_declare_simd_clone);
2075 hstate.add_ptr (variant->ctx);
2076 hstate.add_int (variant->matches);
2077 }
2078 return hstate.end ();
2079 }
2080
2081 bool
equal(omp_declare_variant_base_entry * x,omp_declare_variant_base_entry * y)2082 omp_declare_variant_hasher::equal (omp_declare_variant_base_entry *x,
2083 omp_declare_variant_base_entry *y)
2084 {
2085 if (x->base != y->base
2086 || x->variants->length () != y->variants->length ())
2087 return false;
2088 omp_declare_variant_entry *variant;
2089 unsigned int i;
2090 FOR_EACH_VEC_SAFE_ELT (x->variants, i, variant)
2091 if (variant->variant != (*y->variants)[i].variant
2092 || variant->score != (*y->variants)[i].score
2093 || (variant->score_in_declare_simd_clone
2094 != (*y->variants)[i].score_in_declare_simd_clone)
2095 || variant->ctx != (*y->variants)[i].ctx
2096 || variant->matches != (*y->variants)[i].matches)
2097 return false;
2098 return true;
2099 }
2100
2101 static GTY(()) hash_table<omp_declare_variant_hasher> *omp_declare_variants;
2102
2103 struct omp_declare_variant_alt_hasher
2104 : ggc_ptr_hash<omp_declare_variant_base_entry> {
2105 static hashval_t hash (omp_declare_variant_base_entry *);
2106 static bool equal (omp_declare_variant_base_entry *,
2107 omp_declare_variant_base_entry *);
2108 };
2109
2110 hashval_t
hash(omp_declare_variant_base_entry * x)2111 omp_declare_variant_alt_hasher::hash (omp_declare_variant_base_entry *x)
2112 {
2113 return DECL_UID (x->node->decl);
2114 }
2115
2116 bool
equal(omp_declare_variant_base_entry * x,omp_declare_variant_base_entry * y)2117 omp_declare_variant_alt_hasher::equal (omp_declare_variant_base_entry *x,
2118 omp_declare_variant_base_entry *y)
2119 {
2120 return x->node == y->node;
2121 }
2122
2123 static GTY(()) hash_table<omp_declare_variant_alt_hasher>
2124 *omp_declare_variant_alt;
2125
2126 /* Try to resolve declare variant after gimplification. */
2127
2128 static tree
omp_resolve_late_declare_variant(tree alt)2129 omp_resolve_late_declare_variant (tree alt)
2130 {
2131 cgraph_node *node = cgraph_node::get (alt);
2132 cgraph_node *cur_node = cgraph_node::get (cfun->decl);
2133 if (node == NULL
2134 || !node->declare_variant_alt
2135 || !cfun->after_inlining)
2136 return alt;
2137
2138 omp_declare_variant_base_entry entry;
2139 entry.base = NULL;
2140 entry.node = node;
2141 entry.variants = NULL;
2142 omp_declare_variant_base_entry *entryp
2143 = omp_declare_variant_alt->find_with_hash (&entry, DECL_UID (alt));
2144
2145 unsigned int i, j;
2146 omp_declare_variant_entry *varentry1, *varentry2;
2147 auto_vec <bool, 16> matches;
2148 unsigned int nmatches = 0;
2149 FOR_EACH_VEC_SAFE_ELT (entryp->variants, i, varentry1)
2150 {
2151 if (varentry1->matches)
2152 {
2153 /* This has been checked to be ok already. */
2154 matches.safe_push (true);
2155 nmatches++;
2156 continue;
2157 }
2158 switch (omp_context_selector_matches (varentry1->ctx))
2159 {
2160 case 0:
2161 matches.safe_push (false);
2162 break;
2163 case -1:
2164 return alt;
2165 default:
2166 matches.safe_push (true);
2167 nmatches++;
2168 break;
2169 }
2170 }
2171
2172 if (nmatches == 0)
2173 return entryp->base->decl;
2174
2175 /* A context selector that is a strict subset of another context selector
2176 has a score of zero. */
2177 FOR_EACH_VEC_SAFE_ELT (entryp->variants, i, varentry1)
2178 if (matches[i])
2179 {
2180 for (j = i + 1;
2181 vec_safe_iterate (entryp->variants, j, &varentry2); ++j)
2182 if (matches[j])
2183 {
2184 int r = omp_context_selector_compare (varentry1->ctx,
2185 varentry2->ctx);
2186 if (r == -1)
2187 {
2188 /* ctx1 is a strict subset of ctx2, ignore ctx1. */
2189 matches[i] = false;
2190 break;
2191 }
2192 else if (r == 1)
2193 /* ctx2 is a strict subset of ctx1, remove ctx2. */
2194 matches[j] = false;
2195 }
2196 }
2197
2198 widest_int max_score = -1;
2199 varentry2 = NULL;
2200 FOR_EACH_VEC_SAFE_ELT (entryp->variants, i, varentry1)
2201 if (matches[i])
2202 {
2203 widest_int score
2204 = (cur_node->simdclone ? varentry1->score_in_declare_simd_clone
2205 : varentry1->score);
2206 if (score > max_score)
2207 {
2208 max_score = score;
2209 varentry2 = varentry1;
2210 }
2211 }
2212 return varentry2->variant->decl;
2213 }
2214
2215 /* Hook to adjust hash tables on cgraph_node removal. */
2216
2217 static void
omp_declare_variant_remove_hook(struct cgraph_node * node,void *)2218 omp_declare_variant_remove_hook (struct cgraph_node *node, void *)
2219 {
2220 if (!node->declare_variant_alt)
2221 return;
2222
2223 /* Drop this hash table completely. */
2224 omp_declare_variants = NULL;
2225 /* And remove node from the other hash table. */
2226 if (omp_declare_variant_alt)
2227 {
2228 omp_declare_variant_base_entry entry;
2229 entry.base = NULL;
2230 entry.node = node;
2231 entry.variants = NULL;
2232 omp_declare_variant_alt->remove_elt_with_hash (&entry,
2233 DECL_UID (node->decl));
2234 }
2235 }
2236
2237 /* Try to resolve declare variant, return the variant decl if it should
2238 be used instead of base, or base otherwise. */
2239
2240 tree
omp_resolve_declare_variant(tree base)2241 omp_resolve_declare_variant (tree base)
2242 {
2243 tree variant1 = NULL_TREE, variant2 = NULL_TREE;
2244 if (cfun && (cfun->curr_properties & PROP_gimple_any) != 0)
2245 return omp_resolve_late_declare_variant (base);
2246
2247 auto_vec <tree, 16> variants;
2248 auto_vec <bool, 16> defer;
2249 bool any_deferred = false;
2250 for (tree attr = DECL_ATTRIBUTES (base); attr; attr = TREE_CHAIN (attr))
2251 {
2252 attr = lookup_attribute ("omp declare variant base", attr);
2253 if (attr == NULL_TREE)
2254 break;
2255 if (TREE_CODE (TREE_PURPOSE (TREE_VALUE (attr))) != FUNCTION_DECL)
2256 continue;
2257 cgraph_node *node = cgraph_node::get (base);
2258 /* If this is already a magic decl created by this function,
2259 don't process it again. */
2260 if (node && node->declare_variant_alt)
2261 return base;
2262 switch (omp_context_selector_matches (TREE_VALUE (TREE_VALUE (attr))))
2263 {
2264 case 0:
2265 /* No match, ignore. */
2266 break;
2267 case -1:
2268 /* Needs to be deferred. */
2269 any_deferred = true;
2270 variants.safe_push (attr);
2271 defer.safe_push (true);
2272 break;
2273 default:
2274 variants.safe_push (attr);
2275 defer.safe_push (false);
2276 break;
2277 }
2278 }
2279 if (variants.length () == 0)
2280 return base;
2281
2282 if (any_deferred)
2283 {
2284 widest_int max_score1 = 0;
2285 widest_int max_score2 = 0;
2286 bool first = true;
2287 unsigned int i;
2288 tree attr1, attr2;
2289 omp_declare_variant_base_entry entry;
2290 entry.base = cgraph_node::get_create (base);
2291 entry.node = NULL;
2292 vec_alloc (entry.variants, variants.length ());
2293 FOR_EACH_VEC_ELT (variants, i, attr1)
2294 {
2295 widest_int score1;
2296 widest_int score2;
2297 bool need_two;
2298 tree ctx = TREE_VALUE (TREE_VALUE (attr1));
2299 need_two = omp_context_compute_score (ctx, &score1, false);
2300 if (need_two)
2301 omp_context_compute_score (ctx, &score2, true);
2302 else
2303 score2 = score1;
2304 if (first)
2305 {
2306 first = false;
2307 max_score1 = score1;
2308 max_score2 = score2;
2309 if (!defer[i])
2310 {
2311 variant1 = attr1;
2312 variant2 = attr1;
2313 }
2314 }
2315 else
2316 {
2317 if (max_score1 == score1)
2318 variant1 = NULL_TREE;
2319 else if (score1 > max_score1)
2320 {
2321 max_score1 = score1;
2322 variant1 = defer[i] ? NULL_TREE : attr1;
2323 }
2324 if (max_score2 == score2)
2325 variant2 = NULL_TREE;
2326 else if (score2 > max_score2)
2327 {
2328 max_score2 = score2;
2329 variant2 = defer[i] ? NULL_TREE : attr1;
2330 }
2331 }
2332 omp_declare_variant_entry varentry;
2333 varentry.variant
2334 = cgraph_node::get_create (TREE_PURPOSE (TREE_VALUE (attr1)));
2335 varentry.score = score1;
2336 varentry.score_in_declare_simd_clone = score2;
2337 varentry.ctx = ctx;
2338 varentry.matches = !defer[i];
2339 entry.variants->quick_push (varentry);
2340 }
2341
2342 /* If there is a clear winner variant with the score which is not
2343 deferred, verify it is not a strict subset of any other context
2344 selector and if it is not, it is the best alternative no matter
2345 whether the others do or don't match. */
2346 if (variant1 && variant1 == variant2)
2347 {
2348 tree ctx1 = TREE_VALUE (TREE_VALUE (variant1));
2349 FOR_EACH_VEC_ELT (variants, i, attr2)
2350 {
2351 if (attr2 == variant1)
2352 continue;
2353 tree ctx2 = TREE_VALUE (TREE_VALUE (attr2));
2354 int r = omp_context_selector_compare (ctx1, ctx2);
2355 if (r == -1)
2356 {
2357 /* The winner is a strict subset of ctx2, can't
2358 decide now. */
2359 variant1 = NULL_TREE;
2360 break;
2361 }
2362 }
2363 if (variant1)
2364 {
2365 vec_free (entry.variants);
2366 return TREE_PURPOSE (TREE_VALUE (variant1));
2367 }
2368 }
2369
2370 static struct cgraph_node_hook_list *node_removal_hook_holder;
2371 if (!node_removal_hook_holder)
2372 node_removal_hook_holder
2373 = symtab->add_cgraph_removal_hook (omp_declare_variant_remove_hook,
2374 NULL);
2375
2376 if (omp_declare_variants == NULL)
2377 omp_declare_variants
2378 = hash_table<omp_declare_variant_hasher>::create_ggc (64);
2379 omp_declare_variant_base_entry **slot
2380 = omp_declare_variants->find_slot (&entry, INSERT);
2381 if (*slot != NULL)
2382 {
2383 vec_free (entry.variants);
2384 return (*slot)->node->decl;
2385 }
2386
2387 *slot = ggc_cleared_alloc<omp_declare_variant_base_entry> ();
2388 (*slot)->base = entry.base;
2389 (*slot)->node = entry.base;
2390 (*slot)->variants = entry.variants;
2391 tree alt = build_decl (DECL_SOURCE_LOCATION (base), FUNCTION_DECL,
2392 DECL_NAME (base), TREE_TYPE (base));
2393 DECL_ARTIFICIAL (alt) = 1;
2394 DECL_IGNORED_P (alt) = 1;
2395 TREE_STATIC (alt) = 1;
2396 tree attributes = DECL_ATTRIBUTES (base);
2397 if (lookup_attribute ("noipa", attributes) == NULL)
2398 {
2399 attributes = tree_cons (get_identifier ("noipa"), NULL, attributes);
2400 if (lookup_attribute ("noinline", attributes) == NULL)
2401 attributes = tree_cons (get_identifier ("noinline"), NULL,
2402 attributes);
2403 if (lookup_attribute ("noclone", attributes) == NULL)
2404 attributes = tree_cons (get_identifier ("noclone"), NULL,
2405 attributes);
2406 if (lookup_attribute ("no_icf", attributes) == NULL)
2407 attributes = tree_cons (get_identifier ("no_icf"), NULL,
2408 attributes);
2409 }
2410 DECL_ATTRIBUTES (alt) = attributes;
2411 DECL_INITIAL (alt) = error_mark_node;
2412 (*slot)->node = cgraph_node::create (alt);
2413 (*slot)->node->declare_variant_alt = 1;
2414 (*slot)->node->create_reference (entry.base, IPA_REF_ADDR);
2415 omp_declare_variant_entry *varentry;
2416 FOR_EACH_VEC_SAFE_ELT (entry.variants, i, varentry)
2417 (*slot)->node->create_reference (varentry->variant, IPA_REF_ADDR);
2418 if (omp_declare_variant_alt == NULL)
2419 omp_declare_variant_alt
2420 = hash_table<omp_declare_variant_alt_hasher>::create_ggc (64);
2421 *omp_declare_variant_alt->find_slot_with_hash (*slot, DECL_UID (alt),
2422 INSERT) = *slot;
2423 return alt;
2424 }
2425
2426 if (variants.length () == 1)
2427 return TREE_PURPOSE (TREE_VALUE (variants[0]));
2428
2429 /* A context selector that is a strict subset of another context selector
2430 has a score of zero. */
2431 tree attr1, attr2;
2432 unsigned int i, j;
2433 FOR_EACH_VEC_ELT (variants, i, attr1)
2434 if (attr1)
2435 {
2436 tree ctx1 = TREE_VALUE (TREE_VALUE (attr1));
2437 FOR_EACH_VEC_ELT_FROM (variants, j, attr2, i + 1)
2438 if (attr2)
2439 {
2440 tree ctx2 = TREE_VALUE (TREE_VALUE (attr2));
2441 int r = omp_context_selector_compare (ctx1, ctx2);
2442 if (r == -1)
2443 {
2444 /* ctx1 is a strict subset of ctx2, remove
2445 attr1 from the vector. */
2446 variants[i] = NULL_TREE;
2447 break;
2448 }
2449 else if (r == 1)
2450 /* ctx2 is a strict subset of ctx1, remove attr2
2451 from the vector. */
2452 variants[j] = NULL_TREE;
2453 }
2454 }
2455 widest_int max_score1 = 0;
2456 widest_int max_score2 = 0;
2457 bool first = true;
2458 FOR_EACH_VEC_ELT (variants, i, attr1)
2459 if (attr1)
2460 {
2461 if (variant1)
2462 {
2463 widest_int score1;
2464 widest_int score2;
2465 bool need_two;
2466 tree ctx;
2467 if (first)
2468 {
2469 first = false;
2470 ctx = TREE_VALUE (TREE_VALUE (variant1));
2471 need_two = omp_context_compute_score (ctx, &max_score1, false);
2472 if (need_two)
2473 omp_context_compute_score (ctx, &max_score2, true);
2474 else
2475 max_score2 = max_score1;
2476 }
2477 ctx = TREE_VALUE (TREE_VALUE (attr1));
2478 need_two = omp_context_compute_score (ctx, &score1, false);
2479 if (need_two)
2480 omp_context_compute_score (ctx, &score2, true);
2481 else
2482 score2 = score1;
2483 if (score1 > max_score1)
2484 {
2485 max_score1 = score1;
2486 variant1 = attr1;
2487 }
2488 if (score2 > max_score2)
2489 {
2490 max_score2 = score2;
2491 variant2 = attr1;
2492 }
2493 }
2494 else
2495 {
2496 variant1 = attr1;
2497 variant2 = attr1;
2498 }
2499 }
2500 /* If there is a disagreement on which variant has the highest score
2501 depending on whether it will be in a declare simd clone or not,
2502 punt for now and defer until after IPA where we will know that. */
2503 return ((variant1 && variant1 == variant2)
2504 ? TREE_PURPOSE (TREE_VALUE (variant1)) : base);
2505 }
2506
2507 void
omp_lto_output_declare_variant_alt(lto_simple_output_block * ob,cgraph_node * node,lto_symtab_encoder_t encoder)2508 omp_lto_output_declare_variant_alt (lto_simple_output_block *ob,
2509 cgraph_node *node,
2510 lto_symtab_encoder_t encoder)
2511 {
2512 gcc_assert (node->declare_variant_alt);
2513
2514 omp_declare_variant_base_entry entry;
2515 entry.base = NULL;
2516 entry.node = node;
2517 entry.variants = NULL;
2518 omp_declare_variant_base_entry *entryp
2519 = omp_declare_variant_alt->find_with_hash (&entry, DECL_UID (node->decl));
2520 gcc_assert (entryp);
2521
2522 int nbase = lto_symtab_encoder_lookup (encoder, entryp->base);
2523 gcc_assert (nbase != LCC_NOT_FOUND);
2524 streamer_write_hwi_stream (ob->main_stream, nbase);
2525
2526 streamer_write_hwi_stream (ob->main_stream, entryp->variants->length ());
2527
2528 unsigned int i;
2529 omp_declare_variant_entry *varentry;
2530 FOR_EACH_VEC_SAFE_ELT (entryp->variants, i, varentry)
2531 {
2532 int nvar = lto_symtab_encoder_lookup (encoder, varentry->variant);
2533 gcc_assert (nvar != LCC_NOT_FOUND);
2534 streamer_write_hwi_stream (ob->main_stream, nvar);
2535
2536 for (widest_int *w = &varentry->score; ;
2537 w = &varentry->score_in_declare_simd_clone)
2538 {
2539 unsigned len = w->get_len ();
2540 streamer_write_hwi_stream (ob->main_stream, len);
2541 const HOST_WIDE_INT *val = w->get_val ();
2542 for (unsigned j = 0; j < len; j++)
2543 streamer_write_hwi_stream (ob->main_stream, val[j]);
2544 if (w == &varentry->score_in_declare_simd_clone)
2545 break;
2546 }
2547
2548 HOST_WIDE_INT cnt = -1;
2549 HOST_WIDE_INT i = varentry->matches ? 1 : 0;
2550 for (tree attr = DECL_ATTRIBUTES (entryp->base->decl);
2551 attr; attr = TREE_CHAIN (attr), i += 2)
2552 {
2553 attr = lookup_attribute ("omp declare variant base", attr);
2554 if (attr == NULL_TREE)
2555 break;
2556
2557 if (varentry->ctx == TREE_VALUE (TREE_VALUE (attr)))
2558 {
2559 cnt = i;
2560 break;
2561 }
2562 }
2563
2564 gcc_assert (cnt != -1);
2565 streamer_write_hwi_stream (ob->main_stream, cnt);
2566 }
2567 }
2568
2569 void
omp_lto_input_declare_variant_alt(lto_input_block * ib,cgraph_node * node,vec<symtab_node * > nodes)2570 omp_lto_input_declare_variant_alt (lto_input_block *ib, cgraph_node *node,
2571 vec<symtab_node *> nodes)
2572 {
2573 gcc_assert (node->declare_variant_alt);
2574 omp_declare_variant_base_entry *entryp
2575 = ggc_cleared_alloc<omp_declare_variant_base_entry> ();
2576 entryp->base = dyn_cast<cgraph_node *> (nodes[streamer_read_hwi (ib)]);
2577 entryp->node = node;
2578 unsigned int len = streamer_read_hwi (ib);
2579 vec_alloc (entryp->variants, len);
2580
2581 for (unsigned int i = 0; i < len; i++)
2582 {
2583 omp_declare_variant_entry varentry;
2584 varentry.variant
2585 = dyn_cast<cgraph_node *> (nodes[streamer_read_hwi (ib)]);
2586 for (widest_int *w = &varentry.score; ;
2587 w = &varentry.score_in_declare_simd_clone)
2588 {
2589 unsigned len2 = streamer_read_hwi (ib);
2590 HOST_WIDE_INT arr[WIDE_INT_MAX_ELTS];
2591 gcc_assert (len2 <= WIDE_INT_MAX_ELTS);
2592 for (unsigned int j = 0; j < len2; j++)
2593 arr[j] = streamer_read_hwi (ib);
2594 *w = widest_int::from_array (arr, len2, true);
2595 if (w == &varentry.score_in_declare_simd_clone)
2596 break;
2597 }
2598
2599 HOST_WIDE_INT cnt = streamer_read_hwi (ib);
2600 HOST_WIDE_INT j = 0;
2601 varentry.ctx = NULL_TREE;
2602 varentry.matches = (cnt & 1) ? true : false;
2603 cnt &= ~HOST_WIDE_INT_1;
2604 for (tree attr = DECL_ATTRIBUTES (entryp->base->decl);
2605 attr; attr = TREE_CHAIN (attr), j += 2)
2606 {
2607 attr = lookup_attribute ("omp declare variant base", attr);
2608 if (attr == NULL_TREE)
2609 break;
2610
2611 if (cnt == j)
2612 {
2613 varentry.ctx = TREE_VALUE (TREE_VALUE (attr));
2614 break;
2615 }
2616 }
2617 gcc_assert (varentry.ctx != NULL_TREE);
2618 entryp->variants->quick_push (varentry);
2619 }
2620 if (omp_declare_variant_alt == NULL)
2621 omp_declare_variant_alt
2622 = hash_table<omp_declare_variant_alt_hasher>::create_ggc (64);
2623 *omp_declare_variant_alt->find_slot_with_hash (entryp, DECL_UID (node->decl),
2624 INSERT) = entryp;
2625 }
2626
2627 /* Encode an oacc launch argument. This matches the GOMP_LAUNCH_PACK
2628 macro on gomp-constants.h. We do not check for overflow. */
2629
2630 tree
oacc_launch_pack(unsigned code,tree device,unsigned op)2631 oacc_launch_pack (unsigned code, tree device, unsigned op)
2632 {
2633 tree res;
2634
2635 res = build_int_cst (unsigned_type_node, GOMP_LAUNCH_PACK (code, 0, op));
2636 if (device)
2637 {
2638 device = fold_build2 (LSHIFT_EXPR, unsigned_type_node,
2639 device, build_int_cst (unsigned_type_node,
2640 GOMP_LAUNCH_DEVICE_SHIFT));
2641 res = fold_build2 (BIT_IOR_EXPR, unsigned_type_node, res, device);
2642 }
2643 return res;
2644 }
2645
2646 /* FIXME: What is the following comment for? */
2647 /* Look for compute grid dimension clauses and convert to an attribute
2648 attached to FN. This permits the target-side code to (a) massage
2649 the dimensions, (b) emit that data and (c) optimize. Non-constant
2650 dimensions are pushed onto ARGS.
2651
2652 The attribute value is a TREE_LIST. A set of dimensions is
2653 represented as a list of INTEGER_CST. Those that are runtime
2654 exprs are represented as an INTEGER_CST of zero.
2655
2656 TODO: Normally the attribute will just contain a single such list. If
2657 however it contains a list of lists, this will represent the use of
2658 device_type. Each member of the outer list is an assoc list of
2659 dimensions, keyed by the device type. The first entry will be the
2660 default. Well, that's the plan. */
2661
2662 /* Replace any existing oacc fn attribute with updated dimensions. */
2663
2664 /* Variant working on a list of attributes. */
2665
2666 tree
oacc_replace_fn_attrib_attr(tree attribs,tree dims)2667 oacc_replace_fn_attrib_attr (tree attribs, tree dims)
2668 {
2669 tree ident = get_identifier (OACC_FN_ATTRIB);
2670
2671 /* If we happen to be present as the first attrib, drop it. */
2672 if (attribs && TREE_PURPOSE (attribs) == ident)
2673 attribs = TREE_CHAIN (attribs);
2674 return tree_cons (ident, dims, attribs);
2675 }
2676
2677 /* Variant working on a function decl. */
2678
2679 void
oacc_replace_fn_attrib(tree fn,tree dims)2680 oacc_replace_fn_attrib (tree fn, tree dims)
2681 {
2682 DECL_ATTRIBUTES (fn)
2683 = oacc_replace_fn_attrib_attr (DECL_ATTRIBUTES (fn), dims);
2684 }
2685
2686 /* Scan CLAUSES for launch dimensions and attach them to the oacc
2687 function attribute. Push any that are non-constant onto the ARGS
2688 list, along with an appropriate GOMP_LAUNCH_DIM tag. */
2689
2690 void
oacc_set_fn_attrib(tree fn,tree clauses,vec<tree> * args)2691 oacc_set_fn_attrib (tree fn, tree clauses, vec<tree> *args)
2692 {
2693 /* Must match GOMP_DIM ordering. */
2694 static const omp_clause_code ids[]
2695 = { OMP_CLAUSE_NUM_GANGS, OMP_CLAUSE_NUM_WORKERS,
2696 OMP_CLAUSE_VECTOR_LENGTH };
2697 unsigned ix;
2698 tree dims[GOMP_DIM_MAX];
2699
2700 tree attr = NULL_TREE;
2701 unsigned non_const = 0;
2702
2703 for (ix = GOMP_DIM_MAX; ix--;)
2704 {
2705 tree clause = omp_find_clause (clauses, ids[ix]);
2706 tree dim = NULL_TREE;
2707
2708 if (clause)
2709 dim = OMP_CLAUSE_EXPR (clause, ids[ix]);
2710 dims[ix] = dim;
2711 if (dim && TREE_CODE (dim) != INTEGER_CST)
2712 {
2713 dim = integer_zero_node;
2714 non_const |= GOMP_DIM_MASK (ix);
2715 }
2716 attr = tree_cons (NULL_TREE, dim, attr);
2717 }
2718
2719 oacc_replace_fn_attrib (fn, attr);
2720
2721 if (non_const)
2722 {
2723 /* Push a dynamic argument set. */
2724 args->safe_push (oacc_launch_pack (GOMP_LAUNCH_DIM,
2725 NULL_TREE, non_const));
2726 for (unsigned ix = 0; ix != GOMP_DIM_MAX; ix++)
2727 if (non_const & GOMP_DIM_MASK (ix))
2728 args->safe_push (dims[ix]);
2729 }
2730 }
2731
2732 /* Verify OpenACC routine clauses.
2733
2734 Returns 0 if FNDECL should be marked with an OpenACC 'routine' directive, 1
2735 if it has already been marked in compatible way, and -1 if incompatible.
2736 Upon returning, the chain of clauses will contain exactly one clause
2737 specifying the level of parallelism. */
2738
2739 int
oacc_verify_routine_clauses(tree fndecl,tree * clauses,location_t loc,const char * routine_str)2740 oacc_verify_routine_clauses (tree fndecl, tree *clauses, location_t loc,
2741 const char *routine_str)
2742 {
2743 tree c_level = NULL_TREE;
2744 tree c_nohost = NULL_TREE;
2745 tree c_p = NULL_TREE;
2746 for (tree c = *clauses; c; c_p = c, c = OMP_CLAUSE_CHAIN (c))
2747 switch (OMP_CLAUSE_CODE (c))
2748 {
2749 case OMP_CLAUSE_GANG:
2750 case OMP_CLAUSE_WORKER:
2751 case OMP_CLAUSE_VECTOR:
2752 case OMP_CLAUSE_SEQ:
2753 if (c_level == NULL_TREE)
2754 c_level = c;
2755 else if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_CODE (c_level))
2756 {
2757 /* This has already been diagnosed in the front ends. */
2758 /* Drop the duplicate clause. */
2759 gcc_checking_assert (c_p != NULL_TREE);
2760 OMP_CLAUSE_CHAIN (c_p) = OMP_CLAUSE_CHAIN (c);
2761 c = c_p;
2762 }
2763 else
2764 {
2765 error_at (OMP_CLAUSE_LOCATION (c),
2766 "%qs specifies a conflicting level of parallelism",
2767 omp_clause_code_name[OMP_CLAUSE_CODE (c)]);
2768 inform (OMP_CLAUSE_LOCATION (c_level),
2769 "... to the previous %qs clause here",
2770 omp_clause_code_name[OMP_CLAUSE_CODE (c_level)]);
2771 /* Drop the conflicting clause. */
2772 gcc_checking_assert (c_p != NULL_TREE);
2773 OMP_CLAUSE_CHAIN (c_p) = OMP_CLAUSE_CHAIN (c);
2774 c = c_p;
2775 }
2776 break;
2777 case OMP_CLAUSE_NOHOST:
2778 /* Don't worry about duplicate clauses here. */
2779 c_nohost = c;
2780 break;
2781 default:
2782 gcc_unreachable ();
2783 }
2784 if (c_level == NULL_TREE)
2785 {
2786 /* Default to an implicit 'seq' clause. */
2787 c_level = build_omp_clause (loc, OMP_CLAUSE_SEQ);
2788 OMP_CLAUSE_CHAIN (c_level) = *clauses;
2789 *clauses = c_level;
2790 }
2791 /* In *clauses, we now have exactly one clause specifying the level of
2792 parallelism. */
2793
2794 tree attr
2795 = lookup_attribute ("omp declare target", DECL_ATTRIBUTES (fndecl));
2796 if (attr != NULL_TREE)
2797 {
2798 /* Diagnose if "#pragma omp declare target" has also been applied. */
2799 if (TREE_VALUE (attr) == NULL_TREE)
2800 {
2801 /* See <https://gcc.gnu.org/PR93465>; the semantics of combining
2802 OpenACC and OpenMP 'target' are not clear. */
2803 error_at (loc,
2804 "cannot apply %<%s%> to %qD, which has also been"
2805 " marked with an OpenMP 'declare target' directive",
2806 routine_str, fndecl);
2807 /* Incompatible. */
2808 return -1;
2809 }
2810
2811 /* If a "#pragma acc routine" has already been applied, just verify
2812 this one for compatibility. */
2813 /* Collect previous directive's clauses. */
2814 tree c_level_p = NULL_TREE;
2815 tree c_nohost_p = NULL_TREE;
2816 for (tree c = TREE_VALUE (attr); c; c = OMP_CLAUSE_CHAIN (c))
2817 switch (OMP_CLAUSE_CODE (c))
2818 {
2819 case OMP_CLAUSE_GANG:
2820 case OMP_CLAUSE_WORKER:
2821 case OMP_CLAUSE_VECTOR:
2822 case OMP_CLAUSE_SEQ:
2823 gcc_checking_assert (c_level_p == NULL_TREE);
2824 c_level_p = c;
2825 break;
2826 case OMP_CLAUSE_NOHOST:
2827 gcc_checking_assert (c_nohost_p == NULL_TREE);
2828 c_nohost_p = c;
2829 break;
2830 default:
2831 gcc_unreachable ();
2832 }
2833 gcc_checking_assert (c_level_p != NULL_TREE);
2834 /* ..., and compare to current directive's, which we've already collected
2835 above. */
2836 tree c_diag;
2837 tree c_diag_p;
2838 /* Matching level of parallelism? */
2839 if (OMP_CLAUSE_CODE (c_level) != OMP_CLAUSE_CODE (c_level_p))
2840 {
2841 c_diag = c_level;
2842 c_diag_p = c_level_p;
2843 goto incompatible;
2844 }
2845 /* Matching 'nohost' clauses? */
2846 if ((c_nohost == NULL_TREE) != (c_nohost_p == NULL_TREE))
2847 {
2848 c_diag = c_nohost;
2849 c_diag_p = c_nohost_p;
2850 goto incompatible;
2851 }
2852 /* Compatible. */
2853 return 1;
2854
2855 incompatible:
2856 if (c_diag != NULL_TREE)
2857 error_at (OMP_CLAUSE_LOCATION (c_diag),
2858 "incompatible %qs clause when applying"
2859 " %<%s%> to %qD, which has already been"
2860 " marked with an OpenACC 'routine' directive",
2861 omp_clause_code_name[OMP_CLAUSE_CODE (c_diag)],
2862 routine_str, fndecl);
2863 else if (c_diag_p != NULL_TREE)
2864 error_at (loc,
2865 "missing %qs clause when applying"
2866 " %<%s%> to %qD, which has already been"
2867 " marked with an OpenACC 'routine' directive",
2868 omp_clause_code_name[OMP_CLAUSE_CODE (c_diag_p)],
2869 routine_str, fndecl);
2870 else
2871 gcc_unreachable ();
2872 if (c_diag_p != NULL_TREE)
2873 inform (OMP_CLAUSE_LOCATION (c_diag_p),
2874 "... with %qs clause here",
2875 omp_clause_code_name[OMP_CLAUSE_CODE (c_diag_p)]);
2876 else
2877 {
2878 /* In the front ends, we don't preserve location information for the
2879 OpenACC routine directive itself. However, that of c_level_p
2880 should be close. */
2881 location_t loc_routine = OMP_CLAUSE_LOCATION (c_level_p);
2882 inform (loc_routine, "... without %qs clause near to here",
2883 omp_clause_code_name[OMP_CLAUSE_CODE (c_diag)]);
2884 }
2885 /* Incompatible. */
2886 return -1;
2887 }
2888
2889 return 0;
2890 }
2891
2892 /* Process the OpenACC 'routine' directive clauses to generate an attribute
2893 for the level of parallelism. All dimensions have a size of zero
2894 (dynamic). TREE_PURPOSE is set to indicate whether that dimension
2895 can have a loop partitioned on it. non-zero indicates
2896 yes, zero indicates no. By construction once a non-zero has been
2897 reached, further inner dimensions must also be non-zero. We set
2898 TREE_VALUE to zero for the dimensions that may be partitioned and
2899 1 for the other ones -- if a loop is (erroneously) spawned at
2900 an outer level, we don't want to try and partition it. */
2901
2902 tree
oacc_build_routine_dims(tree clauses)2903 oacc_build_routine_dims (tree clauses)
2904 {
2905 /* Must match GOMP_DIM ordering. */
2906 static const omp_clause_code ids[]
2907 = {OMP_CLAUSE_GANG, OMP_CLAUSE_WORKER, OMP_CLAUSE_VECTOR, OMP_CLAUSE_SEQ};
2908 int ix;
2909 int level = -1;
2910
2911 for (; clauses; clauses = OMP_CLAUSE_CHAIN (clauses))
2912 for (ix = GOMP_DIM_MAX + 1; ix--;)
2913 if (OMP_CLAUSE_CODE (clauses) == ids[ix])
2914 {
2915 level = ix;
2916 break;
2917 }
2918 gcc_checking_assert (level >= 0);
2919
2920 tree dims = NULL_TREE;
2921
2922 for (ix = GOMP_DIM_MAX; ix--;)
2923 dims = tree_cons (build_int_cst (boolean_type_node, ix >= level),
2924 build_int_cst (integer_type_node, ix < level), dims);
2925
2926 return dims;
2927 }
2928
2929 /* Retrieve the oacc function attrib and return it. Non-oacc
2930 functions will return NULL. */
2931
2932 tree
oacc_get_fn_attrib(tree fn)2933 oacc_get_fn_attrib (tree fn)
2934 {
2935 return lookup_attribute (OACC_FN_ATTRIB, DECL_ATTRIBUTES (fn));
2936 }
2937
2938 /* Return true if FN is an OpenMP or OpenACC offloading function. */
2939
2940 bool
offloading_function_p(tree fn)2941 offloading_function_p (tree fn)
2942 {
2943 tree attrs = DECL_ATTRIBUTES (fn);
2944 return (lookup_attribute ("omp declare target", attrs)
2945 || lookup_attribute ("omp target entrypoint", attrs));
2946 }
2947
2948 /* Extract an oacc execution dimension from FN. FN must be an
2949 offloaded function or routine that has already had its execution
2950 dimensions lowered to the target-specific values. */
2951
2952 int
oacc_get_fn_dim_size(tree fn,int axis)2953 oacc_get_fn_dim_size (tree fn, int axis)
2954 {
2955 tree attrs = oacc_get_fn_attrib (fn);
2956
2957 gcc_assert (axis < GOMP_DIM_MAX);
2958
2959 tree dims = TREE_VALUE (attrs);
2960 while (axis--)
2961 dims = TREE_CHAIN (dims);
2962
2963 int size = TREE_INT_CST_LOW (TREE_VALUE (dims));
2964
2965 return size;
2966 }
2967
2968 /* Extract the dimension axis from an IFN_GOACC_DIM_POS or
2969 IFN_GOACC_DIM_SIZE call. */
2970
2971 int
oacc_get_ifn_dim_arg(const gimple * stmt)2972 oacc_get_ifn_dim_arg (const gimple *stmt)
2973 {
2974 gcc_checking_assert (gimple_call_internal_fn (stmt) == IFN_GOACC_DIM_SIZE
2975 || gimple_call_internal_fn (stmt) == IFN_GOACC_DIM_POS);
2976 tree arg = gimple_call_arg (stmt, 0);
2977 HOST_WIDE_INT axis = TREE_INT_CST_LOW (arg);
2978
2979 gcc_checking_assert (axis >= 0 && axis < GOMP_DIM_MAX);
2980 return (int) axis;
2981 }
2982
2983 /* Build COMPONENT_REF and set TREE_THIS_VOLATILE and TREE_READONLY on it
2984 as appropriate. */
2985
2986 tree
omp_build_component_ref(tree obj,tree field)2987 omp_build_component_ref (tree obj, tree field)
2988 {
2989 tree ret = build3 (COMPONENT_REF, TREE_TYPE (field), obj, field, NULL);
2990 if (TREE_THIS_VOLATILE (field))
2991 TREE_THIS_VOLATILE (ret) |= 1;
2992 if (TREE_READONLY (field))
2993 TREE_READONLY (ret) |= 1;
2994 return ret;
2995 }
2996
2997 #include "gt-omp-general.h"
2998