[committed] declare variant scoring
diff mbox series

Message ID 20191102090827.GY4650@tucnak
State New
Headers show
Series
  • [committed] declare variant scoring
Related show

Commit Message

Jakub Jelinek Nov. 2, 2019, 9:08 a.m. UTC
Hi!

The following patch implements the scoring, except it doesn't handle the
case when a call is nested in 188+ OpenMP constructs yet.

At least in the current standard wording, the score can be a sum of
up to 8 2^X numbers where X is quite arbitrary constant (because the
construct nesting depth can be arbitrary too) plus some user provided
integral constants with supported integral types, so if it can't be represented
in the widest_int, I think I need to find some bit above what the sum of
the user constants (and 1) needs where some bit is unset and thus addition
can't overflow into further bits and simply represent the number as
widest_int + up to 8 2^X numbers represented by those Xs, and then just
write a comparison function for that.

Regtested on x86_64-linux and i686-linux, committed to trunk.

2019-11-02  Jakub Jelinek  <jakub@redhat.com>

	* gimplify.h (omp_construct_selector_matches): Change return
	type to int, add a new SCORES argument.
	* gimplify.c (omp_construct_selector_matches): Likewise.  If
	SCORES is non-NULL, compute scores of each construct.
	* omp-general.h (omp_get_context_selector): Declare.
	* omp-general.c (omp_maybe_offloaded, omp_context_selector_matches):
	Adjust omp_construct_selector_matches callers.
	(omp_get_context_selector): New function, moved from c-family/c-omp.c.
	(omp_context_compute_score): New function.
	(omp_resolve_declare_variant): Compute scores and decide based on
	that.
c-family/
	* c-common.h (c_omp_get_context_selector): Remove.
	* c-omp.c (c_omp_get_context_selector): Moved to omp-general.c
	and renamed to omp_get_context_selector.
c/
	* c-parser.c (c_finish_omp_declare_variant): Use
	omp_get_context_selector instead of c_omp_get_context_selector.
cp/
	* decl.c (omp_declare_variant_finalize_one): Use
	omp_get_context_selector instead of c_omp_get_context_selector.
testsuite/
	* c-c++-common/gomp/declare-variant-12.c: New test.


	Jakub

Patch
diff mbox series

--- gcc/gimplify.h.jj	2019-11-01 22:19:46.253883112 +0100
+++ gcc/gimplify.h	2019-11-02 08:07:55.118710131 +0100
@@ -75,7 +75,7 @@  extern void omp_firstprivatize_variable
 extern enum gimplify_status gimplify_expr (tree *, gimple_seq *, gimple_seq *,
 					   bool (*) (tree), fallback_t);
 
-HOST_WIDE_INT omp_construct_selector_matches (enum tree_code *, int);
+int omp_construct_selector_matches (enum tree_code *, int, int *);
 
 extern void gimplify_type_sizes (tree, gimple_seq *);
 extern void gimplify_one_sizepos (tree *, gimple_seq *);
