c-common.h (c_omp_check_context_selector, [...]): Declare.

c-family/
	* c-common.h (c_omp_check_context_selector,
	c_omp_get_context_selector): Declare.
	* c-omp.c (c_omp_declare_simd_clauses_to_numbers): Fix spelling
	in diagnostic message.
	(c_omp_check_context_selector, c_omp_get_context_selector): New
	functions.
	* c-attribs.c (c_common_attribute_table): Add "omp declare variant"
	attribute.
	(handle_omp_declare_variant_attribute): New function.
c/
	* c-parser.c (c_parser_omp_all_clauses): Add NESTED_P argument, if
	true, terminate processing on closing paren and don't skip to end of
	pragma line.
	(c_parser_omp_declare_simd): Handle also declare variant.
	(omp_construct_selectors, omp_device_selectors,
	omp_implementation_selectors, omp_user_selectors): New variables.
	(c_parser_omp_context_selector,
	c_parser_omp_context_selector_specification,
	c_finish_omp_declare_variant): New functions.
	(c_finish_omp_declare_simd): Handle both declare simd and
	declare variant.
	(c_parser_omp_declare): Handle declare variant.
cp/
	* parser.h (struct cp_omp_declare_simd_data): Add variant_p member.
	* parser.c (cp_ensure_no_omp_declare_simd): Handle both declare simd
	and declare variant.
	(cp_parser_oacc_all_clauses): Formatting fix.
	(cp_parser_omp_all_clauses): Add NESTED_P argument, if true, terminate
	processing on closing paren and don't skip to end of pragma line.
	(cp_parser_omp_declare_simd): Add VARIANT_P argument.  Handle also
	declare variant.
	(omp_construct_selectors, omp_device_selectors,
	omp_implementation_selectors, omp_user_selectors): New variables.
	(cp_parser_omp_context_selector,
	cp_parser_omp_context_selector_specification,
	cp_finish_omp_declare_variant): New functions.
	(cp_parser_late_parsing_omp_declare_simd): Handle also declare variant.
	(cp_parser_omp_declare): Handle declare variant.
testsuite/
	* c-c++-common/gomp/declare-variant-1.c: New test.
	* c-c++-common/gomp/declare-variant-2.c: New test.
	* c-c++-common/gomp/declare-variant-3.c: New test.
	* g++.dg/gomp/this-1.C: Adjust for diagnostic message spelling fix.
	* gcc.dg/gomp/declare-variant-1.c: New test.
	* gcc.dg/gomp/declare-variant-2.c: New test.

From-SVN: r276789

16 files changed, 1570 insertions, 71 deletions

diff --git a/gcc/c-family/ChangeLog b/gcc/c-family/ChangeLog
index 398ba6d..598fea4 100644
--- a/gcc/c-family/ChangeLog
+++ b/gcc/c-family/ChangeLog
@@ -1,3 +1,15 @@
+2019-10-10  Jakub Jelinek  <jakub@redhat.com>
+
+	* c-common.h (c_omp_check_context_selector,
+	c_omp_get_context_selector): Declare.
+	* c-omp.c (c_omp_declare_simd_clauses_to_numbers): Fix spelling
+	in diagnostic message.
+	(c_omp_check_context_selector, c_omp_get_context_selector): New
+	functions.
+	* c-attribs.c (c_common_attribute_table): Add "omp declare variant"
+	attribute.
+	(handle_omp_declare_variant_attribute): New function.
+
 2019-10-09  Martin Sebor  <msebor@redhat.com>
 
 	PR tree-optimization/90879
diff --git a/gcc/c-family/c-attribs.c b/gcc/c-family/c-attribs.c
index 6500b99..917d483 100644
--- a/gcc/c-family/c-attribs.c
+++ b/gcc/c-family/c-attribs.c
@@ -140,6 +140,8 @@ static tree handle_warn_unused_attribute (tree *, tree, tree, int, bool *);
 static tree handle_returns_nonnull_attribute (tree *, tree, tree, int, bool *);
 static tree handle_omp_declare_simd_attribute (tree *, tree, tree, int,
 					       bool *);
+static tree handle_omp_declare_variant_attribute (tree *, tree, tree, int,
+						  bool *);
 static tree handle_simd_attribute (tree *, tree, tree, int, bool *);
 static tree handle_omp_declare_target_attribute (tree *, tree, tree, int,
 						 bool *);
@@ -442,6 +444,8 @@ const struct attribute_spec c_common_attribute_table[] =
 			      handle_returns_nonnull_attribute, NULL },
   { "omp declare simd",       0, -1, true,  false, false, false,
 			      handle_omp_declare_simd_attribute, NULL },
+  { "omp declare variant",    0, -1, true,  false, false, false,
+			      handle_omp_declare_variant_attribute, NULL },
   { "simd",		      0, 1, true,  false, false, false,
 			      handle_simd_attribute, NULL },
   { "omp declare target",     0, -1, true, false, false, false,
@@ -3064,6 +3068,15 @@ handle_omp_declare_simd_attribute (tree *, tree, tree, int, bool *)
   return NULL_TREE;
 }
 
+/* Handle an "omp declare variant" attribute; arguments as in
+   struct attribute_spec.handler.  */
+
+static tree
+handle_omp_declare_variant_attribute (tree *, tree, tree, int, bool *)
+{
+  return NULL_TREE;
+}
+
 /* Handle a "simd" attribute.  */
 
 static tree
diff --git a/gcc/c-family/c-common.h b/gcc/c-family/c-common.h
index 1e13aaa..eabe689 100644
--- a/gcc/c-family/c-common.h
+++ b/gcc/c-family/c-common.h
@@ -1189,6 +1189,8 @@ extern tree c_omp_declare_simd_clauses_to_numbers (tree, tree);
 extern void c_omp_declare_simd_clauses_to_decls (tree, tree);
 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 *);
 
 /* Return next tree in the chain for chain_next walking of tree nodes.  */
 static inline tree
diff --git a/gcc/c-family/c-omp.c b/gcc/c-family/c-omp.c
index 0048289..5426256 100644
--- a/gcc/c-family/c-omp.c
+++ b/gcc/c-family/c-omp.c
@@ -2011,7 +2011,7 @@ c_omp_declare_simd_clauses_to_numbers (tree parms, tree clauses)
 	  if (arg == NULL_TREE)
 	    {
 	      error_at (OMP_CLAUSE_LOCATION (c),
-			"%qD is not an function argument", decl);
+			"%qD is not a function argument", decl);
 	      continue;
 	    }
 	  OMP_CLAUSE_DECL (c) = build_int_cst (integer_type_node, idx);
@@ -2026,7 +2026,7 @@ c_omp_declare_simd_clauses_to_numbers (tree parms, tree clauses)
 	      if (arg == NULL_TREE)
 		{
 		  error_at (OMP_CLAUSE_LOCATION (c),
-			    "%qD is not an function argument", decl);
+			    "%qD is not a function argument", decl);
 		  continue;
 		}
 	      OMP_CLAUSE_LINEAR_STEP (c)
@@ -2120,3 +2120,133 @@ c_omp_predetermined_sharing (tree decl)
 
   return OMP_CLAUSE_DEFAULT_UNSPECIFIED;
 }
+
+/* Diagnose errors in an OpenMP context selector, return CTX if
+   it is correct or error_mark_node otherwise.  */
+
+tree
+c_omp_check_context_selector (location_t loc, tree ctx)
+{
+  /* Each trait-set-selector-name can only be specified once.
+     There are just 4 set names.  */
+  for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1))
+    for (tree t2 = TREE_CHAIN (t1); t2; t2 = TREE_CHAIN (t2))
+      if (TREE_PURPOSE (t1) == TREE_PURPOSE (t2))
+	{
+	  error_at (loc, "selector set %qs specified more than once",
+	  	    IDENTIFIER_POINTER (TREE_PURPOSE (t1)));
+	  return error_mark_node;
+	}
+  for (tree t = ctx; t; t = TREE_CHAIN (t))
+    {
+      /* Each trait-selector-name can only be specified once.  */
+      if (list_length (TREE_VALUE (t)) < 5)
+	{
+	  for (tree t1 = TREE_VALUE (t); t1; t1 = TREE_CHAIN (t1))
+	    for (tree t2 = TREE_CHAIN (t1); t2; t2 = TREE_CHAIN (t2))
+	      if (TREE_PURPOSE (t1) == TREE_PURPOSE (t2))
+		{
+		  error_at (loc,
+			    "selector %qs specified more than once in set %qs",
+			    IDENTIFIER_POINTER (TREE_PURPOSE (t1)),
+			    IDENTIFIER_POINTER (TREE_PURPOSE (t)));
+		  return error_mark_node;
+		}
+	}
+      else
+	{
+	  hash_set<tree> pset;
+	  for (tree t1 = TREE_VALUE (t); t1; t1 = TREE_CHAIN (t1))
+	    if (pset.add (TREE_PURPOSE (t1)))
+	      {
+		error_at (loc,
+			  "selector %qs specified more than once in set %qs",
+			  IDENTIFIER_POINTER (TREE_PURPOSE (t1)),
+			  IDENTIFIER_POINTER (TREE_PURPOSE (t)));
+		return error_mark_node;
+	      }
+	}
+
+      static const char *const kind[] = {
+	"host", "nohost", "cpu", "gpu", "fpga", "any", NULL };
+      static const char *const vendor[] = {
+	"amd", "arm", "bsc", "cray", "fujitsu", "gnu", "ibm", "intel",
+	"llvm", "pgi", "ti", "unknown", NULL };
+      static const char *const extension[] = { NULL };
+      static const char *const atomic_default_mem_order[] = {
+	"seq_cst", "relaxed", "acq_rel", NULL };
+      struct known_properties { const char *set; const char *selector;
+				const char *const *props; };
+      known_properties props[] = {
+	{ "device", "kind", kind },
+	{ "implementation", "vendor", vendor },
+	{ "implementation", "extension", extension },
+	{ "implementation", "atomic_default_mem_order",
+	  atomic_default_mem_order } };
+      for (tree t1 = TREE_VALUE (t); t1; t1 = TREE_CHAIN (t1))
+	for (unsigned i = 0; i < ARRAY_SIZE (props); i++)
+	  if (!strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t1)),
+					   props[i].selector)
+	      && !strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t)),
+					      props[i].set))
+	    for (tree t2 = TREE_VALUE (t1); t2; t2 = TREE_CHAIN (t2))
+	      for (unsigned j = 0; ; j++)
+		{
+		  if (props[i].props[j] == NULL)
+		    {
+		      if (!strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t2)),
+				   " score"))
+			break;
+		      if (props[i].props == atomic_default_mem_order)
+			{
+			  error_at (loc,
+				    "incorrect property %qs of %qs selector",
+				    IDENTIFIER_POINTER (TREE_PURPOSE (t2)),
+				    "atomic_default_mem_order");
+			  return error_mark_node;
+			}
+		      else
+			warning_at (loc, 0,
+				    "unknown property %qs of %qs selector",
+				    IDENTIFIER_POINTER (TREE_PURPOSE (t2)),
+				    props[i].selector);
+		      break;
+		    }
+		  else if (!strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t2)),
+				    props[i].props[j]))
+		    {
+		      if (props[i].props == atomic_default_mem_order
+			  && t2 != TREE_VALUE (t1))
+			{
+			  tree t3 = TREE_VALUE (t1);
+			  if (!strcmp (IDENTIFIER_POINTER (TREE_PURPOSE (t3)),
+				       " score")
+			      && t2 == TREE_CHAIN (TREE_VALUE (t1)))
+			    break;
+			  error_at (loc,
+				    "%qs selector must have a single property",
+				    "atomic_default_mem_order");
+			  return error_mark_node;
+			}
+		      break;
+		    }
+		}
+    }
+  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.  */
+
+tree
+c_omp_get_context_selector (tree ctx, const char *set, const char *sel)
+{
+  tree setid = get_identifier (set);
+  tree selid = get_identifier (sel);
+  for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1))
+    if (TREE_PURPOSE (t1) == setid)
+      for (tree t2 = TREE_VALUE (t1); t2; t2 = TREE_CHAIN (t2))
+	if (TREE_PURPOSE (t2) == selid)
+	  return t2;
+  return NULL_TREE;
+}
diff --git a/gcc/c/ChangeLog b/gcc/c/ChangeLog
index 0f263f7..fa1001c 100644
--- a/gcc/c/ChangeLog
+++ b/gcc/c/ChangeLog
@@ -1,3 +1,18 @@
+2019-10-10  Jakub Jelinek  <jakub@redhat.com>
+
+	* c-parser.c (c_parser_omp_all_clauses): Add NESTED_P argument, if
+	true, terminate processing on closing paren and don't skip to end of
+	pragma line.
+	(c_parser_omp_declare_simd): Handle also declare variant.
+	(omp_construct_selectors, omp_device_selectors,
+	omp_implementation_selectors, omp_user_selectors): New variables.
+	(c_parser_omp_context_selector,
+	c_parser_omp_context_selector_specification,
+	c_finish_omp_declare_variant): New functions.
+	(c_finish_omp_declare_simd): Handle both declare simd and
+	declare variant.
+	(c_parser_omp_declare): Handle declare variant.
+
 2019-10-02  Joseph Myers  <joseph@codesourcery.com>
 
 	* c-parser.c (c_parser_asm_statement): Handle CPP_SCOPE like two
diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c
index 6957297..2daaee8 100644
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -15213,11 +15213,15 @@ c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask,
 }
 
 /* Parse all OpenMP clauses.  The set clauses allowed by the directive
-   is a bitmask in MASK.  Return the list of clauses found.  */
+   is a bitmask in MASK.  Return the list of clauses found.
+   FINISH_P set if c_finish_omp_clauses should be called.
+   NESTED_P set if clauses should be terminated by closing paren instead
+   of end of pragma.  */
 
 static tree
 c_parser_omp_all_clauses (c_parser *parser, omp_clause_mask mask,
-			  const char *where, bool finish_p = true)
+			  const char *where, bool finish_p = true,
+			  bool nested_p = false)
 {
   tree clauses = NULL;
   bool first = true;
@@ -15229,6 +15233,9 @@ c_parser_omp_all_clauses (c_parser *parser, omp_clause_mask mask,
       const char *c_name;
       tree prev = clauses;
 
+      if (nested_p && c_parser_next_token_is (parser, CPP_CLOSE_PAREN))
+	break;
+
       if (!first && c_parser_next_token_is (parser, CPP_COMMA))
 	c_parser_consume_token (parser);
 
@@ -15513,7 +15520,8 @@ c_parser_omp_all_clauses (c_parser *parser, omp_clause_mask mask,
     }
 
  saw_error:
-  c_parser_skip_to_pragma_eol (parser);
+  if (!nested_p)
+    c_parser_skip_to_pragma_eol (parser);
 
   if (finish_p)
     {
@@ -18919,7 +18927,11 @@ check_clauses:
 }
 
 /* OpenMP 4.0:
-   # pragma omp declare simd declare-simd-clauses[optseq] new-line  */
+   # pragma omp declare simd declare-simd-clauses[optseq] new-line
+
+   OpenMP 5.0:
+   # pragma omp declare variant (identifier) match(context-selector) new-line
+   */
 
 #define OMP_DECLARE_SIMD_CLAUSE_MASK				\
 	( (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_SIMDLEN)	\
@@ -18932,6 +18944,12 @@ check_clauses:
 static void
 c_parser_omp_declare_simd (c_parser *parser, enum pragma_context context)
 {
+  c_token *token = c_parser_peek_token (parser);
+  gcc_assert (token->type == CPP_NAME);
+  tree kind = token->value;
+  gcc_assert (strcmp (IDENTIFIER_POINTER (kind), "simd") == 0
+	      || strcmp (IDENTIFIER_POINTER (kind), "variant") == 0);
+
   auto_vec<c_token> clauses;
   while (c_parser_next_token_is_not (parser, CPP_PRAGMA_EOL))
     {
@@ -18949,17 +18967,14 @@ c_parser_omp_declare_simd (c_parser *parser, enum pragma_context context)
 
   while (c_parser_next_token_is (parser, CPP_PRAGMA))
     {
-      if (c_parser_peek_token (parser)->pragma_kind
-	  != PRAGMA_OMP_DECLARE
+      if (c_parser_peek_token (parser)->pragma_kind != PRAGMA_OMP_DECLARE
 	  || c_parser_peek_2nd_token (parser)->type != CPP_NAME
-	  || strcmp (IDENTIFIER_POINTER
-				(c_parser_peek_2nd_token (parser)->value),
-		     "simd") != 0)
+	  || c_parser_peek_2nd_token (parser)->value != kind)
 	{
-	  c_parser_error (parser,
-			  "%<#pragma omp declare simd%> must be followed by "
-			  "function declaration or definition or another "
-			  "%<#pragma omp declare simd%>");
+	  error ("%<#pragma omp declare %s%> must be followed by "
+		 "function declaration or definition or another "
+		 "%<#pragma omp declare %s%>",
+		 IDENTIFIER_POINTER (kind), IDENTIFIER_POINTER (kind));
 	  return;
 	}
       c_parser_consume_pragma (parser);
@@ -19007,8 +19022,9 @@ c_parser_omp_declare_simd (c_parser *parser, enum pragma_context context)
     case pragma_struct:
     case pragma_param:
     case pragma_stmt:
-      c_parser_error (parser, "%<#pragma omp declare simd%> must be followed by "
-			      "function declaration or definition");
+      error ("%<#pragma omp declare %s%> must be followed by "
+	     "function declaration or definition",
+	     IDENTIFIER_POINTER (kind));
       break;
     case pragma_compound:
       if (c_parser_next_token_is (parser, CPP_KEYWORD)
@@ -19034,38 +19050,470 @@ c_parser_omp_declare_simd (c_parser *parser, enum pragma_context context)
 					 NULL, clauses);
 	  break;
 	}
-      c_parser_error (parser, "%<#pragma omp declare simd%> must be followed by "
-			      "function declaration or definition");
+      error ("%<#pragma omp declare %s%> must be followed by "
+	     "function declaration or definition",
+	     IDENTIFIER_POINTER (kind));
       break;
     default:
       gcc_unreachable ();
     }
 }
 
-/* Finalize #pragma omp declare simd clauses after FNDECL has been parsed,
-   and put that into "omp declare simd" attribute.  */
+static const char *const omp_construct_selectors[] = {
+  "simd", "target", "teams", "parallel", "for", NULL };
+static const char *const omp_device_selectors[] = {
+  "kind", "isa", "arch", NULL };
+static const char *const omp_implementation_selectors[] = {
+  "vendor", "extension", "atomic_default_mem_order", "unified_address",
+  "unified_shared_memory", "dynamic_allocators", "reverse_offload", NULL };
+static const char *const omp_user_selectors[] = {
+  "condition", NULL };
+
+/* OpenMP 5.0:
+
+   trait-selector:
+     trait-selector-name[([trait-score:]trait-property[,trait-property[,...]])]
+
+   trait-score:
+     score(score-expression)  */
+
+static tree
+c_parser_omp_context_selector (c_parser *parser, tree set, tree parms)
+{
+  tree ret = NULL_TREE;
+  do
+    {
+      tree selector;
+      if (c_parser_next_token_is (parser, CPP_KEYWORD)
+	  || c_parser_next_token_is (parser, CPP_NAME))
+	selector = c_parser_peek_token (parser)->value;
+      else
+	{
+	  c_parser_error (parser, "expected trait selector name");
+	  return error_mark_node;
+	}
+
+      tree properties = NULL_TREE;
+      const char *const *selectors = NULL;
+      bool allow_score = true;
+      bool allow_user = false;
+      int property_limit = 0;
+      enum { CTX_PROPERTY_NONE, CTX_PROPERTY_USER, CTX_PROPERTY_IDLIST,
+	     CTX_PROPERTY_EXPR, CTX_PROPERTY_SIMD } property_kind
+	= CTX_PROPERTY_NONE;
+      switch (IDENTIFIER_POINTER (set)[0])
+	{
+	case 'c': /* construct */
+	  selectors = omp_construct_selectors;
+	  allow_score = false;
+	  property_limit = 1;
+	  property_kind = CTX_PROPERTY_SIMD;
+	  break;
+	case 'd': /* device */
+	  selectors = omp_device_selectors;
+	  allow_score = false;
+	  allow_user = true;
+	  property_limit = 3;
+	  property_kind = CTX_PROPERTY_IDLIST;
+	  break;
+	case 'i': /* implementation */
+	  selectors = omp_implementation_selectors;
+	  allow_user = true;
+	  property_limit = 3;
+	  property_kind = CTX_PROPERTY_IDLIST;
+	  break;
+	case 'u': /* user */
+	  selectors = omp_user_selectors;
+	  property_limit = 1;
+	  property_kind = CTX_PROPERTY_EXPR;
+	  break;
+	default:
+	  gcc_unreachable ();
+	}
+      for (int i = 0; ; i++)
+	{
+	  if (selectors[i] == NULL)
+	    {
+	      if (allow_user)
+		{
+		  property_kind = CTX_PROPERTY_USER;
+		  break;
+		}
+	      else
+		{
+		  error_at (c_parser_peek_token (parser)->location,
+			    "selector %qs not allowed for context selector "
+			    "set %qs", IDENTIFIER_POINTER (selector),
+			    IDENTIFIER_POINTER (set));
+		  c_parser_consume_token (parser);
+		  return error_mark_node;
+		}
+	    }
+	  if (i == property_limit)
+	    property_kind = CTX_PROPERTY_NONE;
+	  if (strcmp (selectors[i], IDENTIFIER_POINTER (selector)) == 0)
+	    break;
+	}
+
+      c_parser_consume_token (parser);
+
+      if (c_parser_next_token_is (parser, CPP_OPEN_PAREN))
+	{
+	  if (property_kind == CTX_PROPERTY_NONE)
+	    {
+	      error_at (c_parser_peek_token (parser)->location,
+			"selector %qs does not accept any properties",
+			IDENTIFIER_POINTER (selector));
+	      return error_mark_node;
+	    }
+
+	  matching_parens parens;
+	  parens.require_open (parser);
+
+	  c_token *token = c_parser_peek_token (parser);
+	  if (allow_score
+	      && c_parser_next_token_is (parser, CPP_NAME)
+	      && strcmp (IDENTIFIER_POINTER (token->value), "score") == 0
+	      && c_parser_peek_2nd_token (parser)->type == CPP_OPEN_PAREN)
+	    {
+	      c_parser_consume_token (parser);
+
+	      matching_parens parens2;
+	      parens2.require_open (parser);
+	      tree score = c_parser_expr_no_commas (parser, NULL).value;
+	      parens2.skip_until_found_close (parser);
+	      c_parser_require (parser, CPP_COLON, "expected %<:%>");
+	      if (score != error_mark_node)
+		{
+		  mark_exp_read (score);
+		  score = c_fully_fold (score, false, NULL);
+		  if (!INTEGRAL_TYPE_P (TREE_TYPE (score))
+		      || !tree_fits_shwi_p (score))
+		    error_at (token->location, "score argument must be "
+			      "constant integer expression");
+		  else
+		    properties = tree_cons (get_identifier (" score"),
+					    score, properties);
+		}
+	      token = c_parser_peek_token (parser);
+	    }
+
+	  switch (property_kind)
+	    {
+	      tree t;
+	    case CTX_PROPERTY_USER:
+	      do
+		{
+		  t = c_parser_expr_no_commas (parser, NULL).value;
+		  if (TREE_CODE (t) == STRING_CST)
+		    properties = tree_cons (NULL_TREE, t, properties);
+		  else if (t != error_mark_node)
+		    {
+		      mark_exp_read (t);
+		      t = c_fully_fold (t, false, NULL);
+		      if (!INTEGRAL_TYPE_P (TREE_TYPE (t))
+			  || !tree_fits_shwi_p (t))
+			error_at (token->location, "property must be "
+				  "constant integer expression or string "
+				  "literal");
+		      else
+			properties = tree_cons (NULL_TREE, t, properties);
+		    }
+
+		  if (c_parser_next_token_is (parser, CPP_COMMA))
+		    c_parser_consume_token (parser);
+		  else
+		    break;
+		}
+	      while (1);
+	      break;
+	    case CTX_PROPERTY_IDLIST:
+	      do
+		{
+		  tree prop;
+		  if (c_parser_next_token_is (parser, CPP_KEYWORD)
+		      || c_parser_next_token_is (parser, CPP_NAME))
+		    prop = c_parser_peek_token (parser)->value;
+		  else
+		    {
+		      c_parser_error (parser, "expected identifier");
+		      return error_mark_node;
+		    }
+		  c_parser_consume_token (parser);
+
+		  properties = tree_cons (prop, NULL_TREE, properties);
+
+		  if (c_parser_next_token_is (parser, CPP_COMMA))
+		    c_parser_consume_token (parser);
+		  else
+		    break;
+		}
+	      while (1);
+	      break;
+	    case CTX_PROPERTY_EXPR:
+	      t = c_parser_expr_no_commas (parser, NULL).value;
+	      if (t != error_mark_node)
+		{
+		  mark_exp_read (t);
+		  t = c_fully_fold (t, false, NULL);
+		  if (!INTEGRAL_TYPE_P (TREE_TYPE (t))
+		      || !tree_fits_shwi_p (t))
+		    error_at (token->location, "property must be "
+			      "constant integer expression");
+		  else
+		    properties = tree_cons (NULL_TREE, t, properties);
+		}
+	      break;
+	    case CTX_PROPERTY_SIMD:
+	      if (parms == NULL_TREE)
+		{
+		  error_at (token->location, "properties for %<simd%> "
+			    "selector may not be specified in "
+			    "%<metadirective%>");
+		  return error_mark_node;
+		}
+	      tree c;
+	      c = c_parser_omp_all_clauses (parser,
+					    OMP_DECLARE_SIMD_CLAUSE_MASK,
+					    "simd", true, true);
+	      c = c_omp_declare_simd_clauses_to_numbers (parms
+							 == error_mark_node
+							 ? NULL_TREE : parms,
+							 c);
+	      properties = tree_cons (NULL_TREE, c, properties);
+	      break;
+	    default:
+	      gcc_unreachable ();
+	    }
+
+	  parens.skip_until_found_close (parser);
+	  properties = nreverse (properties);
+	}
+      else if (property_kind == CTX_PROPERTY_IDLIST
+	       || property_kind == CTX_PROPERTY_EXPR)
+	{
+	  c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>");
+	  return error_mark_node;
+	}
+
+      ret = tree_cons (selector, properties, ret);
+
+      if (c_parser_next_token_is (parser, CPP_COMMA))
+	c_parser_consume_token (parser);
+      else
+	break;
+    }
+  while (1);
+
+  return nreverse (ret);
+}
+
+/* OpenMP 5.0:
+
+   trait-set-selector[,trait-set-selector[,...]]
+
+   trait-set-selector:
+     trait-set-selector-name = { trait-selector[, trait-selector[, ...]] }
+
+   trait-set-selector-name:
+     constructor
+     device
+     implementation
+     user  */
+
+static tree
+c_parser_omp_context_selector_specification (c_parser *parser, tree parms)
+{
+  tree ret = NULL_TREE;
+  do
+    {
+      const char *setp = "";
+      if (c_parser_next_token_is (parser, CPP_NAME))
+	setp = IDENTIFIER_POINTER (c_parser_peek_token (parser)->value);
+      switch (setp[0])
+	{
+	case 'c':
+	  if (strcmp (setp, "construct") == 0)
+	    setp = NULL;
+	  break;
+	case 'd':
+	  if (strcmp (setp, "device") == 0)
+	    setp = NULL;
+	  break;
+	case 'i':
+	  if (strcmp (setp, "implementation") == 0)
+	    setp = NULL;
+	  break;
+	case 'u':
+	  if (strcmp (setp, "user") == 0)
+	    setp = NULL;
+	  break;
+	default:
+	  break;
+	}
+      if (setp)
+	{
+	  c_parser_error (parser, "expected %<construct%>, %<device%>, "
+				  "%<implementation%> or %<user%>");
+	  return error_mark_node;
+	}
+
+      tree set = c_parser_peek_token (parser)->value;
+      c_parser_consume_token (parser);
+
+      if (!c_parser_require (parser, CPP_EQ, "expected %<=%>"))
+	return error_mark_node;
+
+      matching_braces braces;
+      if (!braces.require_open (parser))
+	return error_mark_node;
+
+      tree selectors = c_parser_omp_context_selector (parser, set, parms);
+      if (selectors == error_mark_node)
+	ret = error_mark_node;
+      else if (ret != error_mark_node)
+	ret = tree_cons (set, selectors, ret);
+
+      braces.skip_until_found_close (parser);
+
+      if (c_parser_next_token_is (parser, CPP_COMMA))
+	c_parser_consume_token (parser);
+      else
+	break;
+    }
+  while (1);
+
+  if (ret == error_mark_node)
+    return ret;
+  return nreverse (ret);
+}
+
+/* Finalize #pragma omp declare variant after FNDECL has been parsed, and put
+   that into "omp declare variant" attribute.  */
+
+static void
+c_finish_omp_declare_variant (c_parser *parser, tree fndecl, tree parms)
+{
+  matching_parens parens;
+  if (!parens.require_open (parser))
+    {
+     fail:
+      c_parser_skip_to_pragma_eol (parser, false);
+      return;
+    }
+
+  if (c_parser_next_token_is_not (parser, CPP_NAME)
+      || c_parser_peek_token (parser)->id_kind != C_ID_ID)
+    {
+      c_parser_error (parser, "expected identifier");
+      goto fail;
+    }
+
+  c_token *token = c_parser_peek_token (parser);
+  tree variant = lookup_name (token->value);
+
+  if (variant == NULL_TREE)
+    {
+      undeclared_variable (token->location, token->value);
+      variant = error_mark_node;
+    }
+
+  c_parser_consume_token (parser);
+
+  parens.require_close (parser);
+
+  const char *clause = "";
+  location_t match_loc = c_parser_peek_token (parser)->location;
+  if (c_parser_next_token_is (parser, CPP_NAME))
+    clause = IDENTIFIER_POINTER (c_parser_peek_token (parser)->value);
+  if (strcmp (clause, "match"))
+    {
+      c_parser_error (parser, "expected %<match%>");
+      goto fail;
+    }
+
+  c_parser_consume_token (parser);
+
+  if (!parens.require_open (parser))
+    goto fail;
+
+  if (parms == NULL_TREE)
+    parms = error_mark_node;
+
+  tree ctx = c_parser_omp_context_selector_specification (parser, parms);
+  if (ctx == error_mark_node)
+    goto fail;
+  ctx = c_omp_check_context_selector (match_loc, ctx);
+  if (ctx != error_mark_node && variant != error_mark_node)
+    {
+      if (TREE_CODE (variant) != FUNCTION_DECL)
+	{
+	  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
+	       && !comptypes (TREE_TYPE (fndecl), TREE_TYPE (variant)))
+	{
+	  error_at (token->location, "variant %qD and base %qD have "
+				     "incompatible types", variant, fndecl);
+	  variant = error_mark_node;
+	}
+      else if (fndecl_built_in_p (variant)
+	       && (strncmp (IDENTIFIER_POINTER (DECL_NAME (variant)),
+			    "__builtin_", strlen ("__builtin_")) == 0
+		   || strncmp (IDENTIFIER_POINTER (DECL_NAME (variant)),
+			       "__sync_", strlen ("__sync_")) == 0
+		   || strncmp (IDENTIFIER_POINTER (DECL_NAME (variant)),
+			       "__atomic_", strlen ("__atomic_")) == 0))
+	{
+	  error_at (token->location, "variant %qD is a built-in", variant);
+	  variant = error_mark_node;
+	}
+      if (variant != error_mark_node)
+	{
+	  C_DECL_USED (variant) = 1;
+	  tree attr = tree_cons (get_identifier ("omp declare variant"),
+				 build_tree_list (variant, ctx),
+				 DECL_ATTRIBUTES (fndecl));
+	  DECL_ATTRIBUTES (fndecl) = attr;
+	}
+    }
+
+  parens.require_close (parser);
+  c_parser_skip_to_pragma_eol (parser);
+}
+
+/* Finalize #pragma omp declare simd or #pragma omp declare variant
+   clauses after FNDECL has been parsed, and put that into "omp declare simd"
+   or "omp declare variant" attribute.  */
 
 static void
 c_finish_omp_declare_simd (c_parser *parser, tree fndecl, tree parms,
 			   vec<c_token> clauses)
 {
-  /* Normally first token is CPP_NAME "simd".  CPP_EOF there indicates
-     error has been reported and CPP_PRAGMA that c_finish_omp_declare_simd
-     has already processed the tokens.  */
+  /* Normally first token is CPP_NAME "simd" or "variant".  CPP_EOF there
+     indicates error has been reported and CPP_PRAGMA that
+     c_finish_omp_declare_simd has already processed the tokens.  */
   if (clauses.exists () && clauses[0].type == CPP_EOF)
     return;
+  const char *kind = "simd";
+  if (clauses.exists ()
+      && (clauses[0].type == CPP_NAME || clauses[0].type == CPP_PRAGMA))
+    kind = IDENTIFIER_POINTER (clauses[0].value);
+  gcc_assert (strcmp (kind, "simd") == 0 || strcmp (kind, "variant") == 0);
   if (fndecl == NULL_TREE || TREE_CODE (fndecl) != FUNCTION_DECL)
     {
-      error ("%<#pragma omp declare simd%> not immediately followed by "
-	     "a function declaration or definition");
+      error ("%<#pragma omp declare %s%> not immediately followed by "
+	     "a function declaration or definition", kind);
       clauses[0].type = CPP_EOF;
       return;
     }
   if (clauses.exists () && clauses[0].type != CPP_NAME)
     {
       error_at (DECL_SOURCE_LOCATION (fndecl),
-		"%<#pragma omp declare simd%> not immediately followed by "
-		"a single function declaration or definition");
+		"%<#pragma omp declare %s%> not immediately followed by "
+		"a single function declaration or definition", kind);
       clauses[0].type = CPP_EOF;
       return;
     }
@@ -19075,7 +19523,6 @@ c_finish_omp_declare_simd (c_parser *parser, tree fndecl, tree parms,
 
   unsigned int tokens_avail = parser->tokens_avail;
   gcc_assert (parser->tokens == &parser->tokens_buf[0]);
-  
 
   parser->tokens = clauses.address ();
   parser->tokens_avail = clauses.length ();
@@ -19085,19 +19532,27 @@ c_finish_omp_declare_simd (c_parser *parser, tree fndecl, tree parms,
     {
       c_token *token = c_parser_peek_token (parser);
       gcc_assert (token->type == CPP_NAME
-		  && strcmp (IDENTIFIER_POINTER (token->value), "simd") == 0);
+		  && strcmp (IDENTIFIER_POINTER (token->value), kind) == 0);
       c_parser_consume_token (parser);
       parser->in_pragma = true;
 
-      tree c = NULL_TREE;
-      c = c_parser_omp_all_clauses (parser, OMP_DECLARE_SIMD_CLAUSE_MASK,
-				      "#pragma omp declare simd");
-      c = c_omp_declare_simd_clauses_to_numbers (parms, c);
-      if (c != NULL_TREE)
-	c = tree_cons (NULL_TREE, c, NULL_TREE);
-      c = build_tree_list (get_identifier ("omp declare simd"), c);
-      TREE_CHAIN (c) = DECL_ATTRIBUTES (fndecl);
-      DECL_ATTRIBUTES (fndecl) = c;
+      if (strcmp (kind, "simd") == 0)
+	{
+	  tree c;
+	  c = c_parser_omp_all_clauses (parser, OMP_DECLARE_SIMD_CLAUSE_MASK,
+					"#pragma omp declare simd");
+	  c = c_omp_declare_simd_clauses_to_numbers (parms, c);
+	  if (c != NULL_TREE)
+	    c = tree_cons (NULL_TREE, c, NULL_TREE);
+	  c = build_tree_list (get_identifier ("omp declare simd"), c);
+	  TREE_CHAIN (c) = DECL_ATTRIBUTES (fndecl);
+	  DECL_ATTRIBUTES (fndecl) = c;
+	}
+      else
+	{
+	  gcc_assert (strcmp (kind, "variant") == 0);
+	  c_finish_omp_declare_variant (parser, fndecl, parms);
+	}
     }
 
   parser->tokens = &parser->tokens_buf[0];
@@ -19612,7 +20067,10 @@ c_parser_omp_declare_reduction (c_parser *parser, enum pragma_context context)
    #pragma omp declare simd declare-simd-clauses[optseq] new-line
    #pragma omp declare reduction (reduction-id : typename-list : expression) \
       initializer-clause[opt] new-line
-   #pragma omp declare target new-line  */
+   #pragma omp declare target new-line
+
+   OpenMP 5.0
+   #pragma omp declare variant (identifier) match (context-selector)  */
 
 static void
 c_parser_omp_declare (c_parser *parser, enum pragma_context context)
@@ -19645,10 +20103,17 @@ c_parser_omp_declare (c_parser *parser, enum pragma_context context)
 	  c_parser_omp_declare_target (parser);
 	  return;
 	}
+      if (strcmp (p, "variant") == 0)
+	{
+	  /* c_parser_consume_token (parser); done in
+	     c_parser_omp_declare_simd.  */
+	  c_parser_omp_declare_simd (parser, context);
+	  return;
+	}
     }
 
-  c_parser_error (parser, "expected %<simd%> or %<reduction%> "
-			  "or %<target%>");
+  c_parser_error (parser, "expected %<simd%>, %<reduction%>, "
+			  "%<target%> or %<variant%>");
   c_parser_skip_to_pragma_eol (parser);
 }
 
diff --git a/gcc/cp/ChangeLog b/gcc/cp/ChangeLog
index 2a35bf2..9ed7311 100644
--- a/gcc/cp/ChangeLog
+++ b/gcc/cp/ChangeLog
@@ -1,3 +1,21 @@
+2019-10-10  Jakub Jelinek  <jakub@redhat.com>
+
+	* parser.h (struct cp_omp_declare_simd_data): Add variant_p member.
+	* parser.c (cp_ensure_no_omp_declare_simd): Handle both declare simd
+	and declare variant.
+	(cp_parser_oacc_all_clauses): Formatting fix.
+	(cp_parser_omp_all_clauses): Add NESTED_P argument, if true, terminate
+	processing on closing paren and don't skip to end of pragma line.
+	(cp_parser_omp_declare_simd): Add VARIANT_P argument.  Handle also
+	declare variant.
+	(omp_construct_selectors, omp_device_selectors,
+	omp_implementation_selectors, omp_user_selectors): New variables.
+	(cp_parser_omp_context_selector,
+	cp_parser_omp_context_selector_specification,
+	cp_finish_omp_declare_variant): New functions.
+	(cp_parser_late_parsing_omp_declare_simd): Handle also declare variant.
+	(cp_parser_omp_declare): Handle declare variant.
+
 2019-10-09  Jason Merrill  <jason@redhat.com>
 
 	* cp-tree.h (template_info_decl_check): Check ENABLE_TREE_CHECKING.
diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c
index b6e738c..3ee8da7 100644
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -1376,8 +1376,9 @@ cp_ensure_no_omp_declare_simd (cp_parser *parser)
 {
   if (parser->omp_declare_simd && !parser->omp_declare_simd->error_seen)
     {
-      error ("%<#pragma omp declare simd%> not immediately followed by "
-	     "function declaration or definition");
+      error ("%<#pragma omp declare %s%> not immediately followed by "
+	     "function declaration or definition",
+	     parser->omp_declare_simd->variant_p ? "variant" : "simd");
       parser->omp_declare_simd = NULL;
     }
 }
@@ -35878,8 +35879,8 @@ cp_parser_oacc_clause_async (cp_parser *parser, tree list)
 
 static tree
 cp_parser_oacc_all_clauses (cp_parser *parser, omp_clause_mask mask,
-			   const char *where, cp_token *pragma_tok,
-			   bool finish_p = true)
+			    const char *where, cp_token *pragma_tok,
+			    bool finish_p = true)
 {
   tree clauses = NULL;
   bool first = true;
@@ -36075,13 +36076,15 @@ cp_parser_oacc_all_clauses (cp_parser *parser, omp_clause_mask mask,
 }
 
 /* Parse all OpenMP clauses.  The set clauses allowed by the directive
-   is a bitmask in MASK.  Return the list of clauses found; the result
-   of clause default goes in *pdefault.  */
+   is a bitmask in MASK.  Return the list of clauses found.
+   FINISH_P set if finish_omp_clauses should be called.
+   NESTED_P set if clauses should be terminated by closing paren instead
+   of end of pragma.  */
 
 static tree
 cp_parser_omp_all_clauses (cp_parser *parser, omp_clause_mask mask,
 			   const char *where, cp_token *pragma_tok,
-			   bool finish_p = true)
+			   bool finish_p = true, bool nested_p = false)
 {
   tree clauses = NULL;
   bool first = true;
@@ -36096,6 +36099,9 @@ cp_parser_omp_all_clauses (cp_parser *parser, omp_clause_mask mask,
       const char *c_name;
       tree prev = clauses;
 
+      if (nested_p && cp_lexer_next_token_is (parser->lexer, CPP_CLOSE_PAREN))
+	break;
+
       if (!first && cp_lexer_next_token_is (parser->lexer, CPP_COMMA))
 	cp_lexer_consume_token (parser->lexer);
 
@@ -36415,7 +36421,8 @@ cp_parser_omp_all_clauses (cp_parser *parser, omp_clause_mask mask,
 	}
     }
  saw_error:
-  cp_parser_skip_to_pragma_eol (parser, pragma_tok);
+  if (!nested_p)
+    cp_parser_skip_to_pragma_eol (parser, pragma_tok);
   if (finish_p)
     {
       if ((mask & (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_UNIFORM)) != 0)
@@ -40241,7 +40248,8 @@ cp_parser_oacc_wait (cp_parser *parser, cp_token *pragma_tok)
 
 static void
 cp_parser_omp_declare_simd (cp_parser *parser, cp_token *pragma_tok,
-			    enum pragma_context context)
+			    enum pragma_context context,
+			    bool variant_p)
 {
   bool first_p = parser->omp_declare_simd == NULL;
   cp_omp_declare_simd_data data;
@@ -40249,12 +40257,22 @@ cp_parser_omp_declare_simd (cp_parser *parser, cp_token *pragma_tok,
     {
       data.error_seen = false;
       data.fndecl_seen = false;
+      data.variant_p = variant_p;
       data.tokens = vNULL;
       data.clauses = NULL_TREE;
       /* It is safe to take the address of a local variable; it will only be
 	 used while this scope is live.  */
       parser->omp_declare_simd = &data;
     }
+  else if (parser->omp_declare_simd->variant_p != variant_p)
+    {
+      error_at (pragma_tok->location,
+		"%<#pragma omp declare %s%> followed by "
+		"%<#pragma omp declare %s%>",
+		parser->omp_declare_simd->variant_p ? "variant" : "simd",
+		parser->omp_declare_simd->variant_p ? "simd" : "variant");
+      parser->omp_declare_simd->error_seen = true;
+    }
 
   /* Store away all pragma tokens.  */
   while (cp_lexer_next_token_is_not (parser->lexer, CPP_PRAGMA_EOL)
@@ -40290,13 +40308,427 @@ cp_parser_omp_declare_simd (cp_parser *parser, cp_token *pragma_tok,
 	  && !parser->omp_declare_simd->error_seen
 	  && !parser->omp_declare_simd->fndecl_seen)
 	error_at (pragma_tok->location,
-		  "%<#pragma omp declare simd%> not immediately followed by "
-		  "function declaration or definition");
+		  "%<#pragma omp declare %s%> not immediately followed by "
+		  "function declaration or definition",
+		  parser->omp_declare_simd->variant_p ? "variant" : "simd");
       data.tokens.release ();
       parser->omp_declare_simd = NULL;
     }
 }
 
+static const char *const omp_construct_selectors[] = {
+  "simd", "target", "teams", "parallel", "for", NULL };
+static const char *const omp_device_selectors[] = {
+  "kind", "isa", "arch", NULL };
+static const char *const omp_implementation_selectors[] = {
+  "vendor", "extension", "atomic_default_mem_order", "unified_address",
+  "unified_shared_memory", "dynamic_allocators", "reverse_offload", NULL };
+static const char *const omp_user_selectors[] = {
+  "condition", NULL };
+
+/* OpenMP 5.0:
+
+   trait-selector:
+     trait-selector-name[([trait-score:]trait-property[,trait-property[,...]])]
+
+   trait-score:
+     score(score-expression)  */
+
+static tree
+cp_parser_omp_context_selector (cp_parser *parser, tree set, bool has_parms_p)
+{
+  tree ret = NULL_TREE;
+  do
+    {
+      tree selector;
+      if (cp_lexer_next_token_is (parser->lexer, CPP_KEYWORD)
+	  || cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+	selector = cp_lexer_peek_token (parser->lexer)->u.value;
+      else
+	{
+	  cp_parser_error (parser, "expected trait selector name");
+	  return error_mark_node;
+	}
+
+      tree properties = NULL_TREE;
+      const char *const *selectors = NULL;
+      bool allow_score = true;
+      bool allow_user = false;
+      int property_limit = 0;
+      enum { CTX_PROPERTY_NONE, CTX_PROPERTY_USER, CTX_PROPERTY_IDLIST,
+	     CTX_PROPERTY_EXPR, CTX_PROPERTY_SIMD } property_kind
+	= CTX_PROPERTY_NONE;
+      switch (IDENTIFIER_POINTER (set)[0])
+	{
+	case 'c': /* construct */
+	  selectors = omp_construct_selectors;
+	  allow_score = false;
+	  property_limit = 1;
+	  property_kind = CTX_PROPERTY_SIMD;
+	  break;
+	case 'd': /* device */
+	  selectors = omp_device_selectors;
+	  allow_score = false;
+	  allow_user = true;
+	  property_limit = 3;
+	  property_kind = CTX_PROPERTY_IDLIST;
+	  break;
+	case 'i': /* implementation */
+	  selectors = omp_implementation_selectors;
+	  allow_user = true;
+	  property_limit = 3;
+	  property_kind = CTX_PROPERTY_IDLIST;
+	  break;
+	case 'u': /* user */
+	  selectors = omp_user_selectors;
+	  property_limit = 1;
+	  property_kind = CTX_PROPERTY_EXPR;
+	  break;
+	default:
+	  gcc_unreachable ();
+	}
+      for (int i = 0; ; i++)
+	{
+	  if (selectors[i] == NULL)
+	    {
+	      if (allow_user)
+		{
+		  property_kind = CTX_PROPERTY_USER;
+		  break;
+		}
+	      else
+		{
+		  error ("selector %qs not allowed for context selector "
+			 "set %qs", IDENTIFIER_POINTER (selector),
+			 IDENTIFIER_POINTER (set));
+		  cp_lexer_consume_token (parser->lexer);
+		  return error_mark_node;
+		}
+	    }
+	  if (i == property_limit)
+	    property_kind = CTX_PROPERTY_NONE;
+	  if (strcmp (selectors[i], IDENTIFIER_POINTER (selector)) == 0)
+	    break;
+	}
+
+      cp_lexer_consume_token (parser->lexer);
+
+      if (cp_lexer_next_token_is (parser->lexer, CPP_OPEN_PAREN))
+	{
+	  if (property_kind == CTX_PROPERTY_NONE)
+	    {
+	      error ("selector %qs does not accept any properties",
+		     IDENTIFIER_POINTER (selector));
+	      return error_mark_node;
+	    }
+
+	  matching_parens parens;
+	  parens.consume_open (parser);
+
+	  cp_token *token = cp_lexer_peek_token (parser->lexer);
+	  if (allow_score
+	      && cp_lexer_next_token_is (parser->lexer, CPP_NAME)
+	      && strcmp (IDENTIFIER_POINTER (token->u.value), "score") == 0
+	      && cp_lexer_nth_token_is (parser->lexer, 2, CPP_OPEN_PAREN))
+	    {
+	      cp_lexer_save_tokens (parser->lexer);
+	      cp_lexer_consume_token (parser->lexer);
+	      cp_lexer_consume_token (parser->lexer);
+	      if (cp_parser_skip_to_closing_parenthesis (parser, false, false,
+							 true)
+		  && cp_lexer_next_token_is (parser->lexer, CPP_COLON))
+		{
+		  cp_lexer_rollback_tokens (parser->lexer);
+		  cp_lexer_consume_token (parser->lexer);
+
+		  matching_parens parens2;
+		  parens2.require_open (parser);
+		  tree score = cp_parser_constant_expression (parser);
+		  if (!parens2.require_close (parser))
+		    cp_parser_skip_to_closing_parenthesis (parser, true,
+							   false, true);
+		  cp_parser_require (parser, CPP_COLON, RT_COLON);
+		  if (score != error_mark_node)
+		    {
+		      score = fold_non_dependent_expr (score);
+		      if (!value_dependent_expression_p (score)
+			  && (!INTEGRAL_TYPE_P (TREE_TYPE (score))
+			      || !tree_fits_shwi_p (score)))
+			error_at (token->location, "score argument must be "
+				  "constant integer expression");
+		      else
+			properties = tree_cons (get_identifier (" score"),
+						score, properties);
+		    }
+		}
+	      else
+		cp_lexer_rollback_tokens (parser->lexer);
+
+	      token = cp_lexer_peek_token (parser->lexer);
+	    }
+
+	  switch (property_kind)
+	    {
+	      tree t;
+	    case CTX_PROPERTY_USER:
+	      do
+		{
+		  t = cp_parser_constant_expression (parser);
+		  if (t != error_mark_node)
+		    {
+		      t = fold_non_dependent_expr (t);
+		      if (TREE_CODE (t) == STRING_CST)
+			properties = tree_cons (NULL_TREE, t, properties);
+		      else if (!value_dependent_expression_p (t)
+			       && (!INTEGRAL_TYPE_P (TREE_TYPE (t))
+				   || !tree_fits_shwi_p (t)))
+			error_at (token->location, "property must be "
+				  "constant integer expression or string "
+				  "literal");
+		      else
+			properties = tree_cons (NULL_TREE, t, properties);
+		    }
+
+		  if (cp_lexer_next_token_is (parser->lexer, CPP_COMMA))
+		    cp_lexer_consume_token (parser->lexer);
+		  else
+		    break;
+		}
+	      while (1);
+	      break;
+	    case CTX_PROPERTY_IDLIST:
+	      do
+		{
+		  tree prop;
+		  if (cp_lexer_next_token_is (parser->lexer, CPP_KEYWORD)
+		      || cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+		    prop = cp_lexer_peek_token (parser->lexer)->u.value;
+		  else
+		    {
+		      cp_parser_error (parser, "expected identifier");
+		      return error_mark_node;
+		    }
+		  cp_lexer_consume_token (parser->lexer);
+
+		  properties = tree_cons (prop, NULL_TREE, properties);
+
+		  if (cp_lexer_next_token_is (parser->lexer, CPP_COMMA))
+		    cp_lexer_consume_token (parser->lexer);
+		  else
+		    break;
+		}
+	      while (1);
+	      break;
+	    case CTX_PROPERTY_EXPR:
+	      t = cp_parser_constant_expression (parser);
+	      if (t != error_mark_node)
+		{
+		  t = fold_non_dependent_expr (t);
+		  if (!value_dependent_expression_p (t)
+		      && (!INTEGRAL_TYPE_P (TREE_TYPE (t))
+			  || !tree_fits_shwi_p (t)))
+		    error_at (token->location, "property must be "
+			      "constant integer expression");
+		  else
+		    properties = tree_cons (NULL_TREE, t, properties);
+		}
+	      break;
+	    case CTX_PROPERTY_SIMD:
+	      if (!has_parms_p)
+		{
+		  error_at (token->location, "properties for %<simd%> "
+			    "selector may not be specified in "
+			    "%<metadirective%>");
+		  return error_mark_node;
+		}
+	      tree c;
+	      c = cp_parser_omp_all_clauses (parser,
+					     OMP_DECLARE_SIMD_CLAUSE_MASK,
+					     "simd", NULL, true, true);
+	      properties = tree_cons (NULL_TREE, c, properties);
+	      break;
+	    default:
+	      gcc_unreachable ();
+	    }
+
+	  if (!parens.require_close (parser))
+	    cp_parser_skip_to_closing_parenthesis (parser, true, false, true);
+
+	  properties = nreverse (properties);
+	}
+      else if (property_kind == CTX_PROPERTY_IDLIST
+	       || property_kind == CTX_PROPERTY_EXPR)
+	{
+	  cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN);
+	  return error_mark_node;
+	}
+
+      ret = tree_cons (selector, properties, ret);
+
+      if (cp_lexer_next_token_is (parser->lexer, CPP_COMMA))
+	cp_lexer_consume_token (parser->lexer);
+      else
+	break;
+    }
+  while (1);
+
+  return nreverse (ret);
+}
+
+/* OpenMP 5.0:
+
+   trait-set-selector[,trait-set-selector[,...]]
+
+   trait-set-selector:
+     trait-set-selector-name = { trait-selector[, trait-selector[, ...]] }
+
+   trait-set-selector-name:
+     constructor
+     device
+     implementation
+     user  */
+
+static tree
+cp_parser_omp_context_selector_specification (cp_parser *parser,
+					      bool has_parms_p)
+{
+  tree ret = NULL_TREE;
+  do
+    {
+      const char *setp = "";
+      if (cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+	setp
+	  = IDENTIFIER_POINTER (cp_lexer_peek_token (parser->lexer)->u.value);
+      switch (setp[0])
+	{
+	case 'c':
+	  if (strcmp (setp, "construct") == 0)
+	    setp = NULL;
+	  break;
+	case 'd':
+	  if (strcmp (setp, "device") == 0)
+	    setp = NULL;
+	  break;
+	case 'i':
+	  if (strcmp (setp, "implementation") == 0)
+	    setp = NULL;
+	  break;
+	case 'u':
+	  if (strcmp (setp, "user") == 0)
+	    setp = NULL;
+	  break;
+	default:
+	  break;
+	}
+      if (setp)
+	{
+	  cp_parser_error (parser, "expected %<construct%>, %<device%>, "
+				   "%<implementation%> or %<user%>");
+	  return error_mark_node;
+	}
+
+      tree set = cp_lexer_peek_token (parser->lexer)->u.value;
+      cp_lexer_consume_token (parser->lexer);
+
+      if (!cp_parser_require (parser, CPP_EQ, RT_EQ))
+	return error_mark_node;
+
+      matching_braces braces;
+      if (!braces.require_open (parser))
+	return error_mark_node;
+
+      tree selectors
+	= cp_parser_omp_context_selector (parser, set, has_parms_p);
+      if (selectors == error_mark_node)
+	{
+	  cp_parser_skip_to_closing_brace (parser);
+	  ret = error_mark_node;
+	}
+      else if (ret != error_mark_node)
+	ret = tree_cons (set, selectors, ret);
+
+      braces.require_close (parser);
+
+      if (cp_lexer_next_token_is (parser->lexer, CPP_COMMA))
+	cp_lexer_consume_token (parser->lexer);
+      else
+	break;
+    }
+  while (1);
+
+  if (ret == error_mark_node)
+    return ret;
+  return nreverse (ret);
+}
+
+/* Finalize #pragma omp declare variant after a fndecl has been parsed, and put
+   that into "omp declare variant" attribute.  */
+
+static tree
+cp_finish_omp_declare_variant (cp_parser *parser, cp_token *pragma_tok,
+			       tree attrs)
+{
+  matching_parens parens;
+  if (!parens.require_open (parser))
+    {
+     fail:
+      cp_parser_skip_to_pragma_eol (parser, pragma_tok);
+      return attrs;
+    }
+
+  cp_token *token = cp_lexer_peek_token (parser->lexer);
+  tree variant;
+  tree name = cp_parser_id_expression (parser, /*template_p=*/false,
+				       /*check_dependency_p=*/true,
+				       /*template_p=*/NULL,
+				       /*declarator_p=*/false,
+				       /*optional_p=*/false);
+  if (identifier_p (name))
+    variant = cp_parser_lookup_name_simple (parser, name, token->location);
+  else
+    variant = name;
+  if (variant == error_mark_node)
+    {
+      cp_parser_name_lookup_error (parser, name, variant, NLE_NULL,
+				   token->location);
+      variant = error_mark_node;
+    }
+
+  parens.require_close (parser);
+
+  const char *clause = "";
+  location_t match_loc = cp_lexer_peek_token (parser->lexer)->location;
+  if (cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+    clause = IDENTIFIER_POINTER (cp_lexer_peek_token (parser->lexer)->u.value);
+  if (strcmp (clause, "match"))
+    {
+      cp_parser_error (parser, "expected %<match%>");
+      goto fail;
+    }
+
+  cp_lexer_consume_token (parser->lexer);
+
+  if (!parens.require_open (parser))
+    goto fail;
+
+  tree ctx = cp_parser_omp_context_selector_specification (parser, true);
+  if (ctx == error_mark_node)
+    goto fail;
+  ctx = c_omp_check_context_selector (match_loc, ctx);
+  if (ctx != error_mark_node && variant != error_mark_node)
+    {
+      attrs = tree_cons (get_identifier ("omp declare variant"),
+			 build_tree_list (variant, ctx), attrs);
+      if (processing_template_decl)
+	ATTR_IS_DEPENDENT (attrs) = 1;
+    }
+
+  parens.require_close (parser);
+  cp_parser_skip_to_pragma_eol (parser, pragma_tok);
+  return attrs;
+}
+
+
 /* Finalize #pragma omp declare simd clauses after direct declarator has
    been parsed, and put that into "omp declare simd" attribute.  */
 
@@ -40309,8 +40741,9 @@ cp_parser_late_parsing_omp_declare_simd (cp_parser *parser, tree attrs)
 
   if (!data->error_seen && data->fndecl_seen)
     {
-      error ("%<#pragma omp declare simd%> not immediately followed by "
-	     "a single function declaration or definition");
+      error ("%<#pragma omp declare %s%> not immediately followed by "
+	     "a single function declaration or definition",
+	     data->variant_p ? "variant" : "simd");
       data->error_seen = true;
     }
   if (data->error_seen)
@@ -40324,17 +40757,28 @@ cp_parser_late_parsing_omp_declare_simd (cp_parser *parser, tree attrs)
       parser->lexer->in_pragma = true;
       gcc_assert (cp_lexer_peek_token (parser->lexer)->type == CPP_PRAGMA);
       cp_token *pragma_tok = cp_lexer_consume_token (parser->lexer);
+      tree id = cp_lexer_peek_token (parser->lexer)->u.value;
+      const char *kind = IDENTIFIER_POINTER (id);
       cp_lexer_consume_token (parser->lexer);
-      cl = cp_parser_omp_all_clauses (parser, OMP_DECLARE_SIMD_CLAUSE_MASK,
-				      "#pragma omp declare simd", pragma_tok);
+      if (strcmp (kind, "simd") == 0)
+	{
+	  cl = cp_parser_omp_all_clauses (parser, OMP_DECLARE_SIMD_CLAUSE_MASK,
+					  "#pragma omp declare simd",
+					  pragma_tok);
+	  if (cl)
+	    cl = tree_cons (NULL_TREE, cl, NULL_TREE);
+	  c = build_tree_list (get_identifier ("omp declare simd"), cl);
+	  TREE_CHAIN (c) = attrs;
+	  if (processing_template_decl)
+	    ATTR_IS_DEPENDENT (c) = 1;
+	  attrs = c;
+	}
+      else
+	{
+	  gcc_assert (strcmp (kind, "variant") == 0);
+	  attrs = cp_finish_omp_declare_variant (parser, pragma_tok, attrs);
+	}
       cp_parser_pop_lexer (parser);
-      if (cl)
-	cl = tree_cons (NULL_TREE, cl, NULL_TREE);
-      c = build_tree_list (get_identifier ("omp declare simd"), cl);
-      TREE_CHAIN (c) = attrs;
-      if (processing_template_decl)
-	ATTR_IS_DEPENDENT (c) = 1;
-      attrs = c;
     }
 
   data->fndecl_seen = true;
@@ -40898,7 +41342,10 @@ cp_parser_omp_declare_reduction (cp_parser *parser, cp_token *pragma_tok,
    #pragma omp declare simd declare-simd-clauses[optseq] new-line
    #pragma omp declare reduction (reduction-id : typename-list : expression) \
       initializer-clause[opt] new-line
-   #pragma omp declare target new-line  */
+   #pragma omp declare target new-line
+
+   OpenMP 5.0
+   #pragma omp declare variant (identifier) match (context-selector)  */
 
 static bool
 cp_parser_omp_declare (cp_parser *parser, cp_token *pragma_tok,
@@ -40913,7 +41360,14 @@ cp_parser_omp_declare (cp_parser *parser, cp_token *pragma_tok,
 	{
 	  cp_lexer_consume_token (parser->lexer);
 	  cp_parser_omp_declare_simd (parser, pragma_tok,
-				      context);
+				      context, false);
+	  return true;
+	}
+      if (flag_openmp && strcmp (p, "variant") == 0)
+	{
+	  cp_lexer_consume_token (parser->lexer);
+	  cp_parser_omp_declare_simd (parser, pragma_tok,
+				      context, true);
 	  return true;
 	}
       cp_ensure_no_omp_declare_simd (parser);
@@ -40936,8 +41390,8 @@ cp_parser_omp_declare (cp_parser *parser, cp_token *pragma_tok,
 	  return false;
 	}
     }
-  cp_parser_error (parser, "expected %<simd%> or %<reduction%> "
-			   "or %<target%>");
+  cp_parser_error (parser, "expected %<simd%>, %<reduction%>, "
+			   "%<target%> or %<variant%>");
   cp_parser_require_pragma_eol (parser, pragma_tok);
   return false;
 }
diff --git a/gcc/cp/parser.h b/gcc/cp/parser.h
index 2890788..91b5916 100644
--- a/gcc/cp/parser.h
+++ b/gcc/cp/parser.h
@@ -202,10 +202,11 @@ struct GTY (()) cp_parser_context {
 };
 
 
-/* Helper data structure for parsing #pragma omp declare simd.  */
+/* Helper data structure for parsing #pragma omp declare {simd,variant}.  */
 struct cp_omp_declare_simd_data {
   bool error_seen; /* Set if error has been reported.  */
   bool fndecl_seen; /* Set if one fn decl/definition has been seen already.  */
+  bool variant_p; /* Set for #pragma omp declare variant.  */
   vec<cp_token_cache_ptr> tokens;
   tree clauses;
 };
diff --git a/gcc/testsuite/ChangeLog b/gcc/testsuite/ChangeLog
index d03827d..0a4961b 100644
--- a/gcc/testsuite/ChangeLog
+++ b/gcc/testsuite/ChangeLog
@@ -1,3 +1,12 @@
+2019-10-10  Jakub Jelinek  <jakub@redhat.com>
+
+	* c-c++-common/gomp/declare-variant-1.c: New test.
+	* c-c++-common/gomp/declare-variant-2.c: New test.
+	* c-c++-common/gomp/declare-variant-3.c: New test.
+	* g++.dg/gomp/this-1.C: Adjust for diagnostic message spelling fix.
+	* gcc.dg/gomp/declare-variant-1.c: New test.
+	* gcc.dg/gomp/declare-variant-2.c: New test.
+
 2019-10-09  Martin Sebor  <msebor@redhat.com>
 
 	PR tree-optimization/90879
diff --git a/gcc/testsuite/c-c++-common/gomp/declare-variant-1.c b/gcc/testsuite/c-c++-common/gomp/declare-variant-1.c
new file mode 100644
index 0000000..8b3cd7f
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/declare-variant-1.c
@@ -0,0 +1,54 @@
+int foo (int, int, int *);
+int bar (int, int, int *);
+#pragma omp declare variant (foo) \
+  match (construct={parallel,for},\
+	 device={isa(avx512f,avx512vl),kind(host,cpu)},\
+	 implementation={vendor(score(0):gnu),unified_shared_memory},\
+	 user={condition(score(0):0)})
+#pragma omp declare variant (bar) \
+  match (device={arch(x86_64,powerpc64),isa(avx512f,popcntb)}, \
+	 implementation={atomic_default_mem_order(seq_cst),made_up_selector("foo", 13, "bar")}, \
+	 user={condition(3-3)})
+int baz (int, int, int *);
+
+int
+qux (void)
+{
+  int i = 3;
+  return baz (1, 2, &i);
+}
+
+int quux (int);
+
+void
+corge (void)
+{
+  int i;
+  #pragma omp declare variant (quux) match (construct={parallel,for})
+  extern int waldo (int);
+  waldo (5);
+  #pragma omp parallel for
+  for (i = 0; i < 3; i++)
+    waldo (6);
+  #pragma omp parallel
+  #pragma omp taskgroup
+  #pragma omp for
+  for (i = 0; i < 3; i++)
+    waldo (7);
+  #pragma omp parallel
+  #pragma omp master    
+  waldo (8);
+}
+
+#pragma omp declare variant (bar) match \
+  (implementation={atomic_default_mem_order(relaxed), \
+		   unified_address, unified_shared_memory, \
+		   dynamic_allocators, reverse_offload})
+int baz2 (int x, int y, int *z)
+{
+  return x + y + *z;
+}
+
+#pragma omp declare variant (bar) match \
+  (implementation={atomic_default_mem_order(score(3): acq_rel)})
+int baz3 (int, int, int *);
diff --git a/gcc/testsuite/c-c++-common/gomp/declare-variant-2.c b/gcc/testsuite/c-c++-common/gomp/declare-variant-2.c
new file mode 100644
index 0000000..bc12398
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/declare-variant-2.c
@@ -0,0 +1,122 @@
+void f1 (void);
+#pragma omp declare variant	/* { dg-error "expected '\\(' before end of line" } */
+void f2 (void);
+#pragma omp declare variant (	/* { dg-error "" } */
+void f3 (void);
+#pragma omp declare variant ()	/* { dg-error "" } */
+void f4 (void);
+#pragma omp declare variant match(user={condition(0)})	/* { dg-error "expected '\\(' before 'match'" } */
+void f5 (void);
+#pragma omp declare variant (f1)	/* { dg-error "expected 'match' before end of line" } */
+void f6 (void);
+#pragma omp declare variant (f1) simd	/* { dg-error "expected 'match' before 'simd'" } */
+void f7 (void);
+#pragma omp declare variant (f1) match	/* { dg-error "expected '\\(' before end of line" } */
+void f8 (void);
+#pragma omp declare variant (f1) match(	/* { dg-error "expected 'construct', 'device', 'implementation' or 'user' before end of line" } */
+void f9 (void);
+#pragma omp declare variant (f1) match()	/* { dg-error "expected 'construct', 'device', 'implementation' or 'user' before '\\)' token" } */
+void f10 (void);
+#pragma omp declare variant (f1) match(foo)	/* { dg-error "expected 'construct', 'device', 'implementation' or 'user' before 'foo'" } */
+void f11 (void);
+#pragma omp declare variant (f1) match(something={something})	/* { dg-error "expected 'construct', 'device', 'implementation' or 'user' before 'something'" } */
+void f12 (void);
+#pragma omp declare variant (f1) match(user)	/* { dg-error "expected '=' before '\\)' token" } */
+void f13 (void);
+#pragma omp declare variant (f1) match(user=)	/* { dg-error "expected '\\\{' before '\\)' token" } */
+void f14 (void);
+#pragma omp declare variant (f1) match(user=	/* { dg-error "expected '\\\{' before end of line" } */
+void f15 (void);
+#pragma omp declare variant (f1) match(user={)	/* { dg-error "expected trait selector name before '\\)' token" } */
+void f16 (void);				/* { dg-error "expected '\\\}' before" "" { target c++ } .-1 } */
+#pragma omp declare variant (f1) match(user={})	/* { dg-error "expected trait selector name before '\\\}' token" } */
+void f17 (void);
+#pragma omp declare variant (f1) match(user={condition})	/* { dg-error "expected '\\(' before '\\\}' token" } */
+void f18 (void);
+#pragma omp declare variant (f1) match(user={condition(})	/* { dg-error "expected \[^\n\r]*expression before '\\\}' token" } */
+void f19 (void);						/* { dg-error "expected '\\)' before '\\\}' token" "" { target c++ } .-1 } */
+#pragma omp declare variant (f1) match(user={condition()})	/* { dg-error "expected \[^\n\r]*expression before '\\)' token" } */
+void f20 (void);
+#pragma omp declare variant (f1) match(user={condition(f1)})	/* { dg-error "property must be constant integer expression" "" { target { c || c++11 } } } */
+void f21 (void);						/* { dg-error "cannot appear in a constant-expression" "" { target c++98_only } .-1 } */
+#pragma omp declare variant (f1) match(user={condition(1, 2, 3)})	/* { dg-error "expected '\\)' before ',' token" } */
+void f22 (void);
+#pragma omp declare variant (f1) match(construct={master})	/* { dg-error "selector 'master' not allowed for context selector set 'construct'" } */
+void f23 (void);
+#pragma omp declare variant (f1) match(construct={teams,parallel,master,for})	/* { dg-error "selector 'master' not allowed for context selector set 'construct'" } */
+void f24 (void);						/* { dg-error "expected '\\\}' before ',' token" "" { target c } .-1 } */
+#pragma omp declare variant (f1) match(construct={parallel(1	/* { dg-error "selector 'parallel' does not accept any properties" } */
+void f25 (void);						/* { dg-error "expected '\\\}' before end of line" "" { target c++ } .-1 } */
+								/* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-2 } */
+#pragma omp declare variant (f1) match(construct={parallel(1)})	/* { dg-error "selector 'parallel' does not accept any properties" } */
+void f26 (void);							/* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */
+#pragma omp declare variant (f1) match(construct={simd(12)})	/* { dg-error "expected \[^\n\r]* clause before" } */
+void f27 (void);						/* { dg-error "'\\)' before numeric constant" "" { target c++ } .-1 } */
+#pragma omp declare variant (f1) match(construct={parallel},construct={for})	/* { dg-error "selector set 'construct' specified more than once" } */
+void f28 (void);
+#pragma omp declare variant (f1) match(construct={parallel},construct={parallel})	/* { dg-error "selector set 'construct' specified more than once" } */
+void f29 (void);
+#pragma omp declare variant (f1) match(user={condition(0)},construct={target},user={condition(0)})	/* { dg-error "selector set 'user' specified more than once" } */
+void f30 (void);
+#pragma omp declare variant (f1) match(user={condition(0)},user={condition(1)})	/* { dg-error "selector set 'user' specified more than once" } */
+void f31 (void);
+#pragma omp declare variant (f1) match(device={kind})	/* { dg-error "expected '\\(' before '\\\}' token" } */
+void f32 (void);
+#pragma omp declare variant (f1) match(device={isa})	/* { dg-error "expected '\\(' before '\\\}' token" } */
+void f33 (void);
+#pragma omp declare variant (f1) match(device={arch})	/* { dg-error "expected '\\(' before '\\\}' token" } */
+void f34 (void);
+#pragma omp declare variant (f1) match(device={kind,isa,arch})	/* { dg-error "expected '\\(' before ',' token" } */
+void f35 (void);
+#pragma omp declare variant (f1) match(device={kind(})	/* { dg-error "expected identifier before '\\\}' token" } */
+void f36 (void);
+#pragma omp declare variant (f1) match(device={kind(unknown)})	/* { dg-warning "unknown property 'unknown' of 'kind' selector" } */
+void f37 (void);
+#pragma omp declare variant (f1) match(device={kind(unknown,foobar)})	/* { dg-warning "unknown property 'unknown' of 'kind' selector" } */
+void f38 (void);							/* { dg-warning "unknown property 'foobar' of 'kind' selector" "" { target *-*-* } .-1 } */
+#pragma omp declare variant (f1) match(device={isa(1)})	/* { dg-error "expected identifier before numeric constant" } */
+void f39 (void);
+#pragma omp declare variant (f1) match(device={arch(17)})	/* { dg-error "expected identifier before numeric constant" } */
+void f40 (void);
+#pragma omp declare variant (f1) match(device={foobar(3)})
+void f41 (void);
+#pragma omp declare variant (f1) match(device={arch(x86_64)},device={isa(avx512vl)})	/* { dg-error "selector set 'device' specified more than once" } */
+void f42 (void);
+#pragma omp declare variant (f1) match(implementation={foobar(3)})
+void f43 (void);
+#pragma omp declare variant (f1) match(implementation={vendor})	/* { dg-error "expected '\\(' before '\\\}' token" } */
+void f44 (void);
+#pragma omp declare variant (f1) match(implementation={extension})	/* { dg-error "expected '\\(' before '\\\}' token" } */
+void f45 (void);
+#pragma omp declare variant (f1) match(implementation={vendor()})	/* { dg-error "expected identifier before '\\)' token" } */
+void f45 (void);
+#pragma omp declare variant (f1) match(implementation={vendor(123-234)})	/* { dg-error "expected identifier before numeric constant" } */
+void f46 (void);
+#pragma omp declare variant (f1) match(implementation={vendor("x86_64")})	/* { dg-error "expected identifier before string constant" } */
+void f47 (void);
+#pragma omp declare variant (f1) match(implementation={unified_address(yes)})	/* { dg-error "selector 'unified_address' does not accept any properties" } */
+void f48 (void);								/* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */
+#pragma omp declare variant (f1) match(implementation={unified_shared_memory(no)})	/* { dg-error "selector 'unified_shared_memory' does not accept any properties" } */
+void f49 (void);									/* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */
+#pragma omp declare variant (f1) match(implementation={dynamic_allocators(42)})	/* { dg-error "selector 'dynamic_allocators' does not accept any properties" } */
+void f50 (void);								/* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */
+#pragma omp declare variant (f1) match(implementation={reverse_offload()})	/* { dg-error "selector 'reverse_offload' does not accept any properties" } */
+void f51 (void);								/* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */
+#pragma omp declare variant (f1) match(implementation={atomic_default_mem_order})	/* { dg-error "expected '\\(' before '\\\}' token" } */
+void f52 (void);
+#pragma omp declare variant (f1) match(implementation={atomic_default_mem_order(acquire)})	/* { dg-error "incorrect property 'acquire' of 'atomic_default_mem_order' selector" } */
+void f53 (void);
+#pragma omp declare variant (f1) match(implementation={atomic_default_mem_order(release)})	/* { dg-error "incorrect property 'release' of 'atomic_default_mem_order' selector" } */
+void f54 (void);
+#pragma omp declare variant (f1) match(implementation={atomic_default_mem_order(foobar)})	/* { dg-error "incorrect property 'foobar' of 'atomic_default_mem_order' selector" } */
+void f55 (void);
+#pragma omp declare variant (f1) match(implementation={atomic_default_mem_order(relaxed,seq_cst)})	/* { dg-error "'atomic_default_mem_order' selector must have a single property" } */
+void f56 (void);
+#pragma omp declare variant (f1) match(implementation={atomic_default_mem_order(relaxed)},implementation={atomic_default_mem_order(relaxed)})	/* { dg-error "selector set 'implementation' specified more than once" } */
+void f57 (void);
+#pragma omp declare variant (f1) match(user={foobar(3)})	/* { dg-error "selector 'foobar' not allowed for context selector set 'user'" } */
+void f58 (void);						/* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */
+#pragma omp declare variant (f1) match(construct={foobar(3)})	/* { dg-error "selector 'foobar' not allowed for context selector set 'construct'" } */
+void f59 (void);						/* { dg-error "expected '\\\}' before '\\(' token" "" { target c } .-1 } */
+#pragma omp declare variant (f1) match(construct={parallel},foobar={bar})	/* { dg-error "expected 'construct', 'device', 'implementation' or 'user' before 'foobar'" } */
+void f60 (void);
diff --git a/gcc/testsuite/c-c++-common/gomp/declare-variant-3.c b/gcc/testsuite/c-c++-common/gomp/declare-variant-3.c
new file mode 100644
index 0000000..34a2a06
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/declare-variant-3.c
@@ -0,0 +1,141 @@
+void f1 (void);
+#pragma omp declare variant (f1) match (construct={target})
+void f2 (void);
+void f3 (void);
+#pragma omp declare variant (f3) match (construct={teams})
+void f4 (void);
+void f5 (void);
+#pragma omp declare variant (f5) match (construct={parallel})
+void f6 (void);
+void f7 (void);
+#pragma omp declare variant (f7) match (construct={for})
+void f8 (void);
+void f9 (void);
+#pragma omp declare variant (f9) match (construct={target,teams,parallel,for})
+void f10 (void);
+void f11 (void);
+#pragma omp declare variant (f11) match (construct={teams,for,parallel})
+void f12 (void);
+void f13 (void);
+#pragma omp declare variant (f13) match (device={kind(any)})
+void f14 (void);
+#pragma omp declare variant (f13) match (device={kind(host)})
+void f15 (void);
+#pragma omp declare variant (f13) match (device={kind(nohost)})
+void f16 (void);
+#pragma omp declare variant (f13) match (device={kind(cpu)})
+void f17 (void);
+#pragma omp declare variant (f13) match (device={kind(gpu)})
+void f18 (void);
+#pragma omp declare variant (f13) match (device={kind(fpga)})
+void f19 (void);
+#pragma omp declare variant (f13) match (device={kind(any,any)})
+void f20 (void);
+#pragma omp declare variant (f13) match (device={kind(host,nohost)})
+void f21 (void);
+#pragma omp declare variant (f13) match (device={kind(cpu,gpu,fpga)})
+void f22 (void);
+#pragma omp declare variant (f13) match (device={kind(any,cpu,nohost)})
+void f23 (void);
+#pragma omp declare variant (f13) match (device={isa(avx)})
+void f24 (void);
+#pragma omp declare variant (f13) match (device={isa(sse4,avx512f,avx512vl,avx512bw)})
+void f25 (void);
+#pragma omp declare variant (f13) match (device={arch(x86_64)})
+void f26 (void);
+#pragma omp declare variant (f13) match (device={arch(riscv64)})
+void f27 (void);
+#pragma omp declare variant (f13) match (device={arch(nvptx)})
+void f28 (void);
+#pragma omp declare variant (f13) match (device={arch(x86_64),isa(avx512f,avx512vl),kind(cpu)})
+void f29 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(amd)})
+void f30 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(arm)})
+void f31 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(bsc)})
+void f32 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(cray)})
+void f33 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(fujitsu)})
+void f34 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(gnu)})
+void f35 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(ibm)})
+void f36 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(intel)})
+void f37 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(llvm)})
+void f38 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(pgi)})
+void f39 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(ti)})
+void f40 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(unknown)})
+void f41 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(gnu,llvm,intel,ibm)})
+void f42 (void);
+#pragma omp declare variant (f13) match (implementation={extension(my_cute_extension)})	/* { dg-warning "unknown property 'my_cute_extension' of 'extension' selector" } */
+void f43 (void);
+#pragma omp declare variant (f13) match (implementation={extension(some_other_ext,another_ext)})	/* { dg-warning "unknown property 'some_other_ext' of 'extension' selector" } */
+void f44 (void);											/* { dg-warning "unknown property 'another_ext' of 'extension' selector" "" { target *-*-* } .-1 } */
+#pragma omp declare variant (f13) match (implementation={unified_shared_memory})
+void f45 (void);
+#pragma omp declare variant (f13) match (implementation={unified_address})
+void f46 (void);
+#pragma omp declare variant (f13) match (implementation={dynamic_allocators})
+void f47 (void);
+#pragma omp declare variant (f13) match (implementation={reverse_offload})
+void f48 (void);
+#pragma omp declare variant (f13) match (implementation={atomic_default_mem_order(seq_cst)})
+void f49 (void);
+#pragma omp declare variant (f13) match (implementation={atomic_default_mem_order(relaxed)})
+void f50 (void);
+#pragma omp declare variant (f13) match (implementation={atomic_default_mem_order(acq_rel)})
+void f51 (void);
+#pragma omp declare variant (f14) match (implementation={atomic_default_mem_order(acq_rel),vendor(gnu),unified_address,extension(foobar)})	/* { dg-warning "unknown property 'foobar' of 'extension' selector" } */
+void f52 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(3):amd)})
+void f53 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(4):arm)})
+void f54 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(5):bsc)})
+void f55 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(6):cray)})
+void f56 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(7):fujitsu)})
+void f57 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(8):gnu)})
+void f58 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(9):ibm)})
+void f59 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(10):intel)})
+void f60 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(11):llvm)})
+void f61 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(12):pgi)})
+void f62 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(13):ti)})
+void f63 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(14):unknown)})
+void f64 (void);
+#pragma omp declare variant (f13) match (implementation={vendor(score(15):gnu,llvm,intel,ibm)})
+void f65 (void);
+#pragma omp declare variant (f13) match (implementation={extension(score(16):my_cute_extension)})	/* { dg-warning "unknown property 'my_cute_extension' of 'extension' selector" } */
+void f66 (void);
+#pragma omp declare variant (f13) match (implementation={extension(score(17):some_other_ext,another_ext)})	/* { dg-warning "unknown property 'some_other_ext' of 'extension' selector" } */
+void f67 (void);												/* { dg-warning "unknown property 'another_ext' of 'extension' selector" "" { target *-*-* } .-1 } */
+#pragma omp declare variant (f13) match (implementation={atomic_default_mem_order(score(18):seq_cst)})
+void f68 (void);
+#pragma omp declare variant (f13) match (implementation={atomic_default_mem_order(score(19):relaxed)})
+void f69 (void);
+#pragma omp declare variant (f13) match (implementation={atomic_default_mem_order(score(20):acq_rel)})
+void f70 (void);
+#pragma omp declare variant (f13) match (implementation={atomic_default_mem_order(score(21):acq_rel),vendor(score(22):gnu),unified_address,extension(score(22):foobar)})	/* { dg-warning "unknown property 'foobar' of 'extension' selector" } */
+void f71 (void);
+#pragma omp declare variant (f13) match (user={condition(0)})
+void f72 (void);
+#pragma omp declare variant (f13) match (user={condition(272-272*1)})
+void f73 (void);
+#pragma omp declare variant (f13) match (user={condition(score(25):1)})
+void f74 (void);
diff --git a/gcc/testsuite/g++.dg/gomp/this-1.C b/gcc/testsuite/g++.dg/gomp/this-1.C
index 0e2dd8c..30ab8b3 100644
--- a/gcc/testsuite/g++.dg/gomp/this-1.C
+++ b/gcc/testsuite/g++.dg/gomp/this-1.C
@@ -3,7 +3,7 @@
 
 struct S
 {
-  #pragma omp declare simd linear(this)		// { dg-error "is not an function argument" }
+  #pragma omp declare simd linear(this)		// { dg-error "is not a function argument" }
   static void foo ();
   void bar ();
 };