--- gcc/gimplify.c.jj	2019-11-01 22:19:46.222883585 +0100
+++ gcc/gimplify.c	2019-11-02 08:07:55.123710053 +0100
@@ -10381,14 +10381,24 @@  gimplify_adjust_omp_clauses (gimple_seq
 
 /* Return 0 if CONSTRUCTS selectors don't match the OpenMP context,
    -1 if unknown yet (simd is involved, won't be known until vectorization)
-   and positive number if they do, the number is then the number of constructs
-   in the OpenMP context.  */
+   and 1 if they do.  If SCORES is non-NULL, it should point to an array
+   of at least 2*NCONSTRUCTS+2 ints, and will be filled with the positions
+   of the CONSTRUCTS (position -1 if it will never match) followed by
+   number of constructs in the OpenMP context construct trait.  If the
+   score depends on whether it will be in a declare simd clone or not,
+   the function returns 2 and there will be two sets of the scores, the first
+   one for the case that it is not in a declare simd clone, the other
+   that it is in a declare simd clone.  */
 
-HOST_WIDE_INT
-omp_construct_selector_matches (enum tree_code *constructs, int nconstructs)
+int
+omp_construct_selector_matches (enum tree_code *constructs, int nconstructs,
+				int *scores)
 {
   int matched = 0, cnt = 0;
   bool simd_seen = false;
+  bool target_seen = false;
+  int declare_simd_cnt = -1;
+  auto_vec<enum tree_code, 16> codes;
   for (struct gimplify_omp_ctx *ctx = gimplify_omp_ctxp; ctx;)
     {
       if (((ctx->region_type & ORT_PARALLEL) && ctx->code == OMP_PARALLEL)
@@ -10401,7 +10411,9 @@  omp_construct_selector_matches (enum tre
 	      && !omp_find_clause (ctx->clauses, OMP_CLAUSE_BIND)))
 	{
 	  ++cnt;
-	  if (matched < nconstructs && ctx->code == constructs[matched])
+	  if (scores)
+	    codes.safe_push (ctx->code);
+	  else if (matched < nconstructs && ctx->code == constructs[matched])
 	    {
 	      if (ctx->code == OMP_SIMD)
 		{
@@ -10412,7 +10424,12 @@  omp_construct_selector_matches (enum tre
 	      ++matched;
 	    }
 	  if (ctx->code == OMP_TARGET)
-	    return matched < nconstructs ? 0 : simd_seen ? -1 : cnt;
+	    {
+	      if (scores == NULL)
+		return matched < nconstructs ? 0 : simd_seen ? -1 : 1;
+	      target_seen = true;
+	      break;
+	    }
 	}
       else if (ctx->region_type == ORT_WORKSHARE
 	       && ctx->code == OMP_LOOP
@@ -10424,31 +10441,40 @@  omp_construct_selector_matches (enum tre
 	ctx = ctx->outer_context->outer_context;
       ctx = ctx->outer_context;
     }
-  if (cnt == 0
-      && constructs[0] == OMP_SIMD
+  if (!target_seen
       && lookup_attribute ("omp declare simd",
 			   DECL_ATTRIBUTES (current_function_decl)))
     {
       /* Declare simd is a maybe case, it is supposed to be added only to the
 	 omp-simd-clone.c added clones and not to the base function.  */
-      gcc_assert (matched == 0);
-      ++cnt;
-      simd_seen = true;
-      if (++matched == nconstructs)
-	return -1;
+      declare_simd_cnt = cnt++;
+      if (scores)
+	codes.safe_push (OMP_SIMD);
+      else if (cnt == 0
+	       && constructs[0] == OMP_SIMD)
+	{
+	  gcc_assert (matched == 0);
+	  simd_seen = true;
+	  if (++matched == nconstructs)
+	    return -1;
+	}
     }
   if (tree attr = lookup_attribute ("omp declare variant variant",
 				    DECL_ATTRIBUTES (current_function_decl)))
     {
       enum tree_code variant_constructs[5];
-      int variant_nconstructs
-	= omp_constructor_traits_to_codes (TREE_VALUE (attr),
-					   variant_constructs);
+      int variant_nconstructs = 0;
+      if (!target_seen)
+	variant_nconstructs
+	  = omp_constructor_traits_to_codes (TREE_VALUE (attr),
+					     variant_constructs);
       for (int i = 0; i < variant_nconstructs; i++)
 	{
 	  ++cnt;
-	  if (matched < nconstructs
-	      && variant_constructs[i] == constructs[matched])
+	  if (scores)
+	    codes.safe_push (variant_constructs[i]);
+	  else if (matched < nconstructs
+		   && variant_constructs[i] == constructs[matched])
 	    {
 	      if (variant_constructs[i] == OMP_SIMD)
 		{
@@ -10460,15 +10486,38 @@  omp_construct_selector_matches (enum tre
 	    }
 	}
     }
-  if (lookup_attribute ("omp declare target block",
-			DECL_ATTRIBUTES (current_function_decl)))
+  if (!target_seen
+      && lookup_attribute ("omp declare target block",
+			   DECL_ATTRIBUTES (current_function_decl)))
     {
-      ++cnt;
-      if (matched < nconstructs && constructs[matched] == OMP_TARGET)
+      if (scores)
+	codes.safe_push (OMP_TARGET);
+      else if (matched < nconstructs && constructs[matched] == OMP_TARGET)
 	++matched;
     }
+  if (scores)
+    {
+      for (int pass = 0; pass < (declare_simd_cnt == -1 ? 1 : 2); pass++)
+	{
+	  int j = codes.length () - 1;
+	  for (int i = nconstructs - 1; i >= 0; i--)
+	    {
+	      while (j >= 0
+		     && (pass != 0 || declare_simd_cnt != j)
+		     && constructs[i] != codes[j])
+		--j;
+	      if (pass == 0 && declare_simd_cnt != -1 && j > declare_simd_cnt)
+		*scores++ = j - 1;
+	      else
+		*scores++ = j;
+	    }
+	  *scores++ = ((pass == 0 && declare_simd_cnt != -1)
+		       ? codes.length () - 1 : codes.length ());
+	}
+      return declare_simd_cnt == -1 ? 1 : 2;
+    }
   if (matched == nconstructs)
-    return simd_seen ? -1 : cnt;
+    return simd_seen ? -1 : 1;
   return 0;
 }
 
--- gcc/omp-general.h.jj	2019-11-02 00:29:22.967517001 +0100
+++ gcc/omp-general.h	2019-11-02 08:07:55.124710037 +0100
@@ -87,6 +87,7 @@  extern int omp_max_simt_vf (void);
 extern int omp_constructor_traits_to_codes (tree, enum tree_code *);
 extern int omp_context_selector_matches (tree);
 extern int omp_context_selector_set_compare (const char *, tree, tree);
+extern tree omp_get_context_selector (tree, const char *, const char *);
 extern tree omp_resolve_declare_variant (tree);
 extern tree oacc_launch_pack (unsigned code, tree device, unsigned op);
 extern tree oacc_replace_fn_attrib_attr (tree attribs, tree dims);
--- gcc/omp-general.c.jj	2019-11-02 00:29:22.968516986 +0100
+++ gcc/omp-general.c	2019-11-02 08:07:55.124710037 +0100
@@ -639,7 +639,7 @@  omp_maybe_offloaded (void)
   if (cfun && (cfun->curr_properties & PROP_gimple_any) == 0)
     {
       enum tree_code construct = OMP_TARGET;
-      if (omp_construct_selector_matches (&construct, 1))
+      if (omp_construct_selector_matches (&construct, 1, NULL))
 	return true;
     }
   return false;
@@ -677,8 +677,8 @@  omp_context_selector_matches (tree ctx)
 	  enum tree_code constructs[5];
 	  int nconstructs
 	    = omp_constructor_traits_to_codes (TREE_VALUE (t1), constructs);
-	  HOST_WIDE_INT r
-	    = omp_construct_selector_matches (constructs, nconstructs);
+	  int r = omp_construct_selector_matches (constructs, nconstructs,
+						  NULL);
 	  if (r == 0)
 	    return 0;
 	  if (r == -1)
@@ -1261,13 +1261,93 @@  omp_context_selector_compare (tree ctx1,
   return swapped ? -ret : ret;
 }
 
+/* From context selector CTX, return trait-selector with name SEL in
+   trait-selector-set with name SET if any, or NULL_TREE if not found.
+   If SEL is NULL, return the list of trait-selectors in SET.  */
+
+tree
+omp_get_context_selector (tree ctx, const char *set, const char *sel)
+{
+  tree setid = get_identifier (set);
+  tree selid = sel ? get_identifier (sel) : NULL_TREE;
+  for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1))
+    if (TREE_PURPOSE (t1) == setid)
+      {
+	if (sel == NULL)
+	  return TREE_VALUE (t1);
+	for (tree t2 = TREE_VALUE (t1); t2; t2 = TREE_CHAIN (t2))
+	  if (TREE_PURPOSE (t2) == selid)
+	    return t2;
+      }
+  return NULL_TREE;
+}
+
+/* Compute *SCORE for context selector CTX.  Return true if the score
+   would be different depending on whether it is a declare simd clone or
+   not.  DECLARE_SIMD should be true for the case when it would be
+   a declare simd clone.  */
+
+static bool
+omp_context_compute_score (tree ctx, widest_int *score, bool declare_simd)
+{
+  tree construct = omp_get_context_selector (ctx, "construct", NULL);
+  bool has_kind = omp_get_context_selector (ctx, "device", "kind");
+  bool has_arch = omp_get_context_selector (ctx, "device", "arch");
+  bool has_isa = omp_get_context_selector (ctx, "device", "isa");
+  bool ret = false;
+  *score = 1;
+  for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1))
+    for (tree t2 = TREE_VALUE (t1); t2; t2 = TREE_CHAIN (t2))
+      if (tree t3 = TREE_VALUE (t2))
+	if (TREE_PURPOSE (t3)
+	    && strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t3)), " score") == 0
+	    && TREE_CODE (TREE_VALUE (t3)) == INTEGER_CST)
+	  *score += wi::to_widest (TREE_VALUE (t3));
+  if (construct || has_kind || has_arch || has_isa)
+    {
+      int scores[12];
+      enum tree_code constructs[5];
+      int nconstructs = 0;
+      if (construct)
+	nconstructs = omp_constructor_traits_to_codes (construct, constructs);
+      if (omp_construct_selector_matches (constructs, nconstructs, scores)
+	  == 2)
+	ret = true;
+      int b = declare_simd ? nconstructs + 1 : 0;
+      if (scores[b + nconstructs] + 4U < score->get_precision ())
+	{
+	  for (int n = 0; n < nconstructs; ++n)
+	    {
+	      if (scores[b + n] < 0)
+		{
+		  *score = 0;
+		  return ret;
+		}
+	      *score += wi::shifted_mask <widest_int> (scores[b + n], 1, false);
+	    }
+	  if (has_kind)
+	    *score += wi::shifted_mask <widest_int> (scores[b + nconstructs],
+						     1, false);
+	  if (has_arch)
+	    *score += wi::shifted_mask <widest_int> (scores[b + nconstructs] + 1,
+						     1, false);
+	  if (has_isa)
+	    *score += wi::shifted_mask <widest_int> (scores[b + nconstructs] + 2,
+						     1, false);
+	}
+      else /* FIXME: Implement this.  */
+	gcc_unreachable ();
+    }
+  return ret;
+}
+
 /* Try to resolve declare variant, return the variant decl if it should
    be used instead of base, or base otherwise.  */
 
 tree
 omp_resolve_declare_variant (tree base)
 {
-  tree variant = NULL_TREE;
+  tree variant1 = NULL_TREE, variant2 = NULL_TREE;
   auto_vec <tree, 16> variants;
   for (tree attr = DECL_ATTRIBUTES (base); attr; attr = TREE_CHAIN (attr))
     {
@@ -1319,16 +1399,56 @@  omp_resolve_declare_variant (tree base)
 		variants[j] = NULL_TREE;
 	    }
       }
-  /* FIXME: Scoring not implemented yet, so just resolve it
-     if there is a single variant left.  */
+  widest_int max_score1 = 0;
+  widest_int max_score2 = 0;
+  bool first = true;
   FOR_EACH_VEC_ELT (variants, i, attr1)
     if (attr1)
       {
-	if (variant)
-	  return base;
-	variant = TREE_PURPOSE (TREE_VALUE (attr1));
+	if (variant1)
+	  {
+	    widest_int score1;
+	    widest_int score2;
+	    bool need_two;
+	    tree ctx;
+	    if (first)
+	      {
+		first = false;
+		ctx = TREE_VALUE (TREE_VALUE (variant1));
+		need_two = omp_context_compute_score (ctx, &max_score1, false);
+		if (need_two)
+		  omp_context_compute_score (ctx, &max_score2, true);
+		else
+		  max_score2 = max_score1;
+	      }
+	    ctx = TREE_VALUE (TREE_VALUE (attr1));
+	    need_two = omp_context_compute_score (ctx, &score1, false);
+	    if (need_two)
+	      omp_context_compute_score (ctx, &score2, true);
+	    else
+	      score2 = score1;
+	    if (score1 > max_score1)
+	      {
+		max_score1 = score1;
+		variant1 = attr1;
+	      }
+	    if (score2 > max_score2)
+	      {
+		max_score2 = score2;
+		variant2 = attr1;
+	      }
+	  }
+	else
+	  {
+	    variant1 = attr1;
+	    variant2 = attr1;
+	  }
       }
-  return variant ? variant : base;
+  /* If there is a disagreement on which variant has the highest score
+     depending on whether it will be in a declare simd clone or not,
+     punt for now and defer until after IPA where we will know that.  */
+  return ((variant1 && variant1 == variant2)
+	  ? TREE_PURPOSE (TREE_VALUE (variant1)) : base);
 }
 
 
--- gcc/c-family/c-common.h.jj	2019-11-02 00:26:48.192858544 +0100
+++ gcc/c-family/c-common.h	2019-11-02 08:07:55.124710037 +0100
@@ -1193,7 +1193,6 @@  extern void c_omp_declare_simd_clauses_t
 extern bool c_omp_predefined_variable (tree);
 extern enum omp_clause_default_kind c_omp_predetermined_sharing (tree);
 extern tree c_omp_check_context_selector (location_t, tree);
-extern tree c_omp_get_context_selector (tree, const char *, const char *);
 extern void c_omp_mark_declare_variant (location_t, tree, tree);
 
 /* Return next tree in the chain for chain_next walking of tree nodes.  */
--- gcc/c-family/c-omp.c.jj	2019-11-02 00:29:22.968516986 +0100
+++ gcc/c-family/c-omp.c	2019-11-02 08:07:55.124710037 +0100
@@ -2237,27 +2237,6 @@  c_omp_check_context_selector (location_t
   return ctx;
 }
 
-/* From context selector CTX, return trait-selector with name SEL in
-   trait-selector-set with name SET if any, or NULL_TREE if not found.
-   If SEL is NULL, return the list of trait-selectors in SET.  */
-
-tree
-c_omp_get_context_selector (tree ctx, const char *set, const char *sel)
-{
-  tree setid = get_identifier (set);
-  tree selid = sel ? get_identifier (sel) : NULL_TREE;
-  for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1))
-    if (TREE_PURPOSE (t1) == setid)
-      {
-	if (sel == NULL)
-	  return TREE_VALUE (t1);
-	for (tree t2 = TREE_VALUE (t1); t2; t2 = TREE_CHAIN (t2))
-	  if (TREE_PURPOSE (t2) == selid)
-	    return t2;
-      }
-  return NULL_TREE;
-}
-
 /* Register VARIANT as variant of some base function marked with
    #pragma omp declare variant.  CONSTRUCT is corresponding construct
    selector set.  */