@@ -35,7 +35,7 @@ S::bar ()
 template <int N>
 struct T
 {
-  #pragma omp declare simd linear(this)		// { dg-error "is not an function argument" }
+  #pragma omp declare simd linear(this)		// { dg-error "is not a function argument" }
   static void foo ();
   void bar ();
 };
diff --git a/gcc/testsuite/gcc.dg/gomp/declare-variant-1.c b/gcc/testsuite/gcc.dg/gomp/declare-variant-1.c
new file mode 100644
index 0000000..9b20cfe
--- /dev/null
+++ b/gcc/testsuite/gcc.dg/gomp/declare-variant-1.c
@@ -0,0 +1,41 @@
+/* Test parsing of #pragma omp declare variant */
+/* { dg-do compile } */
+
+int fn0 (int);
+int fn6 (int);
+
+#pragma omp declare variant (fn0) match (user={condition(0)})
+int a;	/* { dg-error "not immediately followed by a function declaration or definition" } */
+
+#pragma omp declare variant (fn0) match (user={condition(0)})
+int fn1 (int a), fn2 (int a);	/* { dg-error "not immediately followed by a single function declaration or definition" } */
+
+#pragma omp declare variant (fn0) match (user={condition(0)})
+int b, fn3 (int a);	/* { dg-error "not immediately followed by a function declaration or definition" } */
+
+#pragma omp declare variant (fn0) match (user={condition(0)})
+int fn4 (int a), c;	/* { dg-error "not immediately followed by a function declaration or definition" } */
+
+int t;
+
+#pragma omp declare variant (fn0) match (user={condition(0)})
+#pragma omp declare variant (fn6) match (implementation={vendor(unknown)})
+#pragma omp threadprivate(t)	/* { dg-error "must be followed by function declaration or definition or another" } */
+int fn5 (int a);
+
+#pragma omp declare variant (1 + 2) match (user={condition(0)})	/* { dg-error "expected identifier before numeric constant" } */
+int fn7 (int);
+
+#pragma omp declare variant (t) match (user={condition(0)})	/* { dg-error "variant 't' is not a function" } */
+int fn8 (int);
+
+long fn9 (char, short);
+
+#pragma omp declare variant (fn9) match (implementation={vendor(unknown)})	/* { dg-error "variant 'fn9' and base 'fn10' have incompatible types" } */
+int fn10 (int, long long);
+
+#pragma omp declare variant (memcpy) match (implementation={vendor(llvm)})	/* { dg-error "'memcpy' undeclared here" } */
+void *fn11 (void *, const void *, __SIZE_TYPE__);
+
+#pragma omp declare variant (__builtin_memmove) match (implementation={vendor(gnu)})	/* { dg-error "variant '__builtin_memmove' is a built-in" } */
+void *fn12 (void *, const void *, __SIZE_TYPE__);
diff --git a/gcc/testsuite/gcc.dg/gomp/declare-variant-2.c b/gcc/testsuite/gcc.dg/gomp/declare-variant-2.c
new file mode 100644
index 0000000..701d83b
--- /dev/null
+++ b/gcc/testsuite/gcc.dg/gomp/declare-variant-2.c
@@ -0,0 +1,22 @@
+/* Test parsing of #pragma omp declare variant */
+/* { dg-do compile } */
+
+int f0 (int, int *, int);
+
+int
+f1 (int x)
+{
+  if (x)
+    #pragma omp declare variant (fn0) match (user={condition(0)})
+    extern int f3 (int a, int *b, int c);	/* { dg-error "must be followed by function declaration or definition" } */
+  while (x < 10)
+    #pragma omp declare variant (fn0) match (user={condition(0)})
+    extern int f4 (int a, int *b, int c);	/* { dg-error "must be followed by function declaration or definition" } */
+  {
+lab:
+    #pragma omp declare variant (fn0) match (user={condition(0)})
+    extern int f5 (int a, int *b, int c);	/* { dg-error "must be followed by function declaration or definition" } */
+    x++;					/* { dg-error "expected expression before" "" { target *-*-* } .-1 } */
+  }
+  return x;
+}