--- gcc/c/c-parser.c.jj	2019-11-01 22:19:46.639877219 +0100
+++ gcc/c/c-parser.c	2019-11-02 08:07:55.122710069 +0100
@@ -19465,8 +19465,7 @@  c_finish_omp_declare_variant (c_parser *
 	  error_at (token->location, "variant %qD is not a function", variant);
 	  variant = error_mark_node;
 	}
-      else if (c_omp_get_context_selector (ctx, "construct", "simd")
-	       == NULL_TREE
+      else if (omp_get_context_selector (ctx, "construct", "simd") == NULL_TREE
 	       && !comptypes (TREE_TYPE (fndecl), TREE_TYPE (variant)))
 	{
 	  error_at (token->location, "variant %qD and base %qD have "
@@ -19487,7 +19486,7 @@  c_finish_omp_declare_variant (c_parser *
       if (variant != error_mark_node)
 	{
 	  C_DECL_USED (variant) = 1;
-	  tree construct = c_omp_get_context_selector (ctx, "construct", NULL);
+	  tree construct = omp_get_context_selector (ctx, "construct", NULL);
 	  c_omp_mark_declare_variant (match_loc, variant, construct);
 	  if (omp_context_selector_matches (ctx))
 	    {
--- gcc/cp/decl.c.jj	2019-11-02 00:26:48.964846870 +0100
+++ gcc/cp/decl.c	2019-11-02 08:07:55.118710131 +0100
@@ -7103,7 +7103,7 @@  omp_declare_variant_finalize_one (tree d
 	       DECL_ARGUMENTS (decl), NULL);
 
   tree ctx = TREE_VALUE (TREE_VALUE (attr));
-  tree simd = c_omp_get_context_selector (ctx, "construct", "simd");
+  tree simd = omp_get_context_selector (ctx, "construct", "simd");
   if (simd)
     {
       TREE_VALUE (simd)
@@ -7202,7 +7202,7 @@  omp_declare_variant_finalize_one (tree d
 	}
       else
 	{
-	  tree construct = c_omp_get_context_selector (ctx, "construct", NULL);
+	  tree construct = omp_get_context_selector (ctx, "construct", NULL);
 	  c_omp_mark_declare_variant (match_loc, variant, construct);
 	  if (!omp_context_selector_matches (ctx))
 	    return true;
--- gcc/testsuite/c-c++-common/gomp/declare-variant-12.c.jj	2019-11-02 08:16:43.455504902 +0100
+++ gcc/testsuite/c-c++-common/gomp/declare-variant-12.c	2019-11-02 09:22:12.718908955 +0100
@@ -0,0 +1,88 @@ 
+/* { dg-do compile } */
+/* { dg-additional-options "-foffload=disable -fdump-tree-gimple" } */
+/* { dg-additional-options "-mavx512bw -mavx512vl" { target { i?86-*-* x86_64-*-* } } } */
+
+#pragma omp requires atomic_default_mem_order(seq_cst)
+void f01 (void);
+void f02 (void);
+void f03 (void);
+#pragma omp declare variant (f01) match (device={isa(avx512f,avx512vl)}) /* 16 */
+#pragma omp declare variant (f02) match (implementation={vendor(score(15):gnu)})
+#pragma omp declare variant (f03) match (user={condition(score(11):1)})
+void f04 (void);
+void f05 (void);
+void f06 (void);
+void f07 (void);
+#pragma omp declare variant (f05) match (device={isa(avx512f,avx512vl)}) /* 16 */
+#pragma omp declare variant (f06) match (implementation={vendor(score(15):gnu)})
+#pragma omp declare variant (f07) match (user={condition(score(17):1)})
+void f08 (void);
+void f09 (void);
+void f10 (void);
+void f11 (void);
+void f12 (void);
+#pragma omp declare variant (f09) match (device={arch(x86_64)},user={condition(score(65):1)}) /* 64+65 */
+#pragma omp declare variant (f10) match (implementation={vendor(score(127):gnu)})
+#pragma omp declare variant (f11) match (device={isa(ssse3)}) /* 128 */
+#pragma omp declare variant (f12) match (implementation={atomic_default_mem_order(score(126):seq_cst)})
+void f13 (void);
+void f14 (void);
+void f15 (void);
+void f16 (void);
+#pragma omp declare variant (f14) match (construct={teams,parallel,for}) /* 16+8+4 */
+#pragma omp declare variant (f15) match (construct={parallel},user={condition(score(19):1)}) /* 8+19 */
+#pragma omp declare variant (f16) match (implementation={atomic_default_mem_order(score(27):seq_cst)})
+void f17 (void);
+void f18 (void);
+void f19 (void);
+void f20 (void);
+#pragma omp declare variant (f18) match (construct={teams,parallel,for}) /* 16+8+4 */
+#pragma omp declare variant (f19) match (construct={for},user={condition(score(25):1)}) /* 4+25 */
+#pragma omp declare variant (f20) match (implementation={atomic_default_mem_order(score(28):seq_cst)})
+void f21 (void);
+void f22 (void);
+void f23 (void);
+void f24 (void);
+#pragma omp declare variant (f22) match (construct={parallel,for}) /* 2+1 */
+#pragma omp declare variant (f23) match (construct={for}) /* 0 */
+#pragma omp declare variant (f24) match (implementation={atomic_default_mem_order(score(2):seq_cst)})
+void f25 (void);
+void f26 (void);
+void f27 (void);
+void f28 (void);
+#pragma omp declare variant (f26) match (construct={parallel,for}) /* 2+1 */
+#pragma omp declare variant (f27) match (construct={for},user={condition(1)}) /* 4 */
+#pragma omp declare variant (f28) match (implementation={atomic_default_mem_order(score(3):seq_cst)})
+void f29 (void);
+
+void
+test1 (void)
+{
+  int i, j;
+  #pragma omp parallel for	/* 2 constructs in OpenMP context, isa has score 2^4.  */
+  for (i = 0; i < 1; i++)
+    f04 ();	/* { dg-final { scan-tree-dump-times "f01 \\\(\\\);" 1 "gimple" { target i?86-*-* x86_64-*-* } } } */
+		/* { dg-final { scan-tree-dump-times "f02 \\\(\\\);" 1 "gimple" { target { ! { i?86-*-* x86_64-*-* } } } } } */
+  #pragma omp target teams	/* 2 constructs in OpenMP context, isa has score 2^4.  */
+  f08 ();	/* { dg-final { scan-tree-dump-times "f07 \\\(\\\);" 1 "gimple" } } */
+  #pragma omp teams
+  #pragma omp parallel for
+  for (i = 0; i < 1; i++)
+    #pragma omp parallel for	/* 5 constructs in OpenMP context, arch is 2^6, isa 2^7.  */
+    for (j = 0; j < 1; j++)
+      {
+	f13 ();	/* { dg-final { scan-tree-dump-times "f09 \\\(\\\);" 1 "gimple" { target { { i?86-*-* x86_64-*-* } && lp64 } } } } */
+		/* { dg-final { scan-tree-dump-times "f11 \\\(\\\);" 1 "gimple" { target { { i?86-*-* x86_64-*-* } && { ! lp64 } } } } } */
+		/* { dg-final { scan-tree-dump-times "f10 \\\(\\\);" 1 "gimple" { target { ! { i?86-*-* x86_64-*-* } } } } } */
+	f17 ();	/* { dg-final { scan-tree-dump-times "f14 \\\(\\\);" 1 "gimple" } } */
+	f21 ();	/* { dg-final { scan-tree-dump-times "f19 \\\(\\\);" 1 "gimple" } } */
+      }
+  #pragma omp for
+  for (i = 0; i < 1; i++)
+    #pragma omp parallel for
+    for (j = 0; j < 1; j++)
+      {
+	f25 ();	/* { dg-final { scan-tree-dump-times "f22 \\\(\\\);" 1 "gimple" } } */
+	f29 ();	/* { dg-final { scan-tree-dump-times "f27 \\\(\\\);" 1 "gimple" } } */
+      }
+}