[3/7] OpenMP: lvalue parsing for map/to/from clauses (C++)

diff --git a/gcc/c-family/c-common.h b/gcc/c-family/c-common.h
index 82b54ebe74f..6d7cda89aa4 100644
--- a/gcc/c-family/c-common.h
+++ b/gcc/c-family/c-common.h
@@ -1367,8 +1367,8 @@  public:
 
   bool map_supported_p ();
 
-  static tree get_origin (tree);
-  static tree maybe_unconvert_ref (tree);
+  tree get_origin (tree);
+  tree maybe_unconvert_ref (tree);
 
   bool maybe_zero_length_array_section (tree);
 
diff --git a/gcc/c-family/c-omp.cc b/gcc/c-family/c-omp.cc
index 291a26293ef..da0fd3ebe3c 100644
--- a/gcc/c-family/c-omp.cc
+++ b/gcc/c-family/c-omp.cc
@@ -2825,6 +2825,9 @@  c_omp_split_clauses (location_t loc, enum tree_code code,
 		    }
 		  else if (TREE_CODE (OMP_CLAUSE_DECL (c)) == TREE_LIST)
 		    {
+		      /* TODO: This can go away once we transition all uses of
+			 TREE_LIST for representing OMP array sections to
+			 OMP_ARRAY_SECTION.  */
 		      tree t;
 		      for (t = OMP_CLAUSE_DECL (c);
 			   TREE_CODE (t) == TREE_LIST; t = TREE_CHAIN (t))
@@ -2833,6 +2836,17 @@  c_omp_split_clauses (location_t loc, enum tree_code code,
 			bitmap_clear_bit (&allocate_head, DECL_UID (t));
 		      break;
 		    }
+		  else if (TREE_CODE (OMP_CLAUSE_DECL (c)) == OMP_ARRAY_SECTION)
+		    {
+		      tree t;
+		      for (t = OMP_CLAUSE_DECL (c);
+			   TREE_CODE (t) == OMP_ARRAY_SECTION;
+			   t = TREE_OPERAND (t, 0))
+			;
+		      if (DECL_P (t))
+			bitmap_clear_bit (&allocate_head, DECL_UID (t));
+		      break;
+		    }
 		  /* FALLTHRU */
 		case OMP_CLAUSE_PRIVATE:
 		case OMP_CLAUSE_FIRSTPRIVATE:
@@ -4027,6 +4041,7 @@  c_omp_address_inspector::map_supported_p ()
 	 || TREE_CODE (t) == SAVE_EXPR
 	 || TREE_CODE (t) == POINTER_PLUS_EXPR
 	 || TREE_CODE (t) == NON_LVALUE_EXPR
+	 || TREE_CODE (t) == OMP_ARRAY_SECTION
 	 || TREE_CODE (t) == NOP_EXPR)
     if (TREE_CODE (t) == COMPOUND_EXPR)
       t = TREE_OPERAND (t, 1);
@@ -4056,7 +4071,8 @@  c_omp_address_inspector::get_origin (tree t)
       else if (TREE_CODE (t) == POINTER_PLUS_EXPR
 	       || TREE_CODE (t) == SAVE_EXPR)
 	t = TREE_OPERAND (t, 0);
-      else if (TREE_CODE (t) == INDIRECT_REF
+      else if (!processing_template_decl_p ()
+	       && TREE_CODE (t) == INDIRECT_REF
 	       && TREE_CODE (TREE_TYPE (TREE_OPERAND (t, 0))) == REFERENCE_TYPE)
 	t = TREE_OPERAND (t, 0);
       else
@@ -4073,7 +4089,10 @@  c_omp_address_inspector::get_origin (tree t)
 tree
 c_omp_address_inspector::maybe_unconvert_ref (tree t)
 {
-  if (TREE_CODE (t) == INDIRECT_REF
+  /* Be careful not to dereference the type if we're processing a
+     template decl, else it might be NULL.  */
+  if (!processing_template_decl_p ()
+      && TREE_CODE (t) == INDIRECT_REF
       && TREE_CODE (TREE_TYPE (TREE_OPERAND (t, 0))) == REFERENCE_TYPE)
     return TREE_OPERAND (t, 0);
 
diff --git a/gcc/cp/constexpr.cc b/gcc/cp/constexpr.cc
index 56329dcc9e0..c015afc3e07 100644
--- a/gcc/cp/constexpr.cc
+++ b/gcc/cp/constexpr.cc
@@ -9593,6 +9593,7 @@  potential_constant_expression_1 (tree t, bool want_rval, bool strict, bool now,
     case OACC_ENTER_DATA:
     case OACC_EXIT_DATA:
     case OACC_UPDATE:
+    case OMP_ARRAY_SECTION:
       /* GCC internal stuff.  */
     case VA_ARG_EXPR:
     case TRANSACTION_EXPR:
diff --git a/gcc/cp/cp-tree.h b/gcc/cp/cp-tree.h
index b87039490dd..cffe92f9046 100644
--- a/gcc/cp/cp-tree.h
+++ b/gcc/cp/cp-tree.h
@@ -6949,6 +6949,7 @@  extern void grokclassfn				(tree, tree,
 						 enum overload_flags);
 extern tree grok_array_decl			(location_t, tree, tree,
 						 vec<tree, va_gc> **, tsubst_flags_t);
+extern tree grok_omp_array_section		(location_t, tree, tree, tree);
 extern tree delete_sanity			(location_t, tree, tree, bool,
 						 int, tsubst_flags_t);
 extern tree check_classfn			(tree, tree, tree);
@@ -8049,6 +8050,7 @@  inline tree build_x_binary_op (const op_location_t &loc,
 }
 extern tree build_x_array_ref			(location_t, tree, tree,
 						 tsubst_flags_t);
+extern tree build_omp_array_section		(location_t, tree, tree, tree);
 extern tree build_x_unary_op			(location_t,
 						 enum tree_code, cp_expr,
 						 tree, tsubst_flags_t);
diff --git a/gcc/cp/decl2.cc b/gcc/cp/decl2.cc
index f1f17d4751c..407f7b090e9 100644
--- a/gcc/cp/decl2.cc
+++ b/gcc/cp/decl2.cc
@@ -615,6 +615,51 @@  grok_array_decl (location_t loc, tree array_expr, tree index_exp,
   return expr;
 }
 
+/* Build an OMP_ARRAY_SECTION expression, handling usage in template
+   definitions, etc.  */
+
+tree
+grok_omp_array_section (location_t loc, tree array_expr, tree index,
+			tree length)
+{
+  tree orig_array_expr = array_expr;
+  tree orig_index = index;
+  tree orig_length = length;
+
+  if (error_operand_p (array_expr)
+      || error_operand_p (index)
+      || error_operand_p (length))
+    return error_mark_node;
+
+  if (processing_template_decl)
+    {
+      if (type_dependent_expression_p (array_expr)
+	  || type_dependent_expression_p (index)
+	  || type_dependent_expression_p (length))
+	return build_min_nt_loc (loc, OMP_ARRAY_SECTION, array_expr, index,
+				 length);
+      array_expr = build_non_dependent_expr (array_expr);
+      if (index)
+	index = build_non_dependent_expr (index);
+      if (length)
+	length = build_non_dependent_expr (length);
+    }
+
+  index = fold_non_dependent_expr (index);
+  length = fold_non_dependent_expr (length);
+
+  /* NOTE: We can pass through invalidly-typed index/length fields
+     here (e.g. if the user tries to use a floating-point index/length).
+     This is diagnosed later in semantics.cc:handle_omp_array_sections_1.  */
+
+  tree expr = build_omp_array_section (loc, array_expr, index, length);
+
+  if (processing_template_decl)
+    expr = build_min_non_dep (OMP_ARRAY_SECTION, expr, orig_array_expr,
+			      orig_index, orig_length);
+  return expr;
+}
+
 /* Given the cast expression EXP, checking out its validity.   Either return
    an error_mark_node if there was an unavoidable error, return a cast to
    void for trying to delete a pointer w/ the value 0, or return the
diff --git a/gcc/cp/error.cc b/gcc/cp/error.cc
index a5d888926a6..ad2e871d372 100644
--- a/gcc/cp/error.cc
+++ b/gcc/cp/error.cc
@@ -2497,6 +2497,15 @@  dump_expr (cxx_pretty_printer *pp, tree t, int flags)
       pp_cxx_right_bracket (pp);
       break;
 
+    case OMP_ARRAY_SECTION:
+      dump_expr (pp, TREE_OPERAND (t, 0), flags);
+      pp_cxx_left_bracket (pp);
+      dump_expr (pp, TREE_OPERAND (t, 1), flags);
+      pp_colon (pp);
+      dump_expr (pp, TREE_OPERAND (t, 2), flags);
+      pp_cxx_right_bracket (pp);
+      break;
+
     case UNARY_PLUS_EXPR:
       dump_unary_op (pp, "+", t, flags);
       break;
diff --git a/gcc/cp/parser.cc b/gcc/cp/parser.cc
index 1e5504b83b8..471431ce393 100644
--- a/gcc/cp/parser.cc
+++ b/gcc/cp/parser.cc
@@ -4377,6 +4377,9 @@  cp_parser_new (cp_lexer *lexer)
   parser->omp_declare_simd = NULL;
   parser->oacc_routine = NULL;
 
+  /* Disallow OpenMP array sections in expressions.  */
+  parser->omp_array_section_p = false;
+
   /* Not declaring an implicit function template.  */
   parser->auto_is_implicit_function_template_parm_p = false;
   parser->fully_implicit_function_template_p = false;
@@ -5361,6 +5364,7 @@  static cp_expr
 cp_parser_statement_expr (cp_parser *parser)
 {
   cp_token_position start = cp_parser_start_tentative_firewall (parser);
+  auto oas = make_temp_override (parser->omp_array_section_p, false);
 
   /* Consume the '('.  */
   location_t start_loc = cp_lexer_peek_token (parser->lexer)->location;
@@ -8193,6 +8197,7 @@  cp_parser_postfix_open_square_expression (cp_parser *parser,
   releasing_vec expression_list = NULL;
   location_t loc = cp_lexer_peek_token (parser->lexer)->location;
   bool saved_greater_than_is_operator_p;
+  bool saved_colon_corrects_to_scope_p;
 
   /* Consume the `[' token.  */
   cp_lexer_consume_token (parser->lexer);
@@ -8200,6 +8205,10 @@  cp_parser_postfix_open_square_expression (cp_parser *parser,
   saved_greater_than_is_operator_p = parser->greater_than_is_operator_p;
   parser->greater_than_is_operator_p = true;
 
+  saved_colon_corrects_to_scope_p = parser->colon_corrects_to_scope_p;
+  if (parser->omp_array_section_p)
+    parser->colon_corrects_to_scope_p = false;
+
   /* Parse the index expression.  */
   /* ??? For offsetof, there is a question of what to allow here.  If
      offsetof is not being used in an integral constant expression context,
@@ -8210,7 +8219,8 @@  cp_parser_postfix_open_square_expression (cp_parser *parser,
      constant expressions here.  */
   if (for_offsetof)
     index = cp_parser_constant_expression (parser);
-  else
+  else if (!parser->omp_array_section_p
+	   || cp_lexer_next_token_is_not (parser->lexer, CPP_COLON))
     {
       if (cxx_dialect >= cxx23
 	  && cp_lexer_next_token_is (parser->lexer, CPP_CLOSE_SQUARE))
@@ -8267,6 +8277,68 @@  cp_parser_postfix_open_square_expression (cp_parser *parser,
 
   parser->greater_than_is_operator_p = saved_greater_than_is_operator_p;
 
+  if (cxx_dialect >= cxx23
+      && parser->omp_array_section_p
+      && expression_list.get () != NULL
+      && vec_safe_length (expression_list) > 1)
+    {
+      error_at (loc, "cannot use multidimensional subscript in OpenMP array "
+		"section");
+      index = error_mark_node;
+    }
+  if (parser->omp_array_section_p
+      && cp_lexer_next_token_is (parser->lexer, CPP_COLON))
+    {
+      cp_lexer_consume_token (parser->lexer);
+      tree length = NULL_TREE;
+      if (cp_lexer_next_token_is_not (parser->lexer, CPP_CLOSE_SQUARE))
+	{
+	  if (cxx_dialect >= cxx23)
+	    {
+	      cp_expr expr
+		= cp_parser_parenthesized_expression_list_elt (parser,
+							       /*cast_p=*/
+							       false,
+							       /*allow_exp_p=*/
+							       true,
+							       /*non_cst_p=*/
+							       NULL);
+
+	      if (expr == error_mark_node)
+		length = error_mark_node;
+	      else
+		length = expr.get_value ();
+
+	      if (cp_lexer_next_token_is (parser->lexer, CPP_COMMA))
+		{
+		  error_at (loc, "cannot use multidimensional subscript in "
+			    "OpenMP array section");
+		  length = error_mark_node;
+		}
+	    }
+	  else
+	    length
+	      = cp_parser_expression (parser, NULL, /*cast_p=*/false,
+				      /*decltype_p=*/false,
+				      /*warn_comma_p=*/warn_comma_subscript);
+	}
+
+      parser->colon_corrects_to_scope_p = saved_colon_corrects_to_scope_p;
+
+      if (index == error_mark_node || length == error_mark_node)
+	{
+	  cp_parser_skip_to_closing_square_bracket (parser);
+	  return error_mark_node;
+	}
+      else
+	cp_parser_require (parser, CPP_CLOSE_SQUARE, RT_CLOSE_SQUARE);
+
+      return grok_omp_array_section (input_location, postfix_expression, index,
+				     length);
+    }
+
+  parser->colon_corrects_to_scope_p = saved_colon_corrects_to_scope_p;
+
   /* Look for the closing `]'.  */
   cp_parser_require (parser, CPP_CLOSE_SQUARE, RT_CLOSE_SQUARE);
 
@@ -8595,6 +8667,7 @@  cp_parser_parenthesized_expression_list (cp_parser* parser,
 {
   vec<tree, va_gc> *expression_list;
   bool saved_greater_than_is_operator_p;
+  bool saved_omp_array_section_p;
 
   /* Assume all the expressions will be constant.  */
   if (non_constant_p)
@@ -8612,6 +8685,9 @@  cp_parser_parenthesized_expression_list (cp_parser* parser,
     = parser->greater_than_is_operator_p;
   parser->greater_than_is_operator_p = true;
 
+  saved_omp_array_section_p = parser->omp_array_section_p;
+  parser->omp_array_section_p = false;
+
   cp_expr expr (NULL_TREE);
 
   /* Consume expressions until there are no more.  */
@@ -8676,12 +8752,14 @@  cp_parser_parenthesized_expression_list (cp_parser* parser,
 	{
 	  parser->greater_than_is_operator_p
 	    = saved_greater_than_is_operator_p;
+	  parser->omp_array_section_p = saved_omp_array_section_p;
 	  return NULL;
 	}
     }
 
   parser->greater_than_is_operator_p
     = saved_greater_than_is_operator_p;
+  parser->omp_array_section_p = saved_omp_array_section_p;
 
   return expression_list;
 }
@@ -11179,6 +11257,7 @@  cp_parser_lambda_expression (cp_parser* parser)
     cp_binding_level* implicit_template_scope = parser->implicit_template_scope;
     bool auto_is_implicit_function_template_parm_p
         = parser->auto_is_implicit_function_template_parm_p;
+    bool saved_omp_array_section_p = parser->omp_array_section_p;
 
     parser->num_template_parameter_lists = 0;
     parser->in_statement = 0;
@@ -11187,6 +11266,7 @@  cp_parser_lambda_expression (cp_parser* parser)
     parser->implicit_template_parms = 0;
     parser->implicit_template_scope = 0;
     parser->auto_is_implicit_function_template_parm_p = false;
+    parser->omp_array_section_p = false;
 
     /* The body of a lambda in a discarded statement is not discarded.  */
     bool discarded = in_discarded_stmt;
@@ -11237,6 +11317,7 @@  cp_parser_lambda_expression (cp_parser* parser)
     parser->implicit_template_scope = implicit_template_scope;
     parser->auto_is_implicit_function_template_parm_p
 	= auto_is_implicit_function_template_parm_p;
+    parser->omp_array_section_p = saved_omp_array_section_p;
   }
 
   /* This field is only used during parsing of the lambda.  */
@@ -25663,6 +25744,7 @@  cp_parser_braced_list (cp_parser* parser, bool* non_constant_p)
 {
   tree initializer;
   location_t start_loc = cp_lexer_peek_token (parser->lexer)->location;
+  auto oas = make_temp_override (parser->omp_array_section_p, false);
 
   /* Consume the `{' token.  */
   matching_braces braces;
@@ -37572,7 +37654,7 @@  static tree
 cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind,
 				tree list, bool *colon,
 				enum c_omp_region_type ort = C_ORT_OMP,
-				bool allow_deref = false)
+				bool map_lvalue = false)
 {
   auto_vec<omp_dim> dims;
   bool array_section_p;
@@ -37589,6 +37671,105 @@  cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind,
 
       if (kind == OMP_CLAUSE_DEPEND || kind == OMP_CLAUSE_AFFINITY)
 	cp_parser_parse_tentatively (parser);
+      /* This condition doesn't include OMP_CLAUSE_DEPEND or
+	 OMP_CLAUSE_AFFINITY since lvalue ("locator list") parsing for those is
+	 handled further down the function.  */
+      else if (map_lvalue
+	       && (kind == OMP_CLAUSE_MAP
+		   || kind == OMP_CLAUSE_TO
+		   || kind == OMP_CLAUSE_FROM))
+	{
+	  auto s = make_temp_override (parser->omp_array_section_p, true);
+	  token = cp_lexer_peek_token (parser->lexer);
+	  location_t loc = token->location;
+	  decl = cp_parser_assignment_expression (parser);
+
+	  /* This code rewrites a parsed expression containing various tree
+	     codes used to represent array accesses into a more uniform nest of
+	     OMP_ARRAY_SECTION nodes before it is processed by
+	     semantics.cc:handle_omp_array_sections_1.  It might be more
+	     efficient to move this logic to that function instead, analysing
+	     the parsed expression directly rather than this preprocessed
+	     form.  */
+	  dims.truncate (0);
+	  if (TREE_CODE (decl) == OMP_ARRAY_SECTION)
+	    {
+	      while (TREE_CODE (decl) == OMP_ARRAY_SECTION)
+		{
+		  tree low_bound = TREE_OPERAND (decl, 1);
+		  tree length = TREE_OPERAND (decl, 2);
+		  dims.safe_push (omp_dim (low_bound, length, loc, false));
+		  decl = TREE_OPERAND (decl, 0);
+		}
+
+	      while (TREE_CODE (decl) == ARRAY_REF
+		     || TREE_CODE (decl) == INDIRECT_REF
+		     || TREE_CODE (decl) == COMPOUND_EXPR)
+		{
+		  if (REFERENCE_REF_P (decl))
+		    break;
+
+		  if (TREE_CODE (decl) == COMPOUND_EXPR)
+		    {
+		      decl = TREE_OPERAND (decl, 1);
+		      STRIP_NOPS (decl);
+		    }
+		  else if (TREE_CODE (decl) == INDIRECT_REF)
+		    {
+		      dims.safe_push (omp_dim (integer_zero_node,
+					       integer_one_node, loc, true));
+		      decl = TREE_OPERAND (decl, 0);
+		    }
+		  else  /* ARRAY_REF. */
+		    {
+		      tree index = TREE_OPERAND (decl, 1);
+		      dims.safe_push (omp_dim (index, integer_one_node, loc,
+					       true));
+		      decl = TREE_OPERAND (decl, 0);
+		    }
+		}
+
+	      /* Bare references have their own special handling, so remove
+		 the explicit dereference added by convert_from_reference.  */
+	      if (REFERENCE_REF_P (decl))
+		decl = TREE_OPERAND (decl, 0);
+
+	      for (int i = dims.length () - 1; i >= 0; i--)
+		decl = grok_omp_array_section (loc, decl, dims[i].low_bound,
+					       dims[i].length);
+	    }
+	  else if (TREE_CODE (decl) == INDIRECT_REF)
+	    {
+	      bool ref_p = REFERENCE_REF_P (decl);
+
+	      /* Turn *foo into foo[0:1].  */
+	      decl = TREE_OPERAND (decl, 0);
+	      STRIP_NOPS (decl);
+
+	      /* If we have "*foo" and
+		 - it's an indirection of a reference, "unconvert" it, i.e.
+		   strip the indirection (to just "foo").
+		 - it's an indirection of a pointer, turn it into
+		   "foo[0:1]".  */
+	      if (!ref_p)
+		decl = grok_omp_array_section (loc, decl, integer_zero_node,
+					       integer_one_node);
+	    }
+	  else if (TREE_CODE (decl) == ARRAY_REF)
+	    {
+	      tree idx = TREE_OPERAND (decl, 1);
+
+	      decl = TREE_OPERAND (decl, 0);
+	      STRIP_NOPS (decl);
+
+	      decl = grok_omp_array_section (loc, decl, idx, integer_one_node);
+	    }
+	  else if (TREE_CODE (decl) == NON_LVALUE_EXPR
+		   || CONVERT_EXPR_P (decl))
+	    decl = TREE_OPERAND (decl, 0);
+
+	  goto build_clause;
+	}
       token = cp_lexer_peek_token (parser->lexer);
       if (kind != 0
 	  && cp_parser_is_keyword (token, RID_THIS))
@@ -37667,8 +37848,7 @@  cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind,
 	    case OMP_CLAUSE_TO:
 	    start_component_ref:
 	      while (cp_lexer_next_token_is (parser->lexer, CPP_DOT)
-		     || (allow_deref
-			 && cp_lexer_next_token_is (parser->lexer, CPP_DEREF)))
+		     || cp_lexer_next_token_is (parser->lexer, CPP_DEREF))
 		{
 		  cpp_ttype ttype
 		    = cp_lexer_next_token_is (parser->lexer, CPP_DOT)
@@ -37755,9 +37935,7 @@  cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind,
 		   || kind == OMP_CLAUSE_TO)
 		  && !array_section_p
 		  && (cp_lexer_next_token_is (parser->lexer, CPP_DOT)
-		      || (allow_deref
-			  && cp_lexer_next_token_is (parser->lexer,
-						     CPP_DEREF))))
+		      || cp_lexer_next_token_is (parser->lexer, CPP_DEREF)))
 		{
 		  for (unsigned i = 0; i < dims.length (); i++)
 		    {
@@ -37769,8 +37947,9 @@  cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind,
 		}
 	      else
 		for (unsigned i = 0; i < dims.length (); i++)
-		  decl = tree_cons (dims[i].low_bound, dims[i].length, decl);
-
+		  decl = build_omp_array_section (input_location, decl,
+						  dims[i].low_bound,
+						  dims[i].length);
 	      break;
 	    default:
 	      break;
@@ -37791,6 +37970,7 @@  cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind,
 		cp_parser_parse_definitely (parser);
 	    }
 
+	build_clause:
 	  tree u = build_omp_clause (token->location, kind);
 	  OMP_CLAUSE_DECL (u) = decl;
 	  OMP_CLAUSE_CHAIN (u) = list;
@@ -37841,11 +38021,11 @@  cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind,
 static tree
 cp_parser_omp_var_list (cp_parser *parser, enum omp_clause_code kind, tree list,
 			enum c_omp_region_type ort = C_ORT_OMP,
-			bool allow_deref = false)
+			bool map_lvalue = false)
 {
   if (cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN))
     return cp_parser_omp_var_list_no_open (parser, kind, list, NULL, ort,
-					   allow_deref);
+					   map_lvalue);
   return list;
 }
 
@@ -37912,7 +38092,7 @@  cp_parser_oacc_data_clause (cp_parser *parser, pragma_omp_clause c_kind,
       gcc_unreachable ();
     }
   tree nl, c;
-  nl = cp_parser_omp_var_list (parser, OMP_CLAUSE_MAP, list, C_ORT_ACC, true);
+  nl = cp_parser_omp_var_list (parser, OMP_CLAUSE_MAP, list, C_ORT_ACC, false);
 
   for (c = nl; c != list; c = OMP_CLAUSE_CHAIN (c))
     OMP_CLAUSE_SET_MAP_KIND (c, kind);
@@ -41081,8 +41261,13 @@  cp_parser_omp_clause_map (cp_parser *parser, tree list)
       cp_lexer_consume_token (parser->lexer);
     }
 
+  /* We introduce a scope here so that errors parsing e.g. "always", "close"
+     tokens do not propagate to later directives that might use them
+     legally.  */
+  begin_scope (sk_omp, NULL);
   nlist = cp_parser_omp_var_list_no_open (parser, OMP_CLAUSE_MAP, list,
 					  NULL, C_ORT_OMP, true);
+  finish_scope ();
 
   for (c = nlist; c != list; c = OMP_CLAUSE_CHAIN (c))
     OMP_CLAUSE_SET_MAP_KIND (c, kind);
diff --git a/gcc/cp/parser.h b/gcc/cp/parser.h
index 6cbb9a8e031..40009a42e96 100644
--- a/gcc/cp/parser.h
+++ b/gcc/cp/parser.h
@@ -407,6 +407,9 @@  struct GTY(()) cp_parser {
   /* TRUE if omp::directive or omp::sequence attributes may not appear.  */
   bool omp_attrs_forbidden_p;
 
+  /* TRUE if an OpenMP array section is allowed.  */
+  bool omp_array_section_p;
+
   /* Tracks the function's template parameter list when declaring a function
      using generic type parameters.  This is either a new chain in the case of a
      fully implicit function template or an extension of the function's existing
diff --git a/gcc/cp/pt.cc b/gcc/cp/pt.cc
index 2d8d5d16ad5..bb1902b2fe5 100644
--- a/gcc/cp/pt.cc
+++ b/gcc/cp/pt.cc
@@ -16727,6 +16727,7 @@  tsubst (tree t, tree args, tsubst_flags_t complain, tree in_decl)
     case CALL_EXPR:
     case ARRAY_REF:
     case SCOPE_REF:
+    case OMP_ARRAY_SECTION:
       /* We should use one of the expression tsubsts for these codes.  */
       gcc_unreachable ();
 
@@ -17745,6 +17746,17 @@  tsubst_copy (tree t, tree args, tsubst_flags_t complain, tree in_decl)
 	return build_nt (ARRAY_REF, op0, op1, NULL_TREE, NULL_TREE);
       }
 
+    case OMP_ARRAY_SECTION:
+      {
+	tree op0 = tsubst_copy (TREE_OPERAND (t, 0), args, complain, in_decl);
+	tree op1 = NULL_TREE, op2 = NULL_TREE;
+	if (TREE_OPERAND (t, 1))
+	  op1 = tsubst_copy (TREE_OPERAND (t, 1), args, complain, in_decl);
+	if (TREE_OPERAND (t, 2))
+	  op2 = tsubst_copy (TREE_OPERAND (t, 2), args, complain, in_decl);
+	return build_nt (OMP_ARRAY_SECTION, op0, op1, op2);
+      }
+
     case CALL_EXPR:
       {
 	int n = VL_EXP_OPERAND_LENGTH (t);
@@ -18013,6 +18025,22 @@  tsubst_omp_clause_decl (tree decl, tree args, tsubst_flags_t complain,
 	= OMP_CLAUSE_DOACROSS_SINK_NEGATIVE (decl);
       return ret;
     }
+  else if (TREE_CODE (decl) == OMP_ARRAY_SECTION)
+    {
+      tree low_bound
+	= tsubst_expr (TREE_OPERAND (decl, 1), args, complain, in_decl);
+      tree length = tsubst_expr (TREE_OPERAND (decl, 2), args, complain,
+				 in_decl);
+      tree base = tsubst_omp_clause_decl (TREE_OPERAND (decl, 0), args,
+					   complain, in_decl, NULL);
+      if (TREE_OPERAND (decl, 0) == base
+	  && TREE_OPERAND (decl, 1) == low_bound
+	  && TREE_OPERAND (decl, 2) == length)
+	return decl;
+      tree ret = build3 (OMP_ARRAY_SECTION, TREE_TYPE (base), base, low_bound,
+			 length);
+      return ret;
+    }
   tree ret = tsubst_expr (decl, args, complain, in_decl);
   /* Undo convert_from_reference tsubst_expr could have called.  */
   if (decl
@@ -20795,6 +20823,27 @@  tsubst_copy_and_build (tree t,
 				 RECUR (TREE_OPERAND (t, 1)),
 				 complain|decltype_flag));
 
+    case OMP_ARRAY_SECTION:
+      {
+	tree op0 = RECUR (TREE_OPERAND (t, 0));
+	tree op1 = NULL_TREE, op2 = NULL_TREE;
+	if (op0 == error_mark_node)
+	  RETURN (error_mark_node);
+	if (TREE_OPERAND (t, 1))
+	  {
+	    op1 = RECUR (TREE_OPERAND (t, 1));
+	    if (op1 == error_mark_node)
+	      RETURN (error_mark_node);
+	  }
+	if (TREE_OPERAND (t, 2))
+	  {
+	    op2 = RECUR (TREE_OPERAND (t, 2));
+	    if (op2 == error_mark_node)
+	      RETURN (error_mark_node);
+	  }
+	RETURN (build_omp_array_section (EXPR_LOCATION (t), op0, op1, op2));
+      }
+
     case SIZEOF_EXPR:
       if (PACK_EXPANSION_P (TREE_OPERAND (t, 0))
 	  || ARGUMENT_PACK_P (TREE_OPERAND (t, 0)))
diff --git a/gcc/cp/semantics.cc b/gcc/cp/semantics.cc
index cd62495b31b..f84cd328098 100644
--- a/gcc/cp/semantics.cc
+++ b/gcc/cp/semantics.cc
@@ -5183,7 +5183,7 @@  handle_omp_array_sections_1 (tree c, tree t, vec<tree> &types,
 {
   tree ret, low_bound, length, type;
   bool openacc = (ort & C_ORT_ACC) != 0;
-  if (TREE_CODE (t) != TREE_LIST)
+  if (TREE_CODE (t) != OMP_ARRAY_SECTION)
     {
       if (error_operand_p (t))
 	return error_mark_node;
@@ -5205,7 +5205,9 @@  handle_omp_array_sections_1 (tree c, tree t, vec<tree> &types,
       ret = t_refto;
       if (TREE_CODE (t) == FIELD_DECL)
 	ret = finish_non_static_data_member (t, NULL_TREE, NULL_TREE);
-      else if (!VAR_P (t) && TREE_CODE (t) != PARM_DECL)
+      else if (!VAR_P (t)
+	       && (openacc || !EXPR_P (t))
+	       && TREE_CODE (t) != PARM_DECL)
 	{
 	  if (processing_template_decl && TREE_CODE (t) != OVERLOAD)
 	    return NULL_TREE;
@@ -5238,17 +5240,17 @@  handle_omp_array_sections_1 (tree c, tree t, vec<tree> &types,
       && (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
 	  || OMP_CLAUSE_CODE (c) == OMP_CLAUSE_IN_REDUCTION
 	  || OMP_CLAUSE_CODE (c) == OMP_CLAUSE_TASK_REDUCTION)
-      && TREE_CODE (TREE_CHAIN (t)) == FIELD_DECL)
-    TREE_CHAIN (t) = omp_privatize_field (TREE_CHAIN (t), false);
-  ret = handle_omp_array_sections_1 (c, TREE_CHAIN (t), types,
+      && TREE_CODE (TREE_OPERAND (t, 0)) == FIELD_DECL)
+    TREE_OPERAND (t, 0) = omp_privatize_field (TREE_OPERAND (t, 0), false);
+  ret = handle_omp_array_sections_1 (c, TREE_OPERAND (t, 0), types,
 				     maybe_zero_len, first_non_one,
 				     non_contiguous, ort);
   if (ret == error_mark_node || ret == NULL_TREE)
     return ret;
 
   type = TREE_TYPE (ret);
-  low_bound = TREE_PURPOSE (t);
-  length = TREE_VALUE (t);
+  low_bound = TREE_OPERAND (t, 1);
+  length = TREE_OPERAND (t, 2);
   if ((low_bound && type_dependent_expression_p (low_bound))
       || (length && type_dependent_expression_p (length)))
     return NULL_TREE;
@@ -5458,7 +5460,7 @@  handle_omp_array_sections_1 (tree c, tree t, vec<tree> &types,
 	  tree lb = cp_save_expr (low_bound);
 	  if (lb != low_bound)
 	    {
-	      TREE_PURPOSE (t) = lb;
+	      TREE_OPERAND (t, 1) = lb;
 	      low_bound = lb;
 	    }
 	}
@@ -5489,20 +5491,20 @@  handle_omp_array_sections_1 (tree c, tree t, vec<tree> &types,
 	 array-section-subscript, the array section could be non-contiguous.  */
       if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_AFFINITY
 	  && OMP_CLAUSE_CODE (c) != OMP_CLAUSE_DEPEND
-	  && TREE_CODE (TREE_CHAIN (t)) == TREE_LIST)
+	  && TREE_CODE (TREE_OPERAND (t, 0)) == OMP_ARRAY_SECTION)
 	{
 	  /* If any prior dimension has a non-one length, then deem this
 	     array section as non-contiguous.  */
-	  for (tree d = TREE_CHAIN (t); TREE_CODE (d) == TREE_LIST;
-	       d = TREE_CHAIN (d))
+	  for (tree d = TREE_OPERAND (t, 0); TREE_CODE (d) == OMP_ARRAY_SECTION;
+	       d = TREE_OPERAND (d, 0))
 	    {
-	      tree d_length = TREE_VALUE (d);
+	      tree d_length = TREE_OPERAND (d, 2);
 	      if (d_length == NULL_TREE || !integer_onep (d_length))
 		{
 		  if (openacc && OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP)
 		    {
-		      while (TREE_CODE (d) == TREE_LIST)
-			d = TREE_CHAIN (d);
+		      while (TREE_CODE (d) == OMP_ARRAY_SECTION)
+			d = TREE_OPERAND (d, 0);
 		      if (DECL_P (d))
 			{
 			  /* Note that OpenACC does accept these kinds of
@@ -5537,7 +5539,7 @@  handle_omp_array_sections_1 (tree c, tree t, vec<tree> &types,
   tree lb = cp_save_expr (low_bound);
   if (lb != low_bound)
     {
-      TREE_PURPOSE (t) = lb;
+      TREE_OPERAND (t, 1) = lb;
       low_bound = lb;
     }
   /* Temporarily disable -fstrong-eval-order for array reductions.
@@ -5617,10 +5619,12 @@  handle_omp_array_sections (tree &c, enum c_omp_region_type ort)
 	return false;
 
       for (i = num, t = OMP_CLAUSE_DECL (c); i > 0;
-	   t = TREE_CHAIN (t))
+	   t = TREE_OPERAND (t, 0))
 	{
-	  tree low_bound = TREE_PURPOSE (t);
-	  tree length = TREE_VALUE (t);
+	  gcc_assert (TREE_CODE (t) == OMP_ARRAY_SECTION);
+
+	  tree low_bound = TREE_OPERAND (t, 1);
+	  tree length = TREE_OPERAND (t, 2);
 
 	  if (length)
 	    STRIP_NOPS (length);
@@ -6752,8 +6756,8 @@  cp_oacc_check_attachments (tree c)
       tree t = OMP_CLAUSE_DECL (c);
       tree type;
 
-      while (TREE_CODE (t) == TREE_LIST)
-	t = TREE_CHAIN (t);
+      while (TREE_CODE (t) == OMP_ARRAY_SECTION)
+	t = TREE_OPERAND (t, 0);
 
       type = TREE_TYPE (t);
 
@@ -6861,7 +6865,7 @@  finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 	case OMP_CLAUSE_TASK_REDUCTION:
 	  field_ok = ((ort & C_ORT_OMP_DECLARE_SIMD) == C_ORT_OMP);
 	  t = OMP_CLAUSE_DECL (c);
-	  if (TREE_CODE (t) == TREE_LIST)
+	  if (TREE_CODE (t) == OMP_ARRAY_SECTION)
 	    {
 	      if (handle_omp_array_sections (c, ort))
 		{
@@ -6877,10 +6881,10 @@  finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 		  remove = true;
 		  break;
 		}
-	      if (TREE_CODE (t) == TREE_LIST)
+	      if (TREE_CODE (t) == OMP_ARRAY_SECTION)
 		{
-		  while (TREE_CODE (t) == TREE_LIST)
-		    t = TREE_CHAIN (t);
+		  while (TREE_CODE (t) == OMP_ARRAY_SECTION)
+		    t = TREE_OPERAND (t, 0);
 		}
 	      else
 		{
@@ -8009,7 +8013,7 @@  finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 	  else
 	    last_iterators = NULL_TREE;
 
-	  if (TREE_CODE (t) == TREE_LIST)
+	  if (TREE_CODE (t) == OMP_ARRAY_SECTION)
 	    {
 	      if (handle_omp_array_sections (c, ort))
 		remove = true;
@@ -8169,7 +8173,7 @@  finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 	    auto_vec<omp_addr_token *, 10> addr_tokens;
 
 	    t = OMP_CLAUSE_DECL (c);
-	    if (TREE_CODE (t) == TREE_LIST)
+	    if (TREE_CODE (t) == OMP_ARRAY_SECTION)
 	      {
 		grp_start_p = pc;
 		grp_sentinel = OMP_CLAUSE_CHAIN (c);
@@ -8179,7 +8183,7 @@  finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 		else
 		  {
 		    t = OMP_CLAUSE_DECL (c);
-		    if (TREE_CODE (t) != TREE_LIST
+		    if (TREE_CODE (t) != OMP_ARRAY_SECTION
 			&& !type_dependent_expression_p (t)
 			&& !omp_mappable_type (TREE_TYPE (t)))
 		      {
@@ -8360,7 +8364,8 @@  finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 		if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP
 		    && (OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_POINTER
 			|| OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_ALWAYS_POINTER
-			|| OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_ATTACH_DETACH))
+			|| OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_ATTACH_DETACH
+			|| (!openacc && EXPR_P (t))))
 		  break;
 		if (DECL_P (t))
 		  error_at (OMP_CLAUSE_LOCATION (c),
@@ -8758,15 +8763,15 @@  finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 
 	case OMP_CLAUSE_HAS_DEVICE_ADDR:
 	  t = OMP_CLAUSE_DECL (c);
-	  if (TREE_CODE (t) == TREE_LIST)
+	  if (TREE_CODE (t) == OMP_ARRAY_SECTION)
 	    {
 	      if (handle_omp_array_sections (c, ort))
 		remove = true;
 	      else
 		{
 		  t = OMP_CLAUSE_DECL (c);
-		  while (TREE_CODE (t) == TREE_LIST)
-		    t = TREE_CHAIN (t);
+		  while (TREE_CODE (t) == OMP_ARRAY_SECTION)
+		    t = TREE_OPERAND (t, 0);
 		  while (TREE_CODE (t) == INDIRECT_REF
 			 || TREE_CODE (t) == ARRAY_REF)
 		    t = TREE_OPERAND (t, 0);
@@ -9231,10 +9236,10 @@  finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 		  if (DECL_P (t))
 		    bitmap_clear_bit (&aligned_head, DECL_UID (t));
 		}
-	      else if (TREE_CODE (t) == TREE_LIST)
+	      else if (TREE_CODE (t) == OMP_ARRAY_SECTION)
 		{
-		  while (TREE_CODE (t) == TREE_LIST)
-		    t = TREE_CHAIN (t);
+		  while (TREE_CODE (t) == OMP_ARRAY_SECTION)
+		    t = TREE_OPERAND (t, 0);
 		  if (DECL_P (t))
 		    bitmap_clear_bit (&aligned_head, DECL_UID (t));
 		  t = OMP_CLAUSE_DECL (c);
diff --git a/gcc/cp/typeck.cc b/gcc/cp/typeck.cc
index ff3a49f01b3..6dcbfedf03e 100644
--- a/gcc/cp/typeck.cc
+++ b/gcc/cp/typeck.cc
@@ -4786,6 +4786,56 @@  build_x_array_ref (location_t loc, tree arg1, tree arg2,
   return expr;
 }
 
+/* Build an OpenMP array section reference, creating an exact type for the
+   resulting expression based on the element type and bounds if possible.  If
+   we have variable bounds, create an incomplete array type for the result
+   instead.  */
+
+tree
+build_omp_array_section (location_t loc, tree array_expr, tree index,
+			 tree length)
+{
+  tree idxtype;
+
+  /* If we know the integer bounds, create an index type with exact
+     low/high (or zero/length) bounds.  Otherwise, create an incomplete
+     array type.  (This mostly only affects diagnostics.)  */
+  if (index != NULL_TREE
+      && length != NULL_TREE
+      && TREE_CODE (index) == INTEGER_CST
+      && TREE_CODE (length) == INTEGER_CST)
+    {
+      tree low = fold_convert (sizetype, index);
+      tree high = fold_convert (sizetype, length);
+      high = size_binop (PLUS_EXPR, low, high);
+      high = size_binop (MINUS_EXPR, high, size_one_node);
+      idxtype = build_range_type (sizetype, low, high);
+    }
+  else if ((index == NULL_TREE || integer_zerop (index))
+	   && length != NULL_TREE
+	   && TREE_CODE (length) == INTEGER_CST)
+    idxtype = build_index_type (length);
+  else
+    idxtype = NULL_TREE;
+
+  tree type = TREE_TYPE (array_expr);
+  gcc_assert (type);
+  type = non_reference (type);
+
+  tree sectype, eltype = TREE_TYPE (type);
+
+  /* It's not an array or pointer type.  Just reuse the type of the
+     original expression as the type of the array section (an error will be
+     raised anyway, later).  */
+  if (eltype == NULL_TREE)
+    sectype = TREE_TYPE (array_expr);
+  else
+    sectype = build_array_type (eltype, idxtype);
+
+  return build3_loc (loc, OMP_ARRAY_SECTION, sectype, array_expr, index,
+		     length);
+}
+
 /* Return whether OP is an expression of enum type cast to integer
    type.  In C++ even unsigned enum types are cast to signed integer
    types.  We do not want to issue warnings about comparisons between
diff --git a/gcc/gimplify.cc b/gcc/gimplify.cc
index 0e856b903ec..05f4d0d6f6a 100644
--- a/gcc/gimplify.cc
+++ b/gcc/gimplify.cc
@@ -18067,6 +18067,9 @@  gimplify_expr (tree *expr_p, gimple_seq *pre_p, gimple_seq *post_p,
 	case TREE_LIST:
 	  gcc_unreachable ();
 
+	case OMP_ARRAY_SECTION:
+	  gcc_unreachable ();
+
 	case COMPOUND_EXPR:
 	  ret = gimplify_compound_expr (expr_p, pre_p, fallback != fb_none);
 	  break;
diff --git a/gcc/testsuite/c-c++-common/gomp/map-6.c b/gcc/testsuite/c-c++-common/gomp/map-6.c
index 5152d9d7c21..014ed35ab41 100644
--- a/gcc/testsuite/c-c++-common/gomp/map-6.c
+++ b/gcc/testsuite/c-c++-common/gomp/map-6.c
@@ -30,12 +30,12 @@  foo (void)
 
 
   #pragma omp target map (close a) /* { dg-error "'close' undeclared" "" { target c } } */ 
-  /* { dg-error "'close' has not been declared" "" { target c++ } .-1 } */ 
+  /* { dg-error "'close' was not declared in this scope" "" { target c++ } .-1 } */ 
   /* { dg-error "expected '\\)' before 'a'" "" { target *-*-* } .-2 } */
   ;
 
   #pragma omp target map (always a) /* { dg-error "'always' undeclared" "" { target c } } */
-  /* { dg-error "'always' has not been declared" "" { target c++ } .-1 } */ 
+  /* { dg-error "'always' was not declared in this scope" "" { target c++ } .-1 } */ 
   /* { dg-error "expected '\\)' before 'a'" "" { target *-*-* } .-2 } */
   ;
 
diff --git a/gcc/testsuite/g++.dg/gomp/array-section-1.C b/gcc/testsuite/g++.dg/gomp/array-section-1.C
new file mode 100644
index 00000000000..023706b15c5
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/array-section-1.C
@@ -0,0 +1,38 @@ 
+// { dg-do compile }
+// { dg-additional-options "-fdump-tree-original" }
+
+int x;
+
+template<int C, int D>
+void foo()
+{
+  int arr1[40];
+#pragma omp target map(arr1[x ? C : D])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[SAVE_EXPR <x != 0 \? 3 : 5>\] \[len: [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: \(long int\) &arr1\[SAVE_EXPR <x != 0 \? 3 : 5>\] - \(long int\) &arr1\]\)} "original" } }
+  { }
+#pragma omp target map(arr1[x ? C : D : D])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[SAVE_EXPR <x != 0 \? 3 : 5>\] \[len: [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: \(long int\) &arr1\[SAVE_EXPR <x != 0 \? 3 : 5>\] - \(long int\) &arr1\]\)} "original" } }
+  { }
+#pragma omp target map(arr1[1 : x ? C : D])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[1\] \[len: x != 0 \? [0-9]+ : [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: [0-9]+\]\)} "original" } }
+  { }
+}
+
+int main()
+{
+  int arr1[40];
+#pragma omp target map(arr1[x ? 3 : 5])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[SAVE_EXPR <x != 0 \? 3 : 5>\] \[len: [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: \(long int\) &arr1\[SAVE_EXPR <x != 0 \? 3 : 5>\] - \(long int\) &arr1\]\)} "original" } }
+  { }
+#pragma omp target map(arr1[x ? 3 : 5 : 5])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[SAVE_EXPR <x != 0 \? 3 : 5>\] \[len: [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: \(long int\) &arr1\[SAVE_EXPR <x != 0 \? 3 : 5>\] - \(long int\) &arr1\]\)} "original" } }
+  { }
+#pragma omp target map(arr1[1 : x ? 3 : 5])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[1\] [len: x != 0 ? [0-9]+ : [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: [0-9]+\]\)} "original" } }
+  { }
+
+  foo<3, 5> ();
+
+  return 0;
+}
+
diff --git a/gcc/testsuite/g++.dg/gomp/array-section-2.C b/gcc/testsuite/g++.dg/gomp/array-section-2.C
new file mode 100644
index 00000000000..072108d1f89
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/array-section-2.C
@@ -0,0 +1,63 @@ 
+// { dg-do compile }
+// { dg-additional-options "-fdump-tree-original" }
+
+int x, y;
+
+class C {
+  int x, y;
+
+public:
+  int foo();
+};
+
+int C::foo()
+{
+  int arr1[40];
+  /* There is a parsing ambiguity here without the space.  We don't try to
+     resolve that automatically (though maybe we could, in theory).  */
+#pragma omp target map(arr1[::x: ::y])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[SAVE_EXPR <x>\] \[len: \(sizetype\) y \* [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: \(long int\) &arr1\[SAVE_EXPR <x>\] - \(long int\) &arr1\]\)} "original" } }
+  { }
+#pragma omp target map(arr1[::x:])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[SAVE_EXPR <x>\] \[len: \(40 - \(sizetype\) SAVE_EXPR <x>\) \* [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: \(long int\) &arr1\[SAVE_EXPR <x>\] - \(long int\) &arr1\]\)} "original" } }
+  { }
+#pragma omp target map(arr1[: ::y])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[0\] \[len: \(sizetype\) y \* [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: 0\]\)} "original" } }
+  { }
+  return ::x + ::y;
+}
+
+template<typename T>
+class Ct {
+  T x, y;
+
+public:
+  void foo();
+};
+
+template<typename T>
+void Ct<T>::foo()
+{
+  int arr1[40];
+#pragma omp target map(arr1[::x: ::y])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[SAVE_EXPR <x>\] \[len: \(sizetype\) y \* [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: \(long int\) &arr1\[SAVE_EXPR <x>\] - \(long int\) &arr1\]\)} "original" } }
+  { }
+#pragma omp target map(arr1[::x:])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[SAVE_EXPR <x>\] \[len: \(40 - \(sizetype\) SAVE_EXPR <x>\) \* [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: \(long int\) &arr1\[SAVE_EXPR <x>\] - \(long int\) &arr1\]\)} "original" } }
+  { }
+#pragma omp target map(arr1[: ::y])
+// { dg-final { scan-tree-dump {map\(tofrom:arr1\[0\] \[len: \(sizetype\) y \* [0-9]+\]\) map\(firstprivate:arr1 \[pointer assign, bias: 0\]\)} "original" } }
+  { }
+}
+
+int main()
+{
+  C c;
+  Ct<int> ct;
+
+  c.foo ();
+  ct.foo ();
+
+  return 0;
+}
+
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-1.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-1.C
new file mode 100644
index 00000000000..7e7e9580206
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-1.C
@@ -0,0 +1,35 @@ 
+// { dg-do compile }
+
+int foo (int *ptr);
+
+template<typename T>
+T baz (T *ptr);
+
+template<typename T>
+void bar()
+{
+  T arr[20];
+
+#pragma omp target map(baz(arr[3:5]))
+// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 }
+// { dg-error {expected '\)' before ':' token} "" { target *-*-* } .-2 }
+// { dg-error {expected '\)' before '\]' token} "" { target *-*-* } .-3 }
+// { dg-error {expected an OpenMP clause before '\]' token} "" { target *-*-* } .-4 }
+  { }
+}
+
+int main()
+{
+  int arr[20];
+  // Reject array section as function argument.
+#pragma omp target map(foo(arr[3:5]))
+// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 }
+// { dg-error {expected '\)' before ':' token} "" { target *-*-* } .-2 }
+// { dg-error {expected '\)' before '\]' token} "" { target *-*-* } .-3 }
+// { dg-error {expected an OpenMP clause before '\]' token} "" { target *-*-* } .-4 }
+  { }
+
+  bar<int> ();
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-10.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-10.C
new file mode 100644
index 00000000000..393b0fefe51
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-10.C
@@ -0,0 +1,35 @@ 
+// { dg-do compile }
+
+template<int C, int D>
+void foo()
+{
+  int arr1[40];
+#pragma omp target map(arr1[4,C:])
+// { dg-warning "top-level comma expression in array subscript is deprecated" "" { target c++20_only } .-1 }
+  { }
+#pragma omp target map(arr1[4,5:C,7])
+// { dg-warning "top-level comma expression in array subscript is deprecated" "" { target c++20_only } .-1 }
+  { }
+#pragma omp target map(arr1[:8,C,10])
+// { dg-warning "top-level comma expression in array subscript is deprecated" "" { target c++20_only } .-1 }
+  { }
+}
+
+int main()
+{
+  int arr1[40];
+#pragma omp target map(arr1[4,5:])
+// { dg-warning "top-level comma expression in array subscript is deprecated" "" { target c++20_only } .-1 }
+  { }
+#pragma omp target map(arr1[4,5:6,7])
+// { dg-warning "top-level comma expression in array subscript is deprecated" "" { target c++20_only } .-1 }
+  { }
+#pragma omp target map(arr1[:8,9,10])
+// { dg-warning "top-level comma expression in array subscript is deprecated" "" { target c++20_only } .-1 }
+  { }
+
+  foo<6, 9> ();
+
+  return 0;
+}
+
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-11.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-11.C
new file mode 100644
index 00000000000..dea3b4428f0
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-11.C
@@ -0,0 +1,36 @@ 
+// { dg-do compile }
+// { dg-additional-options "-std=c++23" }
+
+template<int C, int D>
+void foo()
+{
+  int arr1[40];
+#pragma omp target map(arr1[4,C:])
+// { dg-error "cannot use multidimensional subscript in OpenMP array section" "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr1[4,5:C,7])
+// { dg-error "cannot use multidimensional subscript in OpenMP array section" "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr1[:8,C,10])
+// { dg-error "cannot use multidimensional subscript in OpenMP array section" "" { target *-*-* } .-1 }
+  { }
+}
+
+int main()
+{
+  int arr1[40];
+#pragma omp target map(arr1[4,5:])
+// { dg-error "cannot use multidimensional subscript in OpenMP array section" "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr1[4,5:6,7])
+// { dg-error "cannot use multidimensional subscript in OpenMP array section" "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr1[:8,9,10])
+// { dg-error "cannot use multidimensional subscript in OpenMP array section" "" { target *-*-* } .-1 }
+  { }
+
+  foo<6, 9> ();
+
+  return 0;
+}
+
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-2.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-2.C
new file mode 100644
index 00000000000..811d1fee5a0
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-2.C
@@ -0,0 +1,33 @@ 
+// { dg-do compile }
+// { dg-additional-options "-std=c++11" }
+
+template<typename T>
+void foo()
+{
+  T arr[20];
+  // Reject array section in lambda function.
+#pragma omp target map([&](const int x) -> T* { return arr[0:x]; } (5))
+// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 }
+// { dg-error {invalid conversion from 'int' to 'int\*'} "" { target *-*-* } .-2 }
+// { dg-error {expected ';' before ':' token} "" { target *-*-* } .-3 }
+// { dg-error {expected primary-expression before ':' token} "" { target *-*-* } .-4 }
+// { dg-message {sorry, unimplemented: unsupported map expression '<lambda closure object>foo<int>\(\)::<lambda\(int\)>\{arr\}.foo<int>\(\)::<lambda\(int\)>\(5\)'} "" { target *-*-* } .-5 }
+  { }
+}
+
+int main()
+{
+  int arr[20];
+  // Reject array section in lambda function.
+#pragma omp target map([&](const int x) -> int* { return arr[0:x]; } (5))
+// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 }
+// { dg-error {invalid conversion from 'int' to 'int\*'} "" { target *-*-* } .-2 }
+// { dg-error {expected ';' before ':' token} "" { target *-*-* } .-3 }
+// { dg-error {expected primary-expression before ':' token} "" { target *-*-* } .-4 }
+// { dg-message {sorry, unimplemented: unsupported map expression '<lambda closure object>main\(\)::<lambda\(int\)>\{arr\}.main\(\)::<lambda\(int\)>\(5\)'} "" { target *-*-* } .-5 }
+  { }
+
+  foo<int> ();
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-3.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-3.C
new file mode 100644
index 00000000000..d1f067af2e9
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-3.C
@@ -0,0 +1,28 @@ 
+// { dg-do compile }
+
+template<typename T>
+void foo()
+{
+  T arr[20];
+  // Reject array section in statement expression.
+#pragma omp target map( ({ int x = 5; arr[0:x]; }) )
+// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 }
+// { dg-error {expected ';' before ':' token} "" { target *-*-* } .-2 }
+// { dg-message {sorry, unimplemented: unsupported map expression '\(\{\.\.\.\}\)'} "" { target *-*-* } .-3 }
+  { }
+}
+
+int main()
+{
+  int arr[20];
+  // Reject array section in statement expression.
+#pragma omp target map( ({ int x = 5; arr[0:x]; }) )
+// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 }
+// { dg-error {expected ';' before ':' token} "" { target *-*-* } .-2 }
+// { dg-message {sorry, unimplemented: unsupported map expression '\(\{\.\.\.\}\)'} "" { target *-*-* } .-3 }
+  { }
+
+  foo<int> ();
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-4.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-4.C
new file mode 100644
index 00000000000..707c2c31cb2
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-4.C
@@ -0,0 +1,50 @@ 
+// { dg-do compile }
+
+template<typename T>
+struct St {
+  T *ptr;
+};
+
+template<typename T>
+void foo()
+{
+  T arr[20];
+
+  // Reject array section in compound initialiser.
+#pragma omp target map( (struct St<T>) { .ptr = (T *) arr[5:5] } )
+// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 }
+// { dg-error {expected primary-expression before 'struct'} "" { target *-*-* } .-2 }
+// { dg-error {expected '\)' before 'struct'} "" { target *-*-* } .-3 }
+  { }
+
+  // ...and this is unsupported too (probably not useful anyway).
+#pragma omp target map( (struct St<T>) { .ptr = &arr[5] } )
+// { dg-message {sorry, unimplemented: unsupported map expression 'St<int>\{\(\& arr\[5\]\)\}'} "" { target *-*-* } .-1 }
+  { }
+}
+
+struct S {
+  int *ptr;
+};
+
+int main()
+{
+  int arr[20];
+
+  // Reject array section in compound initialiser.
+#pragma omp target map( (struct S) { .ptr = (int *) arr[5:5] } )
+// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 }
+// { dg-warning {cast to pointer from integer of different size} "" { target *-*-* } .-2 }
+// { dg-error {expected primary-expression before 'struct'} "" { target *-*-* } .-3 }
+// { dg-error {expected '\)' before 'struct'} "" { target *-*-* } .-4 }
+  { }
+
+  // ...and this is unsupported too (probably not useful anyway).
+#pragma omp target map( (struct S) { .ptr = &arr[5] } )
+// { dg-message {sorry, unimplemented: unsupported map expression 'S\{\(\& arr\[5\]\)\}'} "" { target *-*-* } .-1 }
+  { }
+
+  foo<int> ();
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-5.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-5.C
new file mode 100644
index 00000000000..f9c27d48f0e
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-5.C
@@ -0,0 +1,50 @@ 
+// { dg-do compile }
+
+int x;
+
+template<typename T>
+void foo()
+{
+  T arr[20];
+  T *ptr;
+  /* "arr[1:10]" looks like it might be an expression of array type, hence
+     able to be indexed (again).  This isn't allowed, though.  */
+#pragma omp target map(arr[1:10][2])
+// { dg-error {'arr\[1\]' does not have pointer or array type} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr[1:x][2])
+// { dg-error {'arr\[1\]' does not have pointer or array type} "" { target *-*-* } .-1 }
+  { }
+  // ...and nor is this.
+#pragma omp target map(ptr[1:10][2])
+// { dg-error {'\*\(ptr \+ [0-9]+\)' does not have pointer or array type} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(ptr[1:x][2])
+// { dg-error {'\*\(ptr \+ [0-9]+\)' does not have pointer or array type} "" { target *-*-* } .-1 }
+  { }
+}
+
+int main()
+{
+  int arr[20];
+  int *ptr;
+  /* "arr[1:10]" looks like it might be an expression of array type, hence
+     able to be indexed (again).  This isn't allowed, though.  */
+#pragma omp target map(arr[1:10][2])
+// { dg-error {'arr\[1\]' does not have pointer or array type} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr[1:x][2])
+// { dg-error {'arr\[1\]' does not have pointer or array type} "" { target *-*-* } .-1 }
+  { }
+  // ...and nor is this.
+#pragma omp target map(ptr[1:10][2])
+// { dg-error {'\*\(ptr \+ [0-9]+\)' does not have pointer or array type} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(ptr[1:x][2])
+// { dg-error {'\*\(ptr \+ [0-9]+\)' does not have pointer or array type} "" { target *-*-* } .-1 }
+  { }
+
+  foo<int> ();
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-6.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-6.C
new file mode 100644
index 00000000000..418ee80431f
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-6.C
@@ -0,0 +1,24 @@ 
+// { dg-do compile }
+
+bool partly = false;
+
+template<typename T>
+void foo()
+{
+  T arr[20];
+#pragma omp target map(partly ? arr[5:5] : arr)
+// { dg-message {sorry, unimplemented: unsupported map expression '\(partly \? \(\(int\*\)\(\& arr\[5:5\]\)\) : \(\(int\*\)\(\& arr\)\)\)'} "" { target *-*-* } .-1 }
+  { }
+}
+
+int main()
+{
+  int arr[20];
+#pragma omp target map(partly ? arr[5:5] : arr)
+// { dg-message {sorry, unimplemented: unsupported map expression '\(partly \? \(\(int\*\)\(\& arr\[5:5\]\)\) : \(\(int\*\)\(\& arr\)\)\)'} "" { target *-*-* } .-1 }
+  { }
+
+  foo<int> ();
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-7.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-7.C
new file mode 100644
index 00000000000..24ac165e2bd
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-7.C
@@ -0,0 +1,36 @@ 
+// { dg-do compile }
+
+int x;
+
+template<typename T>
+void foo()
+{
+  T arr[20];
+  // Here we know the type of the array section (the upper bound is reported)...
+#pragma omp target map(arr[5:5] * 2)
+// { dg-error {invalid operands of types 'int \[10\]' and 'int'} "" { target *-*-* } .-1 }
+  { }
+  /* ...but here, we have an incomplete array type because of the variable
+     low bound 'x'.  */
+#pragma omp target map(arr[x:5] * 2)
+// { dg-error {invalid operands of types 'int \[\]' and 'int'} "" { target *-*-* } .-1 }
+  { }
+}
+
+int main()
+{
+  int arr[20];
+  // Here we know the type of the array section (the upper bound is reported)...
+#pragma omp target map(arr[5:5] * 2)
+// { dg-error {invalid operands of types 'int \[10\]' and 'int'} "" { target *-*-* } .-1 }
+  { }
+  /* ...but here, we have an incomplete array type because of the variable
+     low bound 'x'.  */
+#pragma omp target map(arr[x:5] * 2)
+// { dg-error {invalid operands of types 'int \[\]' and 'int'} "" { target *-*-* } .-1 }
+  { }
+
+  foo<int> ();
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-8.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-8.C
new file mode 100644
index 00000000000..2353722a581
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-8.C
@@ -0,0 +1,53 @@ 
+// { dg-do compile }
+
+int x;
+
+template<typename X>
+struct Tt {
+  X arr[20];
+};
+
+template<typename X>
+struct St {
+  X *tvec;
+};
+
+template<typename T>
+void foo()
+{
+  struct St<Tt<T> > *s;
+  // You can't use an array section like this.  Make sure sensible errors are
+  // reported.
+#pragma omp target map(s->tvec[3:5].arr[0:20])
+// { dg-error {request for member 'arr' in 's->St<Tt<int> >::tvec\[3:5\]', which is of non-class type 'Tt<int> \[8\]'} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(s->tvec[5:x].arr[0:20])
+// { dg-error {invalid use of array with unspecified bounds} "" { target *-*-* } .-1 }
+  { }
+}
+
+struct T {
+  int arr[20];
+};
+
+struct S {
+  struct T *tvec;
+};
+
+int main()
+{
+  struct S *s;
+  // You can't use an array section like this.  Make sure sensible errors are
+  // reported.
+#pragma omp target map(s->tvec[3:5].arr[0:20])
+// { dg-error {request for member 'arr' in 's->S::tvec\[3:5\]', which is of non-class type 'T \[8\]'} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(s->tvec[5:x].arr[0:20])
+// { dg-error {invalid use of array with unspecified bounds} "" { target *-*-* } .-1 }
+// { dg-error {expected '\)' before 'arr'} "" { target *-*-* } .-2 }
+  { }
+
+  foo<int> ();
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-9.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-9.C
new file mode 100644
index 00000000000..bba7772a3c9
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-9.C
@@ -0,0 +1,39 @@ 
+// { dg-do compile }
+
+int x;
+
+template<typename T>
+void foo()
+{
+  T arr1[40];
+  T arr2[40];
+#pragma omp target map(arr1[arr2[4:5]:arr2[6:7]])
+// { dg-error {low bound 'arr2\[4:5\]' of array section does not have integral type} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr1[arr2[:1]:arr2[6:1]])
+// { dg-error {low bound 'arr2\[:1\]' of array section does not have integral type} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr1[x:arr2[6:1]])
+// { dg-error {length 'arr2\[6:1\]' of array section does not have integral type} "" { target *-*-* } .-1 }
+  { }
+}
+
+int main()
+{
+  int arr1[40];
+  int arr2[40];
+#pragma omp target map(arr1[arr2[4:5]:arr2[6:7]])
+// { dg-error {low bound 'arr2\[4:5\]' of array section does not have integral type} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr1[arr2[:1]:arr2[6:1]])
+// { dg-error {low bound 'arr2\[:1\]' of array section does not have integral type} "" { target *-*-* } .-1 }
+  { }
+#pragma omp target map(arr1[x:arr2[6:1]])
+// { dg-error {length 'arr2\[6:1\]' of array section does not have integral type} "" { target *-*-* } .-1 }
+  { }
+
+  foo<int> ();
+
+  return 0;
+}
+
diff --git a/gcc/testsuite/g++.dg/gomp/has_device_addr-non-lvalue-1.C b/gcc/testsuite/g++.dg/gomp/has_device_addr-non-lvalue-1.C
new file mode 100644
index 00000000000..3d778538d3a
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/has_device_addr-non-lvalue-1.C
@@ -0,0 +1,36 @@ 
+// { dg-do compile }
+
+#include <cstdio>
+#include <cstring>
+#include <cassert>
+
+typedef struct {
+  int arr[100];
+} S;
+
+int main()
+{
+  S *s = new S;
+
+  memset (s->arr, '\0', sizeof s->arr); 
+
+#pragma omp target enter data map(to: (*s).arr)
+  /* You can't do this, at least as of OpenMP 5.2.  "has_device_addr" takes
+     a "variable list" item type
+     (OpenMP 5.2, "5.4.9 has_device_addr Clause").  */
+#pragma omp target has_device_addr((*s).arr[5:20])
+// { dg-error {expected unqualified-id before '\(' token} "" { target *-*-* } .-1 }
+  {
+    for (int i = 5; i < 25; i++)
+      s->arr[i] = i; 
+  }
+
+#pragma omp target exit data map(from: (*s).arr)
+
+  for (int i = 0; i < 100; i++)
+    assert (i >= 5 && i < 25 ? s->arr[i] == i : s->arr[i] == 0);
+
+  delete s;
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/ind-base-3.C b/gcc/testsuite/g++.dg/gomp/ind-base-3.C
new file mode 100644
index 00000000000..7695b1f907e
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/ind-base-3.C
@@ -0,0 +1,37 @@ 
+#include <cassert>
+
+struct S {
+  int x[10];
+};
+
+S *
+choose (S *a, S *b, int c)
+{
+  if (c < 5)
+    return a;
+  else
+    return b; 
+}
+
+int main (int argc, char *argv[])
+{
+  S a, b;
+
+  for (int i = 0; i < 10; i++)
+    a.x[i] = b.x[i] = 0;
+
+  for (int i = 0; i < 10; i++)
+    {
+#pragma omp target map(choose(&a, &b, i)->x[:10])
+/* { dg-message {sorry, unimplemented: unsupported map expression 'choose\(\(& a\), \(& b\), i\)->S::x\[0\]'} "" { target *-*-* } .-1 } */
+      for (int j = 0; j < 10; j++)
+        choose (&a, &b, i)->x[j]++;
+    }
+
+  for (int i = 0; i < 10; i++)
+    assert (a.x[i] == 5 && b.x[i] == 5);
+
+  return 0;
+}
+
+
diff --git a/gcc/testsuite/g++.dg/gomp/map-assignment-1.C b/gcc/testsuite/g++.dg/gomp/map-assignment-1.C
new file mode 100644
index 00000000000..5979ec379f1
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/map-assignment-1.C
@@ -0,0 +1,12 @@ 
+#include <cassert>
+
+int main (int argc, char *argv[])
+{
+  int a = 5, b = 2;
+#pragma omp target map(a += b)
+  /* { dg-message {sorry, unimplemented: unsupported map expression '\(a = \(a \+ b\)\)'} "" { target *-*-* } .-1 } */
+  {
+    a++;
+  }
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/map-inc-1.C b/gcc/testsuite/g++.dg/gomp/map-inc-1.C
new file mode 100644
index 00000000000..b469a4bd548
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/map-inc-1.C
@@ -0,0 +1,10 @@ 
+int main (int argc, char *argv[])
+{
+  int a = 5;
+#pragma omp target map(++a)
+  /* { dg-message {sorry, unimplemented: unsupported map expression '\+\+ a'} "" { target *-*-* } .-1 } */
+  {
+    a++;
+  }
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/map-lvalue-ref-1.C b/gcc/testsuite/g++.dg/gomp/map-lvalue-ref-1.C
new file mode 100644
index 00000000000..d720d4318ae
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/map-lvalue-ref-1.C
@@ -0,0 +1,19 @@ 
+#include <cassert>
+
+int glob = 10;
+
+int& foo ()
+{
+  return glob;
+}
+
+int main (int argc, char *argv[])
+{
+#pragma omp target map(foo())
+  /* { dg-message {sorry, unimplemented: unsupported map expression 'foo\(\)'} "" { target *-*-* } .-1 } */
+  {
+    foo()++;
+  }
+  assert (glob == 11);
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/map-ptrmem-1.C b/gcc/testsuite/g++.dg/gomp/map-ptrmem-1.C
new file mode 100644
index 00000000000..c4023f59fc6
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/map-ptrmem-1.C
@@ -0,0 +1,37 @@ 
+#include <cassert>
+
+struct S {
+  int x;
+  int *ptr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  S s;
+  int S::* xp = &S::x;
+  int* S::* ptrp = &S::ptr;
+
+  s.ptr = new int[64];
+
+  s.*xp = 6;
+  for (int i = 0; i < 64; i++)
+    (s.*ptrp)[i] = i;
+
+#pragma omp target map(s.*xp, s.*ptrp, (s.*ptrp)[:64])
+  /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\*\(\(\(int\*\*\)\(& s\)\) \+ \(\(sizetype\)ptrp\)\)\)' not supported} "" { target *-*-* } .-1 } */
+  /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\(\(int\*\*\)\(& s\)\) \+ \(\(sizetype\)ptrp\)\)' not supported} "" { target *-*-* } .-2 } */
+  /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\(\(int\*\)\(& s\)\) \+ \(\(sizetype\)xp\)\)' not supported} "" { target *-*-* } .-3 } */
+#pragma omp teams distribute parallel for
+  for (int i = 0; i < 64; i++)
+    {
+      (s.*xp)++;
+      (s.*ptrp)[i]++;
+    }
+
+  assert (s.*xp == 70);
+  for (int i = 0; i < 64; i++)
+    assert ((s.*ptrp)[i] == i + 1);
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C b/gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C
new file mode 100644
index 00000000000..fbf379da0eb
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C
@@ -0,0 +1,40 @@ 
+#include <cassert>
+
+struct S {
+  int x;
+  int *ptr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  S *s = new S;
+  int S::* xp = &S::x;
+  int* S::* ptrp = &S::ptr;
+
+  s->ptr = new int[64];
+
+  s->*xp = 4;
+  for (int i = 0; i < 64; i++)
+    (s->*ptrp)[i] = i;
+
+#pragma omp target map(s->*xp, s->*ptrp, (s->*ptrp)[:64])
+  /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\(\(int\*\*\)s\) \+ \(\(sizetype\)ptrp\)\)' not supported} "" { target *-*-* } .-1 } */
+  /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\(\(int\*\)s\) \+ \(\(sizetype\)xp\)\)' not supported} "" { target *-*-* } .-2 } */
+  /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\*\(\(\(int\*\*\)s\) \+ \(\(sizetype\)ptrp\)\)\)' not supported} "" { target *-*-* } .-3 } */
+#pragma omp teams distribute parallel for
+  for (int i = 0; i < 64; i++)
+    {
+      (s->*xp)++;
+      (s->*ptrp)[i]++;
+    }
+
+  assert (s->*xp == 68);
+  for (int i = 0; i < 64; i++)
+    assert ((s->*ptrp)[i] == i + 1);
+
+  delete s->ptr;
+  delete s;
+
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/map-static-cast-lvalue-1.C b/gcc/testsuite/g++.dg/gomp/map-static-cast-lvalue-1.C
new file mode 100644
index 00000000000..3af9668202c
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/map-static-cast-lvalue-1.C
@@ -0,0 +1,17 @@ 
+#include <cassert>
+
+int foo (int x)
+{
+#pragma omp target map(static_cast<int&>(x))
+  /* { dg-message {sorry, unimplemented: unsupported map expression '& x'} "" { target *-*-* } .-1 } */
+  {
+    x += 3;
+  }
+  return x;
+}
+
+int main (int argc, char *argv[])
+{
+  assert (foo (5) == 8);
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/map-ternary-1.C b/gcc/testsuite/g++.dg/gomp/map-ternary-1.C
new file mode 100644
index 00000000000..7b365a909bb
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/map-ternary-1.C
@@ -0,0 +1,20 @@ 
+#include <cassert>
+
+int foo (bool yesno)
+{
+  int x = 5, y = 7;
+#pragma omp target map(yesno ? x : y)
+  /* { dg-message {sorry, unimplemented: unsupported map expression '\(yesno \?  x :  y\)'} "" { target *-*-* } .-1 } */
+  {
+    x += 3;
+    y += 5;
+  }
+  return yesno ? x : y;
+}
+
+int main (int argc, char *argv[])
+{
+  assert (foo (true) == 8);
+  assert (foo (false) == 12);
+  return 0;
+}
diff --git a/gcc/testsuite/g++.dg/gomp/member-array-2.C b/gcc/testsuite/g++.dg/gomp/member-array-2.C
new file mode 100644
index 00000000000..caf8ece4262
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/member-array-2.C
@@ -0,0 +1,91 @@ 
+#include <cassert>
+
+typedef int intarr100[100];
+
+class C {
+  int arr[100];
+  int *ptr;
+
+public:
+  C();
+  ~C();
+  void zero ();
+  void do_operation ();
+  void check (int, int);
+  intarr100 &get_arr () { return arr; }
+  int *get_ptr() { return ptr; }
+};
+
+C::C()
+{
+  ptr = new int[100];
+  for (int i = 0; i < 100; i++)
+    arr[i] = 0;
+}
+
+C::~C()
+{
+  delete ptr;
+}
+
+void
+C::zero ()
+{
+  for (int i = 0; i < 100; i++)
+    arr[i] = ptr[i] = 0;
+}
+
+void
+C::do_operation ()
+{
+#pragma omp target map(arr, ptr, ptr[:100])
+#pragma omp teams distribute parallel for
+  for (int i = 0; i < 100; i++)
+    {
+      arr[i] = arr[i] + 3;
+      ptr[i] = ptr[i] + 5;
+    }
+}
+
+void
+C::check (int arrval, int ptrval)
+{
+  for (int i = 0; i < 100; i++)
+    {
+      assert (arr[i] == arrval);
+      assert (ptr[i] == ptrval);
+    }
+}
+
+int
+main (int argc, char *argv[])
+{
+  C c;
+
+  c.zero ();
+  c.do_operation ();
+  c.check (3, 5);
+
+  /* It might sort of make sense to be able to do this, but we don't support
+     it for now.  */
+  #pragma omp target map(c.get_arr()[:100])
+  /* { dg-message {sorry, unimplemented: unsupported map expression 'c\.C::get_arr\(\)\[0\]'} "" { target *-*-* } .-1 } */
+  #pragma omp teams distribute parallel for
+    for (int i = 0; i < 100; i++)
+      c.get_arr()[i] += 2;
+
+  c.check (5, 5);
+
+  /* Same for this.  */
+  #pragma omp target map(c.get_ptr(), c.get_ptr()[:100])
+  /* { dg-message {sorry, unimplemented: unsupported map expression 'c\.C::get_ptr\(\)'} "" { target *-*-* } .-1 } */
+  /* { dg-message {sorry, unimplemented: unsupported map expression '\* c\.C::get_ptr\(\)'} "" { target *-*-* } .-2 } */
+  #pragma omp teams distribute parallel for
+    for (int i = 0; i < 100; i++)
+      c.get_ptr()[i] += 3;
+
+  c.check (5, 8);
+
+  return 0;
+}
+
diff --git a/gcc/testsuite/g++.dg/gomp/pr67522.C b/gcc/testsuite/g++.dg/gomp/pr67522.C
index da8cb74d1fa..4a901ba68c7 100644
--- a/gcc/testsuite/g++.dg/gomp/pr67522.C
+++ b/gcc/testsuite/g++.dg/gomp/pr67522.C
@@ -12,7 +12,7 @@  foo (void)
   for (int i = 0; i < 16; i++)
     ;
 
-  #pragma omp target map (S[0:10])		// { dg-error "is not a variable in" }
+  #pragma omp target map (S[0:10])		// { dg-error "expected primary-expression before '\\\[' token" }
   ;
 
   #pragma omp task depend (inout: S[0:10])	// { dg-error "is not a variable in" }
diff --git a/gcc/tree-pretty-print.cc b/gcc/tree-pretty-print.cc
index 5f19a3ed41b..197b01892a7 100644
--- a/gcc/tree-pretty-print.cc
+++ b/gcc/tree-pretty-print.cc
@@ -2710,6 +2710,20 @@  dump_generic_node (pretty_printer *pp, tree node, int spc, dump_flags_t flags,
 	}
       break;
 
+    case OMP_ARRAY_SECTION:
+      op0 = TREE_OPERAND (node, 0);
+      if (op_prio (op0) < op_prio (node))
+	pp_left_paren (pp);
+      dump_generic_node (pp, op0, spc, flags, false);
+      if (op_prio (op0) < op_prio (node))
+	pp_right_paren (pp);
+      pp_left_bracket (pp);
+      dump_generic_node (pp, TREE_OPERAND (node, 1), spc, flags, false);
+      pp_colon (pp);
+      dump_generic_node (pp, TREE_OPERAND (node, 2), spc, flags, false);
+      pp_right_bracket (pp);
+      break;
+
     case CONSTRUCTOR:
       {
 	unsigned HOST_WIDE_INT ix;
diff --git a/gcc/tree.def b/gcc/tree.def
index f5f9ae31b7c..32d9e651f0b 100644
--- a/gcc/tree.def
+++ b/gcc/tree.def
@@ -1364,6 +1364,9 @@  DEFTREECODE (OMP_ALLOCATE, "omp allocate", tcc_statement, 1)
 /* OpenMP clauses.  */
 DEFTREECODE (OMP_CLAUSE, "omp_clause", tcc_exceptional, 0)
 
+/* An OpenMP array section.  */
+DEFTREECODE (OMP_ARRAY_SECTION, "omp_array_section", tcc_expression, 3)
+
 /* TRANSACTION_EXPR tree code.
    Operand 0: BODY: contains body of the transaction.  */
 DEFTREECODE (TRANSACTION_EXPR, "transaction_expr", tcc_expression, 1)
diff --git a/libgomp/testsuite/libgomp.c++/baseptrs-4.C b/libgomp/testsuite/libgomp.c++/baseptrs-4.C
index 196029ac186..d5ca79c3344 100644
--- a/libgomp/testsuite/libgomp.c++/baseptrs-4.C
+++ b/libgomp/testsuite/libgomp.c++/baseptrs-4.C
@@ -11,11 +11,9 @@ 
 #define REF2PTR_DECL_BASE
 
 #define ARRAY_DECL_BASE
-// Needs map clause "lvalue"-parsing support.
-//#define REF2ARRAY_DECL_BASE
+#define REF2ARRAY_DECL_BASE
 #define PTR_OFFSET_DECL_BASE
-// Needs map clause "lvalue"-parsing support.
-//#define REF2PTR_OFFSET_DECL_BASE
+#define REF2PTR_OFFSET_DECL_BASE
 
 #define MAP_SECTIONS
 
@@ -30,25 +28,21 @@ 
 
 #define ARRAY_DECL_MEMBER_SLICE
 #define ARRAY_DECL_MEMBER_SLICE_BASEPTR
-// Needs map clause "lvalue"-parsing support.
-//#define REF2ARRAY_DECL_MEMBER_SLICE
-//#define REF2ARRAY_DECL_MEMBER_SLICE_BASEPTR
+#define REF2ARRAY_DECL_MEMBER_SLICE
+#define REF2ARRAY_DECL_MEMBER_SLICE_BASEPTR
 #define PTR_OFFSET_DECL_MEMBER_SLICE
 #define PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
-// Needs map clause "lvalue"-parsing support.
-//#define REF2PTR_OFFSET_DECL_MEMBER_SLICE
-//#define REF2PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+#define REF2PTR_OFFSET_DECL_MEMBER_SLICE
+#define REF2PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
 
 #define PTRARRAY_DECL_MEMBER_SLICE
 #define PTRARRAY_DECL_MEMBER_SLICE_BASEPTR
-// Needs map clause "lvalue"-parsing support.
-//#define REF2PTRARRAY_DECL_MEMBER_SLICE
-//#define REF2PTRARRAY_DECL_MEMBER_SLICE_BASEPTR
+#define REF2PTRARRAY_DECL_MEMBER_SLICE
+#define REF2PTRARRAY_DECL_MEMBER_SLICE_BASEPTR
 #define PTRPTR_OFFSET_DECL_MEMBER_SLICE
 #define PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
-// Needs map clause "lvalue"-parsing support.
-//#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE
-//#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE
+#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
 
 #define NONREF_COMPONENT_BASE
 #define NONREF_COMPONENT_MEMBER_SLICE
diff --git a/libgomp/testsuite/libgomp.c++/baseptrs-6.C b/libgomp/testsuite/libgomp.c++/baseptrs-6.C
new file mode 100644
index 00000000000..0fc93d28664
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/baseptrs-6.C
@@ -0,0 +1,3199 @@ 
+// { dg-do run }
+
+// This is essentially baseptrs-4.C with templates.
+
+#include <cstring>
+#include <cassert>
+
+#define MAP_DECLS
+
+#define NONREF_DECL_BASE
+#define REF_DECL_BASE
+#define PTR_DECL_BASE
+#define REF2PTR_DECL_BASE
+
+#define ARRAY_DECL_BASE
+#define REF2ARRAY_DECL_BASE
+#define PTR_OFFSET_DECL_BASE
+#define REF2PTR_OFFSET_DECL_BASE
+
+#define MAP_SECTIONS
+
+#define NONREF_DECL_MEMBER_SLICE
+#define NONREF_DECL_MEMBER_SLICE_BASEPTR
+#define REF_DECL_MEMBER_SLICE
+#define REF_DECL_MEMBER_SLICE_BASEPTR
+#define PTR_DECL_MEMBER_SLICE
+#define PTR_DECL_MEMBER_SLICE_BASEPTR
+#define REF2PTR_DECL_MEMBER_SLICE
+#define REF2PTR_DECL_MEMBER_SLICE_BASEPTR
+
+#define ARRAY_DECL_MEMBER_SLICE
+#define ARRAY_DECL_MEMBER_SLICE_BASEPTR
+#define REF2ARRAY_DECL_MEMBER_SLICE
+#define REF2ARRAY_DECL_MEMBER_SLICE_BASEPTR
+#define PTR_OFFSET_DECL_MEMBER_SLICE
+#define PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+#define REF2PTR_OFFSET_DECL_MEMBER_SLICE
+#define REF2PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+
+#define PTRARRAY_DECL_MEMBER_SLICE
+#define PTRARRAY_DECL_MEMBER_SLICE_BASEPTR
+#define REF2PTRARRAY_DECL_MEMBER_SLICE
+#define REF2PTRARRAY_DECL_MEMBER_SLICE_BASEPTR
+#define PTRPTR_OFFSET_DECL_MEMBER_SLICE
+#define PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE
+#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+
+#define NONREF_COMPONENT_BASE
+#define NONREF_COMPONENT_MEMBER_SLICE
+#define NONREF_COMPONENT_MEMBER_SLICE_BASEPTR
+
+#define REF_COMPONENT_BASE
+#define REF_COMPONENT_MEMBER_SLICE
+#define REF_COMPONENT_MEMBER_SLICE_BASEPTR
+
+#define PTR_COMPONENT_BASE
+#define PTR_COMPONENT_MEMBER_SLICE
+#define PTR_COMPONENT_MEMBER_SLICE_BASEPTR
+
+#define REF2PTR_COMPONENT_BASE
+#define REF2PTR_COMPONENT_MEMBER_SLICE
+#define REF2PTR_COMPONENT_MEMBER_SLICE_BASEPTR
+
+#ifdef MAP_DECLS
+template<int N>
+void
+map_decls (void)
+{
+  int x = 0;
+  int &y = x;
+  int arr[4];
+  int (&arrref)[N] = arr;
+  int *z = &arr[0];
+  int *&t = z;
+
+  memset (arr, 0, sizeof arr);
+
+  #pragma omp target map(x)
+  {
+    x++;
+  }
+
+  #pragma omp target map(y)
+  {
+    y++;
+  }
+
+  assert (x == 2);
+  assert (y == 2);
+
+  /* "A variable that is of type pointer is treated as if it is the base
+      pointer of a zero-length array section that appeared as a list item in a
+      map clause."  */
+  #pragma omp target map(z)
+  {
+    z++;
+  }
+
+  /* "A variable that is of type reference to pointer is treated as if it had
+      appeared in a map clause as a zero-length array section."
+
+     The pointer here is *not* associated with a target address, so we're not
+     disallowed from modifying it.  */
+  #pragma omp target map(t)
+  {
+    t++;
+  }
+
+  assert (z == &arr[2]);
+  assert (t == &arr[2]);
+
+  #pragma omp target map(arr)
+  {
+    arr[2]++;
+  }
+
+  #pragma omp target map(arrref)
+  {
+    arrref[2]++;
+  }
+
+  assert (arr[2] == 2);
+  assert (arrref[2] == 2);
+}
+#endif
+
+template<typename T>
+struct S {
+  T a;
+  T &b;
+  T *c;
+  T *&d;
+  T e[4];
+  T (&f)[4];
+
+  S(T a1, T &b1, T *c1, T *&d1) :
+    a(a1), b(b1), c(c1), d(d1), f(e)
+  {
+    memset (e, 0, sizeof e);
+  }
+};
+
+#ifdef NONREF_DECL_BASE
+template<typename T>
+void
+nonref_decl_base (void)
+{
+  T a = 0, b = 0, c, *d = &c;
+  S<T> mys(a, b, &c, d);
+
+  #pragma omp target map(mys.a)
+  {
+    mys.a++;
+  }
+
+  #pragma omp target map(mys.b)
+  {
+    mys.b++;
+  }
+
+  assert (mys.a == 1);
+  assert (mys.b == 1);
+
+  #pragma omp target map(mys.c)
+  {
+    mys.c++;
+  }
+
+  #pragma omp target map(mys.d)
+  {
+    mys.d++;
+  }
+
+  assert (mys.c == &c + 1);
+  assert (mys.d == &c + 1);
+
+  #pragma omp target map(mys.e)
+  {
+    mys.e[0]++;
+  }
+
+  #pragma omp target map(mys.f)
+  {
+    mys.f[0]++;
+  }
+
+  assert (mys.e[0] == 2);
+  assert (mys.f[0] == 2);
+}
+#endif
+
+#ifdef REF_DECL_BASE
+template<typename T>
+void
+ref_decl_base (void)
+{
+  T a = 0, b = 0, c, *d = &c;
+  S<T> mys_orig(a, b, &c, d);
+  S<T> &mys = mys_orig;
+
+  #pragma omp target map(mys.a)
+  {
+    mys.a++;
+  }
+
+  #pragma omp target map(mys.b)
+  {
+    mys.b++;
+  }
+
+  assert (mys.a == 1);
+  assert (mys.b == 1);
+
+  #pragma omp target map(mys.c)
+  {
+    mys.c++;
+  }
+
+  #pragma omp target map(mys.d)
+  {
+    mys.d++;
+  }
+
+  assert (mys.c == &c + 1);
+  assert (mys.d == &c + 1);
+
+  #pragma omp target map(mys.e)
+  {
+    mys.e[0]++;
+  }
+
+  #pragma omp target map(mys.f)
+  {
+    mys.f[0]++;
+  }
+
+  assert (mys.e[0] == 2);
+  assert (mys.f[0] == 2);
+}
+#endif
+
+#ifdef PTR_DECL_BASE
+template<typename A>
+void
+ptr_decl_base (void)
+{
+  A a = 0, b = 0, c, *d = &c;
+  S<A> mys_orig(a, b, &c, d);
+  S<A> *mys = &mys_orig;
+
+  #pragma omp target map(mys->a)
+  {
+    mys->a++;
+  }
+
+  #pragma omp target map(mys->b)
+  {
+    mys->b++;
+  }
+
+  assert (mys->a == 1);
+  assert (mys->b == 1);
+
+  #pragma omp target map(mys->c)
+  {
+    mys->c++;
+  }
+
+  #pragma omp target map(mys->d)
+  {
+    mys->d++;
+  }
+
+  assert (mys->c == &c + 1);
+  assert (mys->d == &c + 1);
+
+  #pragma omp target map(mys->e)
+  {
+    mys->e[0]++;
+  }
+
+  #pragma omp target map(mys->f)
+  {
+    mys->f[0]++;
+  }
+
+  assert (mys->e[0] == 2);
+  assert (mys->f[0] == 2);
+}
+#endif
+
+#ifdef REF2PTR_DECL_BASE
+template<typename A>
+void
+ref2ptr_decl_base (void)
+{
+  A a = 0, b = 0, c, *d = &c;
+  S<A> mys_orig(a, b, &c, d);
+  S<A> *mysp = &mys_orig;
+  S<A> *&mys = mysp;
+
+  #pragma omp target map(mys->a)
+  {
+    mys->a++;
+  }
+
+  #pragma omp target map(mys->b)
+  {
+    mys->b++;
+  }
+
+  assert (mys->a == 1);
+  assert (mys->b == 1);
+
+  #pragma omp target map(mys->c)
+  {
+    mys->c++;
+  }
+
+  #pragma omp target map(mys->d)
+  {
+    mys->d++;
+  }
+
+  assert (mys->c == &c + 1);
+  assert (mys->d == &c + 1);
+
+  #pragma omp target map(mys->e)
+  {
+    mys->e[0]++;
+  }
+
+  #pragma omp target map(mys->f)
+  {
+    mys->f[0]++;
+  }
+
+  assert (mys->e[0] == 2);
+  assert (mys->f[0] == 2);
+}
+#endif
+
+#ifdef ARRAY_DECL_BASE
+template<typename A>
+void
+array_decl_base (void)
+{
+  A a = 0, b = 0, c, *d = &c;
+  S<A> mys[4] =
+    {
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d)
+    };
+
+  #pragma omp target map(mys[2].a)
+  {
+    mys[2].a++;
+  }
+
+  #pragma omp target map(mys[2].b)
+  {
+    mys[2].b++;
+  }
+
+  assert (mys[2].a == 1);
+  assert (mys[2].b == 1);
+
+  #pragma omp target map(mys[2].c)
+  {
+    mys[2].c++;
+  }
+
+  #pragma omp target map(mys[2].d)
+  {
+    mys[2].d++;
+  }
+
+  assert (mys[2].c == &c + 1);
+  assert (mys[2].d == &c + 1);
+
+  #pragma omp target map(mys[2].e)
+  {
+    mys[2].e[0]++;
+  }
+
+  #pragma omp target map(mys[2].f)
+  {
+    mys[2].f[0]++;
+  }
+
+  assert (mys[2].e[0] == 2);
+  assert (mys[2].f[0] == 2);
+}
+#endif
+
+#ifdef REF2ARRAY_DECL_BASE
+template<typename A>
+void
+ref2array_decl_base (void)
+{
+  A a = 0, b = 0, c, *d = &c;
+  S<A> mys_orig[4] =
+    {
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d)
+    };
+  S<A> (&mys)[4] = mys_orig;
+
+  #pragma omp target map(mys[2].a)
+  {
+    mys[2].a++;
+  }
+
+  #pragma omp target map(mys[2].b)
+  {
+    mys[2].b++;
+  }
+
+  assert (mys[2].a == 1);
+  assert (mys[2].b == 1);
+
+  #pragma omp target map(mys[2].c)
+  {
+    mys[2].c++;
+  }
+
+  #pragma omp target map(mys[2].d)
+  {
+    mys[2].d++;
+  }
+
+  assert (mys[2].c == &c + 1);
+  assert (mys[2].d == &c + 1);
+
+  #pragma omp target map(mys[2].e)
+  {
+    mys[2].e[0]++;
+  }
+
+  #pragma omp target map(mys[2].f)
+  {
+    mys[2].f[0]++;
+  }
+
+  assert (mys[2].e[0] == 2);
+  assert (mys[2].f[0] == 2);
+}
+#endif
+
+#ifdef PTR_OFFSET_DECL_BASE
+template<typename A>
+void
+ptr_offset_decl_base (void)
+{
+  A a = 0, b = 0, c, *d = &c;
+  S<A> mys_orig[4] =
+    {
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d)
+    };
+  S<A> *mys = &mys_orig[0];
+
+  #pragma omp target map(mys[2].a)
+  {
+    mys[2].a++;
+  }
+
+  #pragma omp target map(mys[2].b)
+  {
+    mys[2].b++;
+  }
+
+  assert (mys[2].a == 1);
+  assert (mys[2].b == 1);
+
+  #pragma omp target map(mys[2].c)
+  {
+    mys[2].c++;
+  }
+
+  #pragma omp target map(mys[2].d)
+  {
+    mys[2].d++;
+  }
+
+  assert (mys[2].c == &c + 1);
+  assert (mys[2].d == &c + 1);
+
+  #pragma omp target map(mys[2].e)
+  {
+    mys[2].e[0]++;
+  }
+
+  #pragma omp target map(mys[2].f)
+  {
+    mys[2].f[0]++;
+  }
+
+  assert (mys[2].e[0] == 2);
+  assert (mys[2].f[0] == 2);
+}
+#endif
+
+#ifdef REF2PTR_OFFSET_DECL_BASE
+template<typename A>
+void
+ref2ptr_offset_decl_base (void)
+{
+  A a = 0, b = 0, c, *d = &c;
+  S<A> mys_orig[4] =
+    {
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d),
+      S<A>(a, b, &c, d)
+    };
+  S<A> *mys_ptr = &mys_orig[0];
+  S<A> *&mys = mys_ptr;
+
+  #pragma omp target map(mys[2].a)
+  {
+    mys[2].a++;
+  }
+
+  #pragma omp target map(mys[2].b)
+  {
+    mys[2].b++;
+  }
+
+  assert (mys[2].a == 1);
+  assert (mys[2].b == 1);
+
+  #pragma omp target map(mys[2].c)
+  {
+    mys[2].c++;
+  }
+
+  #pragma omp target map(mys[2].d)
+  {
+    mys[2].d++;
+  }
+
+  assert (mys[2].c == &c + 1);
+  assert (mys[2].d == &c + 1);
+
+  #pragma omp target map(mys[2].e)
+  {
+    mys[2].e[0]++;
+  }
+
+  #pragma omp target map(mys[2].f)
+  {
+    mys[2].f[0]++;
+  }
+
+  assert (mys[2].e[0] == 2);
+  assert (mys[2].f[0] == 2);
+}
+#endif
+
+#ifdef MAP_SECTIONS
+template<typename A, int B>
+void
+map_sections (void)
+{
+  A arr[B];
+  A *ptr;
+  A (&arrref)[B] = arr;
+  A *&ptrref = ptr;
+
+  ptr = new int[B];
+  memset (ptr, 0, sizeof (int) * B);
+  memset (arr, 0, sizeof (int) * B);
+
+  #pragma omp target map(arr[0:B])
+  {
+    arr[2]++;
+  }
+
+  #pragma omp target map(ptr[0:B])
+  {
+    ptr[2]++;
+  }
+
+  #pragma omp target map(arrref[0:B])
+  {
+    arrref[2]++;
+  }
+
+  #pragma omp target map(ptrref[0:B])
+  {
+    ptrref[2]++;
+  }
+
+  assert (arr[2] == 2);
+  assert (ptr[2] == 2);
+
+  delete ptr;
+}
+#endif
+
+template<typename A>
+struct T {
+  A a[10];
+  A (&b)[10];
+  A *c;
+  A *&d;
+
+  T(A (&b1)[10], A *c1, A *&d1) : b(b1), c(c1), d(d1)
+  {
+    memset (a, 0, sizeof a);
+  }
+};
+
+#ifdef NONREF_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+nonref_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt(c, &c[0], d);
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(myt.a[0:B])
+  {
+    myt.a[2]++;
+  }
+
+  #pragma omp target map(myt.b[0:B])
+  {
+    myt.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt.c)
+
+  #pragma omp target map(myt.c[0:B])
+  {
+    myt.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt.c)
+
+  #pragma omp target enter data map(to: myt.d)
+
+  #pragma omp target map(myt.d[0:B])
+  {
+    myt.d[2]++;
+  }
+
+  #pragma omp target exit data map(from: myt.d)
+
+  assert (myt.a[2] == 1);
+  assert (myt.b[2] == 3);
+  assert (myt.c[2] == 3);
+  assert (myt.d[2] == 3);
+}
+#endif
+
+#ifdef NONREF_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+nonref_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt(c, &c[0], d);
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(to:myt.c) map(myt.c[0:B])
+  {
+    myt.c[2]++;
+  }
+
+  #pragma omp target map(to:myt.d) map(myt.d[0:B])
+  {
+    myt.d[2]++;
+  }
+
+  assert (myt.c[2] == 2);
+  assert (myt.d[2] == 2);
+}
+#endif
+
+#ifdef REF_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ref_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> &myt = myt_real;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(myt.a[0:B])
+  {
+    myt.a[2]++;
+  }
+
+  #pragma omp target map(myt.b[0:B])
+  {
+    myt.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt.c)
+
+  #pragma omp target map(myt.c[0:B])
+  {
+    myt.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt.c)
+
+  #pragma omp target enter data map(to: myt.d)
+
+  #pragma omp target map(myt.d[0:B])
+  {
+    myt.d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt.d)
+
+  assert (myt.a[2] == 1);
+  assert (myt.b[2] == 3);
+  assert (myt.c[2] == 3);
+  assert (myt.d[2] == 3);
+}
+#endif
+
+#ifdef REF_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ref_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> &myt = myt_real;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(to:myt.c) map(myt.c[0:B])
+  {
+    myt.c[2]++;
+  }
+
+  #pragma omp target map(to:myt.d) map(myt.d[0:B])
+  {
+    myt.d[2]++;
+  }
+
+  assert (myt.c[2] == 2);
+  assert (myt.d[2] == 2);
+}
+#endif
+
+#ifdef PTR_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ptr_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt = &myt_real;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target enter data map(to: myt)
+
+  #pragma omp target map(myt->a[0:B])
+  {
+    myt->a[2]++;
+  }
+
+  #pragma omp target map(myt->b[0:B])
+  {
+    myt->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt->c)
+
+  #pragma omp target map(myt->c[0:B])
+  {
+    myt->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt->c)
+
+  #pragma omp target enter data map(to: myt->d)
+
+  #pragma omp target map(myt->d[0:B])
+  {
+    myt->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt, myt->d)
+
+  assert (myt->a[2] == 1);
+  assert (myt->b[2] == 3);
+  assert (myt->c[2] == 3);
+  assert (myt->d[2] == 3);
+}
+#endif
+
+#ifdef PTR_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ptr_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt = &myt_real;
+
+  memset (c, 0, sizeof c);
+
+  // These ones have an implicit firstprivate for 'myt'.
+  #pragma omp target map(to:myt->c) map(myt->c[0:B])
+  {
+    myt->c[2]++;
+  }
+
+  #pragma omp target map(to:myt->d) map(myt->d[0:B])
+  {
+    myt->d[2]++;
+  }
+
+  // These ones have an explicit "TO" mapping for 'myt'.
+  #pragma omp target map(to:myt) map(to:myt->c) map(myt->c[0:B])
+  {
+    myt->c[2]++;
+  }
+
+  #pragma omp target map(to:myt) map(to:myt->d) map(myt->d[0:B])
+  {
+    myt->d[2]++;
+  }
+
+  assert (myt->c[2] == 4);
+  assert (myt->d[2] == 4);
+}
+#endif
+
+#ifdef REF2PTR_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ref2ptr_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt_ptr = &myt_real;
+  T<A> *&myt = myt_ptr;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target enter data map(to: myt)
+
+  #pragma omp target map(myt->a[0:B])
+  {
+    myt->a[2]++;
+  }
+
+  #pragma omp target map(myt->b[0:B])
+  {
+    myt->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt->c)
+
+  #pragma omp target map(myt->c[0:B])
+  {
+    myt->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt->c)
+
+  #pragma omp target enter data map(to: myt->d)
+
+  #pragma omp target map(myt->d[0:B])
+  {
+    myt->d[2]++;
+  }
+
+  #pragma omp target exit data map(from: myt, myt->d)
+
+  assert (myt->a[2] == 1);
+  assert (myt->b[2] == 3);
+  assert (myt->c[2] == 3);
+  assert (myt->d[2] == 3);
+}
+#endif
+
+#ifdef REF2PTR_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ref2ptr_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt_ptr = &myt_real;
+  T<A> *&myt = myt_ptr;
+
+  memset (c, 0, sizeof c);
+
+  // These ones have an implicit firstprivate for 'myt'.
+  #pragma omp target map(to:myt->c) map(myt->c[0:B])
+  {
+    myt->c[2]++;
+  }
+
+  #pragma omp target map(to:myt->d) map(myt->d[0:B])
+  {
+    myt->d[2]++;
+  }
+
+  // These ones have an explicit "TO" mapping for 'myt'.
+  #pragma omp target map(to:myt) map(to:myt->c) map(myt->c[0:B])
+  {
+    myt->c[2]++;
+  }
+
+  #pragma omp target map(to:myt) map(to:myt->d) map(myt->d[0:B])
+  {
+    myt->d[2]++;
+  }
+
+  assert (myt->c[2] == 4);
+  assert (myt->d[2] == 4);
+}
+#endif
+
+#ifdef ARRAY_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+array_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt[4] =
+    {
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d)
+    };
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(myt[2].a[0:B])
+  {
+    myt[2].a[2]++;
+  }
+
+  #pragma omp target map(myt[2].b[0:B])
+  {
+    myt[2].b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt[2].c)
+
+  #pragma omp target map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2].c)
+
+  #pragma omp target enter data map(to: myt[2].d)
+
+  #pragma omp target map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2].d)
+
+  assert (myt[2].a[2] == 1);
+  assert (myt[2].b[2] == 3);
+  assert (myt[2].c[2] == 3);
+  assert (myt[2].d[2] == 3);
+}
+#endif
+
+#ifdef ARRAY_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+array_decl_member_slice_baseptr (void)
+{
+  A c[10];
+  A *d = &c[0];
+  T<A> myt[4] =
+    {
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d)
+    };
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(to:myt[2].c) map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target map(to:myt[2].d) map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  assert (myt[2].c[2] == 2);
+  assert (myt[2].d[2] == 2);
+}
+#endif
+
+#ifdef REF2ARRAY_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ref2array_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real[4] =
+    {
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d)
+    };
+  T<A> (&myt)[4] = myt_real;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(myt[2].a[0:B])
+  {
+    myt[2].a[2]++;
+  }
+
+  #pragma omp target map(myt[2].b[0:B])
+  {
+    myt[2].b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt[2].c)
+
+  #pragma omp target map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2].c)
+
+  #pragma omp target enter data map(to: myt[2].d)
+
+  #pragma omp target map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2].d)
+
+  assert (myt[2].a[2] == 1);
+  assert (myt[2].b[2] == 3);
+  assert (myt[2].c[2] == 3);
+  assert (myt[2].d[2] == 3);
+}
+#endif
+
+#ifdef REF2ARRAY_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ref2array_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real[4] =
+    {
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d)
+    };
+  T<A> (&myt)[4] = myt_real;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(to:myt[2].c) map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target map(to:myt[2].d) map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  assert (myt[2].c[2] == 2);
+  assert (myt[2].d[2] == 2);
+}
+#endif
+
+#ifdef PTR_OFFSET_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ptr_offset_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real[4] =
+    {
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d)
+    };
+  T<A> *myt = &myt_real[0];
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(myt[2].a[0:B])
+  {
+    myt[2].a[2]++;
+  }
+
+  #pragma omp target map(myt[2].b[0:B])
+  {
+    myt[2].b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt[2].c)
+
+  #pragma omp target map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2].c)
+
+  #pragma omp target enter data map(to: myt[2].d)
+
+  #pragma omp target map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2].d)
+
+  assert (myt[2].a[2] == 1);
+  assert (myt[2].b[2] == 3);
+  assert (myt[2].c[2] == 3);
+  assert (myt[2].d[2] == 3);
+}
+#endif
+
+#ifdef PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ptr_offset_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real[4] =
+    {
+      T (c, &c[0], d),
+      T (c, &c[0], d),
+      T (c, &c[0], d),
+      T (c, &c[0], d)
+    };
+  T<A> *myt = &myt_real[0];
+
+  memset (c, 0, sizeof c);
+
+  /* Implicit 'myt'.  */
+  #pragma omp target map(to:myt[2].c) map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target map(to:myt[2].d) map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  /* Explicit 'to'-mapped 'myt'.  */
+  #pragma omp target map(to:myt) map(to:myt[2].c) map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target map(to:myt) map(to:myt[2].d) map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  assert (myt[2].c[2] == 4);
+  assert (myt[2].d[2] == 4);
+}
+#endif
+
+#ifdef REF2PTR_OFFSET_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ref2ptr_offset_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real[4] =
+    {
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d)
+    };
+  T<A> *myt_ptr = &myt_real[0];
+  T<A> *&myt = myt_ptr;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(myt[2].a[0:B])
+  {
+    myt[2].a[2]++;
+  }
+
+  #pragma omp target map(myt[2].b[0:B])
+  {
+    myt[2].b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt[2].c)
+
+  #pragma omp target map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2].c)
+
+  #pragma omp target enter data map(to: myt[2].d)
+
+  #pragma omp target map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2].d)
+
+  assert (myt[2].a[2] == 1);
+  assert (myt[2].b[2] == 3);
+  assert (myt[2].c[2] == 3);
+  assert (myt[2].d[2] == 3);
+}
+#endif
+
+#ifdef REF2PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ref2ptr_offset_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real[4] =
+    {
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d),
+      T<A> (c, &c[0], d)
+    };
+  T<A> *myt_ptr = &myt_real[0];
+  T<A> *&myt = myt_ptr;
+
+  memset (c, 0, sizeof c);
+
+  /* Implicit 'myt'.  */
+  #pragma omp target map(to:myt[2].c) map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target map(to:myt[2].d) map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  /* Explicit 'to'-mapped 'myt'.  */
+  #pragma omp target map(to:myt) map(to:myt[2].c) map(myt[2].c[0:B])
+  {
+    myt[2].c[2]++;
+  }
+
+  #pragma omp target map(to:myt) map(to:myt[2].d) map(myt[2].d[0:B])
+  {
+    myt[2].d[2]++;
+  }
+
+  assert (myt[2].c[2] == 4);
+  assert (myt[2].d[2] == 4);
+}
+#endif
+
+#ifdef PTRARRAY_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ptrarray_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt[4] =
+    {
+      &myt_real,
+      &myt_real,
+      &myt_real,
+      &myt_real
+    };
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target enter data map(to: myt[2])
+
+  #pragma omp target map(myt[2]->a[0:B])
+  {
+    myt[2]->a[2]++;
+  }
+
+  #pragma omp target map(myt[2]->b[0:B])
+  {
+    myt[2]->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt[2]->c)
+
+  #pragma omp target map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target exit data map(from: myt[2]->c)
+
+  #pragma omp target enter data map(to: myt[2]->d)
+
+  #pragma omp target map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  #pragma omp target exit data map(from: myt[2]->d)
+
+  #pragma omp target exit data map(release: myt[2])
+
+  assert (myt[2]->a[2] == 1);
+  assert (myt[2]->b[2] == 3);
+  assert (myt[2]->c[2] == 3);
+  assert (myt[2]->d[2] == 3);
+}
+#endif
+
+#ifdef PTRARRAY_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ptrarray_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt[4] =
+    {
+      &myt_real,
+      &myt_real,
+      &myt_real,
+      &myt_real
+    };
+
+  memset (c, 0, sizeof c);
+
+  // Implicit 'myt'
+  #pragma omp target map(to: myt[2]->c) map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target map(to: myt[2]->d) map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  // One element of 'myt'
+  #pragma omp target map(to:myt[2], myt[2]->c) map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target map(to:myt[2], myt[2]->d) map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  // Explicit map of all of 'myt'
+  #pragma omp target map(to:myt, myt[2]->c) map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target map(to:myt, myt[2]->d) map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  // Explicit map slice of 'myt'
+  #pragma omp target map(to:myt[1:3], myt[2]->c) map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target map(to:myt[1:3], myt[2]->d) map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  assert (myt[2]->c[2] == 8);
+  assert (myt[2]->d[2] == 8);
+}
+#endif
+
+#ifdef REF2PTRARRAY_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ref2ptrarray_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt_ptrarr[4] =
+    {
+      &myt_real,
+      &myt_real,
+      &myt_real,
+      &myt_real
+    };
+  T<A> *(&myt)[4] = myt_ptrarr;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target enter data map(to: myt[2])
+
+  #pragma omp target map(myt[2]->a[0:B])
+  {
+    myt[2]->a[2]++;
+  }
+
+  #pragma omp target map(myt[2]->b[0:B])
+  {
+    myt[2]->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt[2]->c)
+
+  #pragma omp target map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2]->c)
+
+  #pragma omp target enter data map(to: myt[2]->d)
+
+  #pragma omp target map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[2]->d)
+
+  #pragma omp target exit data map(release: myt[2])
+
+  assert (myt[2]->a[2] == 1);
+  assert (myt[2]->b[2] == 3);
+  assert (myt[2]->c[2] == 3);
+  assert (myt[2]->d[2] == 3);
+}
+#endif
+
+#ifdef REF2PTRARRAY_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ref2ptrarray_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt_ptrarr[4] =
+    {
+      &myt_real,
+      &myt_real,
+      &myt_real,
+      &myt_real
+    };
+  T<A> *(&myt)[4] = myt_ptrarr;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(to:myt[2], myt[2]->c) map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target map(to:myt[2], myt[2]->d) map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  #pragma omp target map(to:myt, myt[2]->c) map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target map(to:myt, myt[2]->d) map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  assert (myt[2]->c[2] == 4);
+  assert (myt[2]->d[2] == 4);
+}
+#endif
+
+#ifdef PTRPTR_OFFSET_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ptrptr_offset_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt_ptrarr[4] =
+    {
+      &myt_real,
+      &myt_real,
+      &myt_real,
+      &myt_real
+    };
+  T<A> **myt = &myt_ptrarr[0];
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target enter data map(to: myt[0:3])
+
+  /* NOTE: For the implicit firstprivate 'myt' to work, the zeroth element of
+     myt[] must be mapped above -- otherwise the zero-length array section
+     lookup fails.  */
+  #pragma omp target map(myt[2]->a[0:B])
+  {
+    myt[2]->a[2]++;
+  }
+
+  #pragma omp target map(myt[2]->b[0:B])
+  {
+    myt[2]->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myt[2]->c)
+
+  #pragma omp target map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target exit data map(from: myt[2]->c)
+
+  #pragma omp target enter data map(to: myt[2]->d)
+
+  #pragma omp target map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  #pragma omp target exit data map(from: myt[0:3], myt[2]->d)
+
+  assert (myt[2]->a[2] == 1);
+  assert (myt[2]->b[2] == 3);
+  assert (myt[2]->c[2] == 3);
+  assert (myt[2]->d[2] == 3);
+}
+#endif
+
+#ifdef PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ptrptr_offset_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt_ptrarr[4] =
+    {
+      0,
+      0,
+      0,
+      &myt_real
+    };
+  T<A> **myt = &myt_ptrarr[0];
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(to:myt[3], myt[3]->c) map(myt[3]->c[0:B])
+  {
+    myt[3]->c[2]++;
+  }
+
+  #pragma omp target map(to:myt[3], myt[3]->d) map(myt[3]->d[0:B])
+  {
+    myt[3]->d[2]++;
+  }
+
+  #pragma omp target map(to:myt, myt[3], myt[3]->c) map(myt[3]->c[0:B])
+  {
+    myt[3]->c[2]++;
+  }
+
+  #pragma omp target map(to:myt, myt[3], myt[3]->d) map(myt[3]->d[0:B])
+  {
+    myt[3]->d[2]++;
+  }
+
+  assert (myt[3]->c[2] == 4);
+  assert (myt[3]->d[2] == 4);
+}
+#endif
+
+#ifdef REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE
+template<typename A, int B>
+void
+ref2ptrptr_offset_decl_member_slice (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt_ptrarr[4] =
+    {
+      0,
+      0,
+      &myt_real,
+      0
+    };
+  T<A> **myt_ptrptr = &myt_ptrarr[0];
+  T<A> **&myt = myt_ptrptr;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target enter data map(to: myt[0:3])
+
+  #pragma omp target map(myt[2]->a[0:B])
+  {
+    myt[2]->a[2]++;
+  }
+
+  #pragma omp target map(myt[2]->b[0:B])
+  {
+    myt[2]->b[2]++;
+  }
+
+  #pragma omp target enter data map(to:myt[2]->c)
+
+  #pragma omp target map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target exit data map(release:myt[2]->c)
+
+  #pragma omp target enter data map(to:myt[2]->d)
+
+  #pragma omp target map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myt[0:3], myt[2]->d)
+
+  assert (myt[2]->a[2] == 1);
+  assert (myt[2]->b[2] == 3);
+  assert (myt[2]->c[2] == 3);
+  assert (myt[2]->d[2] == 3);
+}
+#endif
+
+#ifdef REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+template<typename A, int B>
+void
+ref2ptrptr_offset_decl_member_slice_baseptr (void)
+{
+  A c[B];
+  A *d = &c[0];
+  T<A> myt_real(c, &c[0], d);
+  T<A> *myt_ptrarr[4] =
+    {
+      0,
+      0,
+      &myt_real,
+      0
+    };
+  T<A> **myt_ptrptr = &myt_ptrarr[0];
+  T<A> **&myt = myt_ptrptr;
+
+  memset (c, 0, sizeof c);
+
+  #pragma omp target map(to:myt[2], myt[2]->c) map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target map(to:myt[2], myt[2]->d) map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  #pragma omp target map(to:myt, myt[2], myt[2]->c) map(myt[2]->c[0:B])
+  {
+    myt[2]->c[2]++;
+  }
+
+  #pragma omp target map(to:myt, myt[2], myt[2]->d) map(myt[2]->d[0:B])
+  {
+    myt[2]->d[2]++;
+  }
+
+  assert (myt[2]->c[2] == 4);
+  assert (myt[2]->d[2] == 4);
+}
+#endif
+
+template<typename A>
+struct U
+{
+  S<A> s1;
+  T<A> t1;
+  S<A> &s2;
+  T<A> &t2;
+  S<A> *s3;
+  T<A> *t3;
+  S<A> *&s4;
+  T<A> *&t4;
+
+  U(S<A> &sptr1, T<A> &tptr1, S<A> &sptr2, T<A> &tptr2,
+    S<A> *sptr3, T<A> *tptr3, S<A> *&sptr4, T<A> *&tptr4)
+    : s1(sptr1), t1(tptr1), s2(sptr2), t2(tptr2), s3(sptr3), t3(tptr3),
+      s4(sptr4), t4(tptr4)
+  {
+  }
+};
+
+#define INIT_S(N)				\
+  A a##N = 0, b##N = 0, c##N = 0, d##N = 0;	\
+  A *d##N##ptr = &d##N;				\
+  S<A> s##N(a##N, b##N, &c##N, d##N##ptr)
+
+#define INIT_T(N)				\
+  A arr##N[10];					\
+  A *ptr##N = &arr##N[0];			\
+  T<A> t##N(arr##N, &arr##N[0], ptr##N);	\
+  memset (arr##N, 0, sizeof arr##N)
+
+#define INIT_ST					\
+  INIT_S(1);					\
+  INIT_T(1);					\
+  INIT_S(2);					\
+  INIT_T(2);					\
+  INIT_S(3);					\
+  INIT_T(3);					\
+  A a4 = 0, b4 = 0, c4 = 0, d4 = 0;		\
+  A *d4ptr = &d4;				\
+  S<A> *s4 = new S<A>(a4, b4, &c4, d4ptr);	\
+  A arr4[10];					\
+  A *ptr4 = &arr4[0];				\
+  T<A> *t4 = new T<A>(arr4, &arr4[0], ptr4);	\
+  memset (arr4, 0, sizeof arr4)
+
+#ifdef NONREF_COMPONENT_BASE
+template<typename A>
+void
+nonref_component_base (void)
+{
+  INIT_ST;
+  U<A> myu(s1, t1, s2, t2, &s3, &t3, s4, t4);
+
+  #pragma omp target map(myu.s1.a, myu.s1.b, myu.s1.c, myu.s1.d)
+  {
+    myu.s1.a++;
+    myu.s1.b++;
+    myu.s1.c++;
+    myu.s1.d++;
+  }
+
+  assert (myu.s1.a == 1);
+  assert (myu.s1.b == 1);
+  assert (myu.s1.c == &c1 + 1);
+  assert (myu.s1.d == &d1 + 1);
+
+  #pragma omp target map(myu.s2.a, myu.s2.b, myu.s2.c, myu.s2.d)
+  {
+    myu.s2.a++;
+    myu.s2.b++;
+    myu.s2.c++;
+    myu.s2.d++;
+  }
+
+  assert (myu.s2.a == 1);
+  assert (myu.s2.b == 1);
+  assert (myu.s2.c == &c2 + 1);
+  assert (myu.s2.d == &d2 + 1);
+
+  #pragma omp target map(to:myu.s3) \
+		     map(myu.s3->a, myu.s3->b, myu.s3->c, myu.s3->d)
+  {
+    myu.s3->a++;
+    myu.s3->b++;
+    myu.s3->c++;
+    myu.s3->d++;
+  }
+
+  assert (myu.s3->a == 1);
+  assert (myu.s3->b == 1);
+  assert (myu.s3->c == &c3 + 1);
+  assert (myu.s3->d == &d3 + 1);
+
+  #pragma omp target map(to:myu.s4) \
+		     map(myu.s4->a, myu.s4->b, myu.s4->c, myu.s4->d)
+  {
+    myu.s4->a++;
+    myu.s4->b++;
+    myu.s4->c++;
+    myu.s4->d++;
+  }
+
+  assert (myu.s4->a == 1);
+  assert (myu.s4->b == 1);
+  assert (myu.s4->c == &c4 + 1);
+  assert (myu.s4->d == &d4 + 1);
+
+  delete s4;
+  delete t4;
+}
+#endif
+
+#ifdef NONREF_COMPONENT_MEMBER_SLICE
+template<typename A>
+void
+nonref_component_member_slice (void)
+{
+  INIT_ST;
+  U<A> myu(s1, t1, s2, t2, &s3, &t3, s4, t4);
+
+  #pragma omp target map(myu.t1.a[2:5])
+  {
+    myu.t1.a[2]++;
+  }
+
+  #pragma omp target map(myu.t1.b[2:5])
+  {
+    myu.t1.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu.t1.c)
+
+  #pragma omp target map(myu.t1.c[2:5])
+  {
+    myu.t1.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t1.c)
+
+  #pragma omp target enter data map(to: myu.t1.d)
+
+  #pragma omp target map(myu.t1.d[2:5])
+  {
+    myu.t1.d[2]++;
+  }
+
+  #pragma omp target exit data map(from: myu.t1.d)
+
+  assert (myu.t1.a[2] == 1);
+  assert (myu.t1.b[2] == 3);
+  assert (myu.t1.c[2] == 3);
+  assert (myu.t1.d[2] == 3);
+
+  #pragma omp target map(myu.t2.a[2:5])
+  {
+    myu.t2.a[2]++;
+  }
+
+  #pragma omp target map(myu.t2.b[2:5])
+  {
+    myu.t2.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu.t2.c)
+
+  #pragma omp target map(myu.t2.c[2:5])
+  {
+    myu.t2.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t2.c)
+
+  #pragma omp target enter data map(to: myu.t2.d)
+
+  #pragma omp target map(myu.t2.d[2:5])
+  {
+    myu.t2.d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t2.d)
+
+  assert (myu.t2.a[2] == 1);
+  assert (myu.t2.b[2] == 3);
+  assert (myu.t2.c[2] == 3);
+  assert (myu.t2.d[2] == 3);
+
+  #pragma omp target enter data map(to: myu.t3)
+
+  #pragma omp target map(myu.t3->a[2:5])
+  {
+    myu.t3->a[2]++;
+  }
+
+  #pragma omp target map(myu.t3->b[2:5])
+  {
+    myu.t3->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu.t3->c)
+
+  #pragma omp target map(myu.t3->c[2:5])
+  {
+    myu.t3->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t3->c)
+
+  #pragma omp target enter data map(to: myu.t3->d)
+
+  #pragma omp target map(myu.t3->d[2:5])
+  {
+    myu.t3->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t3, myu.t3->d)
+
+  assert (myu.t3->a[2] == 1);
+  assert (myu.t3->b[2] == 3);
+  assert (myu.t3->c[2] == 3);
+  assert (myu.t3->d[2] == 3);
+
+  #pragma omp target enter data map(to: myu.t4)
+
+  #pragma omp target map(myu.t4->a[2:5])
+  {
+    myu.t4->a[2]++;
+  }
+
+  #pragma omp target map(myu.t4->b[2:5])
+  {
+    myu.t4->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu.t4->c)
+
+  #pragma omp target map(myu.t4->c[2:5])
+  {
+    myu.t4->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t4->c)
+
+  #pragma omp target enter data map(to: myu.t4->d)
+
+  #pragma omp target map(myu.t4->d[2:5])
+  {
+    myu.t4->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t4, myu.t4->d)
+
+  assert (myu.t4->a[2] == 1);
+  assert (myu.t4->b[2] == 3);
+  assert (myu.t4->c[2] == 3);
+  assert (myu.t4->d[2] == 3);
+
+  delete s4;
+  delete t4;
+}
+#endif
+
+#ifdef NONREF_COMPONENT_MEMBER_SLICE_BASEPTR
+template<typename A>
+void
+nonref_component_member_slice_baseptr (void)
+{
+  INIT_ST;
+  U<A> myu(s1, t1, s2, t2, &s3, &t3, s4, t4);
+
+  #pragma omp target map(to: myu.t1.c) map(myu.t1.c[2:5])
+  {
+    myu.t1.c[2]++;
+  }
+
+  #pragma omp target map(to: myu.t1.d) map(myu.t1.d[2:5])
+  {
+    myu.t1.d[2]++;
+  }
+
+  assert (myu.t1.c[2] == 2);
+  assert (myu.t1.d[2] == 2);
+
+  #pragma omp target map(to: myu.t2.c) map(myu.t2.c[2:5])
+  {
+    myu.t2.c[2]++;
+  }
+
+  #pragma omp target map(to: myu.t2.d) map(myu.t2.d[2:5])
+  {
+    myu.t2.d[2]++;
+  }
+
+  assert (myu.t2.c[2] == 2);
+  assert (myu.t2.d[2] == 2);
+
+  #pragma omp target map(to: myu.t3, myu.t3->c) map(myu.t3->c[2:5])
+  {
+    myu.t3->c[2]++;
+  }
+
+  #pragma omp target map(to: myu.t3, myu.t3->d) map(myu.t3->d[2:5])
+  {
+    myu.t3->d[2]++;
+  }
+
+  assert (myu.t3->c[2] == 2);
+  assert (myu.t3->d[2] == 2);
+
+  #pragma omp target map(to: myu.t4, myu.t4->c) map(myu.t4->c[2:5])
+  {
+    myu.t4->c[2]++;
+  }
+
+  #pragma omp target map(to: myu.t4, myu.t4->d) map(myu.t4->d[2:5])
+  {
+    myu.t4->d[2]++;
+  }
+
+  assert (myu.t4->c[2] == 2);
+  assert (myu.t4->d[2] == 2);
+
+  delete s4;
+  delete t4;
+}
+#endif
+
+#ifdef REF_COMPONENT_BASE
+template<typename A>
+void
+ref_component_base (void)
+{
+  INIT_ST;
+  U<A> myu_real(s1, t1, s2, t2, &s3, &t3, s4, t4);
+  U<A> &myu = myu_real;
+
+  #pragma omp target map(myu.s1.a, myu.s1.b, myu.s1.c, myu.s1.d)
+  {
+    myu.s1.a++;
+    myu.s1.b++;
+    myu.s1.c++;
+    myu.s1.d++;
+  }
+
+  assert (myu.s1.a == 1);
+  assert (myu.s1.b == 1);
+  assert (myu.s1.c == &c1 + 1);
+  assert (myu.s1.d == &d1 + 1);
+
+  #pragma omp target map(myu.s2.a, myu.s2.b, myu.s2.c, myu.s2.d)
+  {
+    myu.s2.a++;
+    myu.s2.b++;
+    myu.s2.c++;
+    myu.s2.d++;
+  }
+
+  assert (myu.s2.a == 1);
+  assert (myu.s2.b == 1);
+  assert (myu.s2.c == &c2 + 1);
+  assert (myu.s2.d == &d2 + 1);
+
+  #pragma omp target map(to:myu.s3) \
+		     map(myu.s3->a, myu.s3->b, myu.s3->c, myu.s3->d)
+  {
+    myu.s3->a++;
+    myu.s3->b++;
+    myu.s3->c++;
+    myu.s3->d++;
+  }
+
+  assert (myu.s3->a == 1);
+  assert (myu.s3->b == 1);
+  assert (myu.s3->c == &c3 + 1);
+  assert (myu.s3->d == &d3 + 1);
+
+  #pragma omp target map(to:myu.s4) \
+		     map(myu.s4->a, myu.s4->b, myu.s4->c, myu.s4->d)
+  {
+    myu.s4->a++;
+    myu.s4->b++;
+    myu.s4->c++;
+    myu.s4->d++;
+  }
+
+  assert (myu.s4->a == 1);
+  assert (myu.s4->b == 1);
+  assert (myu.s4->c == &c4 + 1);
+  assert (myu.s4->d == &d4 + 1);
+
+  delete s4;
+  delete t4;
+}
+#endif
+
+#ifdef REF_COMPONENT_MEMBER_SLICE
+template<typename A>
+void
+ref_component_member_slice (void)
+{
+  INIT_ST;
+  U<A> myu_real(s1, t1, s2, t2, &s3, &t3, s4, t4);
+  U<A> &myu = myu_real;
+
+  #pragma omp target map(myu.t1.a[2:5])
+  {
+    myu.t1.a[2]++;
+  }
+
+  #pragma omp target map(myu.t1.b[2:5])
+  {
+    myu.t1.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu.t1.c)
+
+  #pragma omp target map(myu.t1.c[2:5])
+  {
+    myu.t1.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t1.c)
+
+  #pragma omp target enter data map(to: myu.t1.d)
+
+  #pragma omp target map(myu.t1.d[2:5])
+  {
+    myu.t1.d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t1.d)
+
+  assert (myu.t1.a[2] == 1);
+  assert (myu.t1.b[2] == 3);
+  assert (myu.t1.c[2] == 3);
+  assert (myu.t1.d[2] == 3);
+
+  #pragma omp target map(myu.t2.a[2:5])
+  {
+    myu.t2.a[2]++;
+  }
+
+  #pragma omp target map(myu.t2.b[2:5])
+  {
+    myu.t2.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu.t2.c)
+
+  #pragma omp target map(myu.t2.c[2:5])
+  {
+    myu.t2.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t2.c)
+
+  #pragma omp target enter data map(to: myu.t2.d)
+
+  #pragma omp target map(myu.t2.d[2:5])
+  {
+    myu.t2.d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t2.d)
+
+  assert (myu.t2.a[2] == 1);
+  assert (myu.t2.b[2] == 3);
+  assert (myu.t2.c[2] == 3);
+  assert (myu.t2.d[2] == 3);
+
+  #pragma omp target enter data map(to: myu.t3)
+
+  #pragma omp target map(myu.t3->a[2:5])
+  {
+    myu.t3->a[2]++;
+  }
+
+  #pragma omp target map(myu.t3->b[2:5])
+  {
+    myu.t3->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu.t3->c)
+
+  #pragma omp target map(myu.t3->c[2:5])
+  {
+    myu.t3->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t3->c)
+
+  #pragma omp target enter data map(to: myu.t3->d)
+
+  #pragma omp target map(myu.t3->d[2:5])
+  {
+    myu.t3->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t3, myu.t3->d)
+
+  assert (myu.t3->a[2] == 1);
+  assert (myu.t3->b[2] == 3);
+  assert (myu.t3->c[2] == 3);
+  assert (myu.t3->d[2] == 3);
+
+  #pragma omp target enter data map(to: myu.t4)
+
+  #pragma omp target map(myu.t4->a[2:5])
+  {
+    myu.t4->a[2]++;
+  }
+
+  #pragma omp target map(myu.t4->b[2:5])
+  {
+    myu.t4->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu.t4->c)
+
+  #pragma omp target map(myu.t4->c[2:5])
+  {
+    myu.t4->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t4->c)
+
+  #pragma omp target enter data map(to: myu.t4->d)
+
+  #pragma omp target map(myu.t4->d[2:5])
+  {
+    myu.t4->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu.t4, myu.t4->d)
+
+  assert (myu.t4->a[2] == 1);
+  assert (myu.t4->b[2] == 3);
+  assert (myu.t4->c[2] == 3);
+  assert (myu.t4->d[2] == 3);
+
+  delete s4;
+  delete t4;
+}
+#endif
+
+#ifdef REF_COMPONENT_MEMBER_SLICE_BASEPTR
+template<typename A>
+void
+ref_component_member_slice_baseptr (void)
+{
+  INIT_ST;
+  U<A> myu_real(s1, t1, s2, t2, &s3, &t3, s4, t4);
+  U<A> &myu = myu_real;
+
+  #pragma omp target map(to: myu.t1.c) map(myu.t1.c[2:5])
+  {
+    myu.t1.c[2]++;
+  }
+
+  #pragma omp target map(to: myu.t1.d) map(myu.t1.d[2:5])
+  {
+    myu.t1.d[2]++;
+  }
+
+  assert (myu.t1.c[2] == 2);
+  assert (myu.t1.d[2] == 2);
+
+  #pragma omp target map(to: myu.t2.c) map(myu.t2.c[2:5])
+  {
+    myu.t2.c[2]++;
+  }
+
+  #pragma omp target map(to: myu.t2.d) map(myu.t2.d[2:5])
+  {
+    myu.t2.d[2]++;
+  }
+
+  assert (myu.t2.c[2] == 2);
+  assert (myu.t2.d[2] == 2);
+
+  #pragma omp target map(to: myu.t3, myu.t3->c) map(myu.t3->c[2:5])
+  {
+    myu.t3->c[2]++;
+  }
+
+  #pragma omp target map(to: myu.t3, myu.t3->d) map(myu.t3->d[2:5])
+  {
+    myu.t3->d[2]++;
+  }
+
+  assert (myu.t3->c[2] == 2);
+  assert (myu.t3->d[2] == 2);
+
+  #pragma omp target map(to: myu.t4, myu.t4->c) map(myu.t4->c[2:5])
+  {
+    myu.t4->c[2]++;
+  }
+
+  #pragma omp target map(to: myu.t4, myu.t4->d) map(myu.t4->d[2:5])
+  {
+    myu.t4->d[2]++;
+  }
+
+  assert (myu.t4->c[2] == 2);
+  assert (myu.t4->d[2] == 2);
+
+  delete s4;
+  delete t4;
+}
+#endif
+
+#ifdef PTR_COMPONENT_BASE
+template<typename A>
+void
+ptr_component_base (void)
+{
+  INIT_ST;
+  U<A> *myu = new U<A>(s1, t1, s2, t2, &s3, &t3, s4, t4);
+
+  #pragma omp target map(myu->s1.a, myu->s1.b, myu->s1.c, myu->s1.d)
+  {
+    myu->s1.a++;
+    myu->s1.b++;
+    myu->s1.c++;
+    myu->s1.d++;
+  }
+
+  assert (myu->s1.a == 1);
+  assert (myu->s1.b == 1);
+  assert (myu->s1.c == &c1 + 1);
+  assert (myu->s1.d == &d1 + 1);
+
+  #pragma omp target map(myu->s2.a, myu->s2.b, myu->s2.c, myu->s2.d)
+  {
+    myu->s2.a++;
+    myu->s2.b++;
+    myu->s2.c++;
+    myu->s2.d++;
+  }
+
+  assert (myu->s2.a == 1);
+  assert (myu->s2.b == 1);
+  assert (myu->s2.c == &c2 + 1);
+  assert (myu->s2.d == &d2 + 1);
+
+  #pragma omp target map(to:myu->s3) \
+		     map(myu->s3->a, myu->s3->b, myu->s3->c, myu->s3->d)
+  {
+    myu->s3->a++;
+    myu->s3->b++;
+    myu->s3->c++;
+    myu->s3->d++;
+  }
+
+  assert (myu->s3->a == 1);
+  assert (myu->s3->b == 1);
+  assert (myu->s3->c == &c3 + 1);
+  assert (myu->s3->d == &d3 + 1);
+
+  #pragma omp target map(to:myu->s4) \
+		     map(myu->s4->a, myu->s4->b, myu->s4->c, myu->s4->d)
+  {
+    myu->s4->a++;
+    myu->s4->b++;
+    myu->s4->c++;
+    myu->s4->d++;
+  }
+
+  assert (myu->s4->a == 1);
+  assert (myu->s4->b == 1);
+  assert (myu->s4->c == &c4 + 1);
+  assert (myu->s4->d == &d4 + 1);
+
+  delete s4;
+  delete t4;
+  delete myu;
+}
+#endif
+
+#ifdef PTR_COMPONENT_MEMBER_SLICE
+template<typename A>
+void
+ptr_component_member_slice (void)
+{
+  INIT_ST;
+  U<A> *myu = new U<A>(s1, t1, s2, t2, &s3, &t3, s4, t4);
+
+  #pragma omp target map(myu->t1.a[2:5])
+  {
+    myu->t1.a[2]++;
+  }
+
+  #pragma omp target map(myu->t1.b[2:5])
+  {
+    myu->t1.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu->t1.c)
+
+  #pragma omp target map(myu->t1.c[2:5])
+  {
+    myu->t1.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t1.c)
+
+  #pragma omp target enter data map(to: myu->t1.d)
+
+  #pragma omp target map(myu->t1.d[2:5])
+  {
+    myu->t1.d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t1.d)
+
+  assert (myu->t1.a[2] == 1);
+  assert (myu->t1.b[2] == 3);
+  assert (myu->t1.c[2] == 3);
+  assert (myu->t1.d[2] == 3);
+
+  #pragma omp target map(myu->t2.a[2:5])
+  {
+    myu->t2.a[2]++;
+  }
+
+  #pragma omp target map(myu->t2.b[2:5])
+  {
+    myu->t2.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu->t2.c)
+
+  #pragma omp target map(myu->t2.c[2:5])
+  {
+    myu->t2.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t2.c)
+
+  #pragma omp target enter data map(to: myu->t2.d)
+
+  #pragma omp target map(myu->t2.d[2:5])
+  {
+    myu->t2.d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t2.d)
+
+  assert (myu->t2.a[2] == 1);
+  assert (myu->t2.b[2] == 3);
+  assert (myu->t2.c[2] == 3);
+  assert (myu->t2.d[2] == 3);
+
+  #pragma omp target enter data map(to: myu->t3)
+
+  #pragma omp target map(myu->t3->a[2:5])
+  {
+    myu->t3->a[2]++;
+  }
+
+  #pragma omp target map(myu->t3->b[2:5])
+  {
+    myu->t3->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu->t3->c)
+
+  #pragma omp target map(myu->t3->c[2:5])
+  {
+    myu->t3->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t3->c)
+
+  #pragma omp target enter data map(to: myu->t3->d)
+
+  #pragma omp target map(myu->t3->d[2:5])
+  {
+    myu->t3->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t3, myu->t3->d)
+
+  assert (myu->t3->a[2] == 1);
+  assert (myu->t3->b[2] == 3);
+  assert (myu->t3->c[2] == 3);
+  assert (myu->t3->d[2] == 3);
+
+  #pragma omp target enter data map(to: myu->t4)
+
+  #pragma omp target map(myu->t4->a[2:5])
+  {
+    myu->t4->a[2]++;
+  }
+
+  #pragma omp target map(myu->t4->b[2:5])
+  {
+    myu->t4->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu->t4->c)
+
+  #pragma omp target map(myu->t4->c[2:5])
+  {
+    myu->t4->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t4->c)
+
+  #pragma omp target enter data map(to: myu->t4->d)
+
+  #pragma omp target map(myu->t4->d[2:5])
+  {
+    myu->t4->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t4, myu->t4->d)
+
+  assert (myu->t4->a[2] == 1);
+  assert (myu->t4->b[2] == 3);
+  assert (myu->t4->c[2] == 3);
+  assert (myu->t4->d[2] == 3);
+
+  delete s4;
+  delete t4;
+  delete myu;
+}
+#endif
+
+#ifdef PTR_COMPONENT_MEMBER_SLICE_BASEPTR
+template<typename A>
+void
+ptr_component_member_slice_baseptr (void)
+{
+  INIT_ST;
+  U<A> *myu = new U<A>(s1, t1, s2, t2, &s3, &t3, s4, t4);
+
+  /* Implicit firstprivate 'myu'.  */
+  #pragma omp target map(to: myu->t1.c) map(myu->t1.c[2:5])
+  {
+    myu->t1.c[2]++;
+  }
+
+  #pragma omp target map(to: myu->t1.d) map(myu->t1.d[2:5])
+  {
+    myu->t1.d[2]++;
+  }
+
+  assert (myu->t1.c[2] == 2);
+  assert (myu->t1.d[2] == 2);
+
+  /* Explicitly-mapped 'myu'.  */
+  #pragma omp target map(to: myu, myu->t1.c) map(myu->t1.c[2:5])
+  {
+    myu->t1.c[2]++;
+  }
+
+  #pragma omp target map(to: myu, myu->t1.d) map(myu->t1.d[2:5])
+  {
+    myu->t1.d[2]++;
+  }
+
+  assert (myu->t1.c[2] == 4);
+  assert (myu->t1.d[2] == 4);
+
+  /* Implicit firstprivate 'myu'.  */
+  #pragma omp target map(to: myu->t2.c) map(myu->t2.c[2:5])
+  {
+    myu->t2.c[2]++;
+  }
+
+  #pragma omp target map(to: myu->t2.d) map(myu->t2.d[2:5])
+  {
+    myu->t2.d[2]++;
+  }
+
+  assert (myu->t2.c[2] == 2);
+  assert (myu->t2.d[2] == 2);
+
+  /* Explicitly-mapped 'myu'.  */
+  #pragma omp target map(to: myu, myu->t2.c) map(myu->t2.c[2:5])
+  {
+    myu->t2.c[2]++;
+  }
+
+  #pragma omp target map(to: myu, myu->t2.d) map(myu->t2.d[2:5])
+  {
+    myu->t2.d[2]++;
+  }
+
+  assert (myu->t2.c[2] == 4);
+  assert (myu->t2.d[2] == 4);
+
+  /* Implicit firstprivate 'myu'.  */
+  #pragma omp target map(to: myu->t3, myu->t3->c) map(myu->t3->c[2:5])
+  {
+    myu->t3->c[2]++;
+  }
+
+  #pragma omp target map(to: myu->t3, myu->t3->d) map(myu->t3->d[2:5])
+  {
+    myu->t3->d[2]++;
+  }
+
+  assert (myu->t3->c[2] == 2);
+  assert (myu->t3->d[2] == 2);
+
+  /* Explicitly-mapped 'myu'.  */
+  #pragma omp target map(to: myu, myu->t3, myu->t3->c) map(myu->t3->c[2:5])
+  {
+    myu->t3->c[2]++;
+  }
+
+  #pragma omp target map(to: myu, myu->t3, myu->t3->d) map(myu->t3->d[2:5])
+  {
+    myu->t3->d[2]++;
+  }
+
+  assert (myu->t3->c[2] == 4);
+  assert (myu->t3->d[2] == 4);
+
+  /* Implicit firstprivate 'myu'.  */
+  #pragma omp target map(to: myu->t4, myu->t4->c) map(myu->t4->c[2:5])
+  {
+    myu->t4->c[2]++;
+  }
+
+  #pragma omp target map(to: myu->t4, myu->t4->d) map(myu->t4->d[2:5])
+  {
+    myu->t4->d[2]++;
+  }
+
+  assert (myu->t4->c[2] == 2);
+  assert (myu->t4->d[2] == 2);
+
+  /* Explicitly-mapped 'myu'.  */
+  #pragma omp target map(to: myu, myu->t4, myu->t4->c) map(myu->t4->c[2:5])
+  {
+    myu->t4->c[2]++;
+  }
+
+  #pragma omp target map(to: myu, myu->t4, myu->t4->d) map(myu->t4->d[2:5])
+  {
+    myu->t4->d[2]++;
+  }
+
+  assert (myu->t4->c[2] == 4);
+  assert (myu->t4->d[2] == 4);
+
+  delete s4;
+  delete t4;
+  delete myu;
+}
+#endif
+
+#ifdef REF2PTR_COMPONENT_BASE
+template<typename A>
+void
+ref2ptr_component_base (void)
+{
+  INIT_ST;
+  U<A> *myu_ptr = new U<A>(s1, t1, s2, t2, &s3, &t3, s4, t4);
+  U<A> *&myu = myu_ptr;
+
+  #pragma omp target map(myu->s1.a, myu->s1.b, myu->s1.c, myu->s1.d)
+  {
+    myu->s1.a++;
+    myu->s1.b++;
+    myu->s1.c++;
+    myu->s1.d++;
+  }
+
+  assert (myu->s1.a == 1);
+  assert (myu->s1.b == 1);
+  assert (myu->s1.c == &c1 + 1);
+  assert (myu->s1.d == &d1 + 1);
+
+  #pragma omp target map(myu->s2.a, myu->s2.b, myu->s2.c, myu->s2.d)
+  {
+    myu->s2.a++;
+    myu->s2.b++;
+    myu->s2.c++;
+    myu->s2.d++;
+  }
+
+  assert (myu->s2.a == 1);
+  assert (myu->s2.b == 1);
+  assert (myu->s2.c == &c2 + 1);
+  assert (myu->s2.d == &d2 + 1);
+
+  #pragma omp target map(to:myu->s3) \
+		     map(myu->s3->a, myu->s3->b, myu->s3->c, myu->s3->d)
+  {
+    myu->s3->a++;
+    myu->s3->b++;
+    myu->s3->c++;
+    myu->s3->d++;
+  }
+
+  assert (myu->s3->a == 1);
+  assert (myu->s3->b == 1);
+  assert (myu->s3->c == &c3 + 1);
+  assert (myu->s3->d == &d3 + 1);
+
+  #pragma omp target map(to:myu->s4) \
+		     map(myu->s4->a, myu->s4->b, myu->s4->c, myu->s4->d)
+  {
+    myu->s4->a++;
+    myu->s4->b++;
+    myu->s4->c++;
+    myu->s4->d++;
+  }
+
+  assert (myu->s4->a == 1);
+  assert (myu->s4->b == 1);
+  assert (myu->s4->c == &c4 + 1);
+  assert (myu->s4->d == &d4 + 1);
+
+  delete s4;
+  delete t4;
+  delete myu_ptr;
+}
+#endif
+
+#ifdef REF2PTR_COMPONENT_MEMBER_SLICE
+template<typename A>
+void
+ref2ptr_component_member_slice (void)
+{
+  INIT_ST;
+  U<A> *myu_ptr = new U<A>(s1, t1, s2, t2, &s3, &t3, s4, t4);
+  U<A> *&myu = myu_ptr;
+
+  #pragma omp target map(myu->t1.a[2:5])
+  {
+    myu->t1.a[2]++;
+  }
+
+  #pragma omp target map(myu->t1.b[2:5])
+  {
+    myu->t1.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu->t1.c)
+
+  #pragma omp target map(myu->t1.c[2:5])
+  {
+    myu->t1.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t1.c)
+
+  #pragma omp target enter data map(to: myu->t1.d)
+
+  #pragma omp target map(myu->t1.d[2:5])
+  {
+    myu->t1.d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t1.d)
+
+  assert (myu->t1.a[2] == 1);
+  assert (myu->t1.b[2] == 3);
+  assert (myu->t1.c[2] == 3);
+  assert (myu->t1.d[2] == 3);
+
+  #pragma omp target map(myu->t2.a[2:5])
+  {
+    myu->t2.a[2]++;
+  }
+
+  #pragma omp target map(myu->t2.b[2:5])
+  {
+    myu->t2.b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu->t2.c)
+
+  #pragma omp target map(myu->t2.c[2:5])
+  {
+    myu->t2.c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t2.c)
+
+  #pragma omp target enter data map(to: myu->t2.d)
+
+  #pragma omp target map(myu->t2.d[2:5])
+  {
+    myu->t2.d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t2.d)
+
+  assert (myu->t2.a[2] == 1);
+  assert (myu->t2.b[2] == 3);
+  assert (myu->t2.c[2] == 3);
+  assert (myu->t2.d[2] == 3);
+
+  #pragma omp target enter data map(to: myu->t3)
+
+  #pragma omp target map(myu->t3->a[2:5])
+  {
+    myu->t3->a[2]++;
+  }
+
+  #pragma omp target map(myu->t3->b[2:5])
+  {
+    myu->t3->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu->t3->c)
+
+  #pragma omp target map(myu->t3->c[2:5])
+  {
+    myu->t3->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t3->c)
+
+  #pragma omp target enter data map(to: myu->t3->d)
+
+  #pragma omp target map(myu->t3->d[2:5])
+  {
+    myu->t3->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t3, myu->t3->d)
+
+  assert (myu->t3->a[2] == 1);
+  assert (myu->t3->b[2] == 3);
+  assert (myu->t3->c[2] == 3);
+  assert (myu->t3->d[2] == 3);
+
+  #pragma omp target enter data map(to: myu->t4)
+
+  #pragma omp target map(myu->t4->a[2:5])
+  {
+    myu->t4->a[2]++;
+  }
+
+  #pragma omp target map(myu->t4->b[2:5])
+  {
+    myu->t4->b[2]++;
+  }
+
+  #pragma omp target enter data map(to: myu->t4->c)
+
+  #pragma omp target map(myu->t4->c[2:5])
+  {
+    myu->t4->c[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t4->c)
+
+  #pragma omp target enter data map(to: myu->t4->d)
+
+  #pragma omp target map(myu->t4->d[2:5])
+  {
+    myu->t4->d[2]++;
+  }
+
+  #pragma omp target exit data map(release: myu->t4, myu->t4->d)
+
+  assert (myu->t4->a[2] == 1);
+  assert (myu->t4->b[2] == 3);
+  assert (myu->t4->c[2] == 3);
+  assert (myu->t4->d[2] == 3);
+
+  delete s4;
+  delete t4;
+  delete myu_ptr;
+}
+#endif
+
+#ifdef REF2PTR_COMPONENT_MEMBER_SLICE_BASEPTR
+template<typename A>
+void
+ref2ptr_component_member_slice_baseptr (void)
+{
+  INIT_ST;
+  U<A> *myu_ptr = new U<A>(s1, t1, s2, t2, &s3, &t3, s4, t4);
+  U<A> *&myu = myu_ptr;
+
+  /* Implicit firstprivate 'myu'.  */
+  #pragma omp target map(to: myu->t1.c) map(myu->t1.c[2:5])
+  {
+    myu->t1.c[2]++;
+  }
+
+  #pragma omp target map(to: myu->t1.d) map(myu->t1.d[2:5])
+  {
+    myu->t1.d[2]++;
+  }
+
+  assert (myu->t1.c[2] == 2);
+  assert (myu->t1.d[2] == 2);
+
+  /* Explicitly-mapped 'myu'.  */
+  #pragma omp target map(to: myu, myu->t1.c) map(myu->t1.c[2:5])
+  {
+    myu->t1.c[2]++;
+  }
+
+  #pragma omp target map(to: myu, myu->t1.d) map(myu->t1.d[2:5])
+  {
+    myu->t1.d[2]++;
+  }
+
+  assert (myu->t1.c[2] == 4);
+  assert (myu->t1.d[2] == 4);
+
+  /* Implicit firstprivate 'myu'.  */
+  #pragma omp target map(to: myu->t2.c) map(myu->t2.c[2:5])
+  {
+    myu->t2.c[2]++;
+  }
+
+  #pragma omp target map(to: myu->t2.d) map(myu->t2.d[2:5])
+  {
+    myu->t2.d[2]++;
+  }
+
+  assert (myu->t2.c[2] == 2);
+  assert (myu->t2.d[2] == 2);
+
+  /* Explicitly-mapped 'myu'.  */
+  #pragma omp target map(to: myu, myu->t2.c) map(myu->t2.c[2:5])
+  {
+    myu->t2.c[2]++;
+  }
+
+  #pragma omp target map(to: myu, myu->t2.d) map(myu->t2.d[2:5])
+  {
+    myu->t2.d[2]++;
+  }
+
+  assert (myu->t2.c[2] == 4);
+  assert (myu->t2.d[2] == 4);
+
+  /* Implicit firstprivate 'myu'.  */
+  #pragma omp target map(to: myu->t3, myu->t3->c) map(myu->t3->c[2:5])
+  {
+    myu->t3->c[2]++;
+  }
+
+  #pragma omp target map(to: myu->t3, myu->t3->d) map(myu->t3->d[2:5])
+  {
+    myu->t3->d[2]++;
+  }
+
+  assert (myu->t3->c[2] == 2);
+  assert (myu->t3->d[2] == 2);
+
+  /* Explicitly-mapped 'myu'.  */
+  #pragma omp target map(to: myu, myu->t3, myu->t3->c) map(myu->t3->c[2:5])
+  {
+    myu->t3->c[2]++;
+  }
+
+  #pragma omp target map(to: myu, myu->t3, myu->t3->d) map(myu->t3->d[2:5])
+  {
+    myu->t3->d[2]++;
+  }
+
+  assert (myu->t3->c[2] == 4);
+  assert (myu->t3->d[2] == 4);
+
+  /* Implicit firstprivate 'myu'.  */
+  #pragma omp target map(to: myu->t4, myu->t4->c) map(myu->t4->c[2:5])
+  {
+    myu->t4->c[2]++;
+  }
+
+  #pragma omp target map(to: myu->t4, myu->t4->d) map(myu->t4->d[2:5])
+  {
+    myu->t4->d[2]++;
+  }
+
+  assert (myu->t4->c[2] == 2);
+  assert (myu->t4->d[2] == 2);
+
+  /* Explicitly-mapped 'myu'.  */
+  #pragma omp target map(to: myu, myu->t4, myu->t4->c) map(myu->t4->c[2:5])
+  {
+    myu->t4->c[2]++;
+  }
+
+  #pragma omp target map(to: myu, myu->t4, myu->t4->d) map(myu->t4->d[2:5])
+  {
+    myu->t4->d[2]++;
+  }
+
+  assert (myu->t4->c[2] == 4);
+  assert (myu->t4->d[2] == 4);
+
+  delete s4;
+  delete t4;
+  delete myu_ptr;
+}
+#endif
+
+int main (int argc, char *argv[])
+{
+#ifdef MAP_DECLS
+  map_decls<4> ();
+#endif
+
+#ifdef NONREF_DECL_BASE
+  nonref_decl_base<int> ();
+#endif
+#ifdef REF_DECL_BASE
+  ref_decl_base<int> ();
+#endif
+#ifdef PTR_DECL_BASE
+  ptr_decl_base<int> ();
+#endif
+#ifdef REF2PTR_DECL_BASE
+  ref2ptr_decl_base<int> ();
+#endif
+
+#ifdef ARRAY_DECL_BASE
+  array_decl_base<int> ();
+#endif
+#ifdef REF2ARRAY_DECL_BASE
+  ref2array_decl_base<int> ();
+#endif
+#ifdef PTR_OFFSET_DECL_BASE
+  ptr_offset_decl_base<int> ();
+#endif
+#ifdef REF2PTR_OFFSET_DECL_BASE
+  ref2ptr_offset_decl_base<int> ();
+#endif
+
+#ifdef MAP_SECTIONS
+  map_sections<int, 10> ();
+#endif
+
+#ifdef NONREF_DECL_MEMBER_SLICE
+  nonref_decl_member_slice<int, 10> ();
+#endif
+#ifdef NONREF_DECL_MEMBER_SLICE_BASEPTR
+  nonref_decl_member_slice_baseptr<int, 10> ();
+#endif
+#ifdef REF_DECL_MEMBER_SLICE
+  ref_decl_member_slice<int, 10> ();
+#endif
+#ifdef REF_DECL_MEMBER_SLICE_BASEPTR
+  ref_decl_member_slice_baseptr<int, 10> ();
+#endif
+#ifdef PTR_DECL_MEMBER_SLICE
+  ptr_decl_member_slice<int, 10> ();
+#endif
+#ifdef PTR_DECL_MEMBER_SLICE_BASEPTR
+  ptr_decl_member_slice_baseptr<int, 10> ();
+#endif
+#ifdef REF2PTR_DECL_MEMBER_SLICE
+  ref2ptr_decl_member_slice<int, 10> ();
+#endif
+#ifdef REF2PTR_DECL_MEMBER_SLICE_BASEPTR
+  ref2ptr_decl_member_slice_baseptr<int, 10> ();
+#endif
+
+#ifdef ARRAY_DECL_MEMBER_SLICE
+  array_decl_member_slice<int, 10> ();
+#endif
+#ifdef ARRAY_DECL_MEMBER_SLICE_BASEPTR
+  array_decl_member_slice_baseptr<int, 10> ();
+#endif
+#ifdef REF2ARRAY_DECL_MEMBER_SLICE
+  ref2array_decl_member_slice<int, 10> ();
+#endif
+#ifdef REF2ARRAY_DECL_MEMBER_SLICE_BASEPTR
+  ref2array_decl_member_slice_baseptr<int, 10> ();
+#endif
+#ifdef PTR_OFFSET_DECL_MEMBER_SLICE
+  ptr_offset_decl_member_slice<int, 10> ();
+#endif
+#ifdef PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+  ptr_offset_decl_member_slice_baseptr<int, 10> ();
+#endif
+#ifdef REF2PTR_OFFSET_DECL_MEMBER_SLICE
+  ref2ptr_offset_decl_member_slice<int, 10> ();
+#endif
+#ifdef REF2PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+  ref2ptr_offset_decl_member_slice_baseptr<int, 10> ();
+#endif
+
+#ifdef PTRARRAY_DECL_MEMBER_SLICE
+  ptrarray_decl_member_slice<int, 10> ();
+#endif
+#ifdef PTRARRAY_DECL_MEMBER_SLICE_BASEPTR
+  ptrarray_decl_member_slice_baseptr<int, 10> ();
+#endif
+#ifdef REF2PTRARRAY_DECL_MEMBER_SLICE
+  ref2ptrarray_decl_member_slice<int, 10> ();
+#endif
+#ifdef REF2PTRARRAY_DECL_MEMBER_SLICE_BASEPTR
+  ref2ptrarray_decl_member_slice_baseptr<int, 10> ();
+#endif
+#ifdef PTRPTR_OFFSET_DECL_MEMBER_SLICE
+  ptrptr_offset_decl_member_slice<int, 10> ();
+#endif
+#ifdef PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+  ptrptr_offset_decl_member_slice_baseptr<int, 10> ();
+#endif
+#ifdef REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE
+  ref2ptrptr_offset_decl_member_slice<int, 10> ();
+#endif
+#ifdef REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR
+  ref2ptrptr_offset_decl_member_slice_baseptr<int, 10> ();
+#endif
+
+#ifdef NONREF_COMPONENT_BASE
+  nonref_component_base<int> ();
+#endif
+#ifdef NONREF_COMPONENT_MEMBER_SLICE
+  nonref_component_member_slice<int> ();
+#endif
+#ifdef NONREF_COMPONENT_MEMBER_SLICE_BASEPTR
+  nonref_component_member_slice_baseptr<int> ();
+#endif
+
+#ifdef REF_COMPONENT_BASE
+  ref_component_base<int> ();
+#endif
+#ifdef REF_COMPONENT_MEMBER_SLICE
+  ref_component_member_slice<int> ();
+#endif
+#ifdef REF_COMPONENT_MEMBER_SLICE_BASEPTR
+  ref_component_member_slice_baseptr<int> ();
+#endif
+
+#ifdef PTR_COMPONENT_BASE
+  ptr_component_base<int> ();
+#endif
+#ifdef PTR_COMPONENT_MEMBER_SLICE
+  ptr_component_member_slice<int> ();
+#endif
+#ifdef PTR_COMPONENT_MEMBER_SLICE_BASEPTR
+  ptr_component_member_slice_baseptr<int> ();
+#endif
+
+#ifdef REF2PTR_COMPONENT_BASE
+  ref2ptr_component_base<int> ();
+#endif
+#ifdef REF2PTR_COMPONENT_MEMBER_SLICE
+  ref2ptr_component_member_slice<int> ();
+#endif
+#ifdef REF2PTR_COMPONENT_MEMBER_SLICE_BASEPTR
+  ref2ptr_component_member_slice_baseptr<int> ();
+#endif
+
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c++/ind-base-1.C b/libgomp/testsuite/libgomp.c++/ind-base-1.C
new file mode 100644
index 00000000000..4566854e60a
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/ind-base-1.C
@@ -0,0 +1,162 @@ 
+// { dg-do run }
+// { dg-options "-fopenmp" }
+
+#include <cassert>
+
+struct S
+{
+  int x[10];
+};
+
+struct T
+{
+  struct S *s;
+};
+
+struct U
+{
+  struct T *t;
+};
+
+void
+foo_siblist (void)
+{
+  U *u = new U;
+  u->t = new T;
+  u->t->s = new S;
+  for (int i = 0; i < 10; i++)
+    u->t->s->x[i] = 0;
+#pragma omp target map(u->t, *(u->t), u->t->s, *u->t->s)
+  for (int i = 0; i < 10; i++)
+    u->t->s->x[i] = i * 3;
+  for (int i = 0; i < 10; i++)
+    assert (u->t->s->x[i] == i * 3);
+  delete u->t->s;
+  delete u->t;
+  delete u;
+}
+
+void
+foo (void)
+{
+  U *u = new U;
+  u->t = new T;
+  u->t->s = new S;
+  for (int i = 0; i < 10; i++)
+    u->t->s->x[i] = 0;
+#pragma omp target map(*u, u->t, *(u->t), u->t->s, *u->t->s)
+  for (int i = 0; i < 10; i++)
+    u->t->s->x[i] = i * 3;
+  for (int i = 0; i < 10; i++)
+    assert (u->t->s->x[i] == i * 3);
+  delete u->t->s;
+  delete u->t;
+  delete u;
+}
+
+void
+foo_tofrom (void)
+{
+  U *u = new U;
+  u->t = new T;
+  u->t->s = new S;
+  for (int i = 0; i < 10; i++)
+    u->t->s->x[i] = 0;
+#pragma omp target map(u, *u, u->t, *(u->t), u->t->s, *u->t->s)
+  for (int i = 0; i < 10; i++)
+    u->t->s->x[i] = i * 3;
+  for (int i = 0; i < 10; i++)
+    assert (u->t->s->x[i] == i * 3);
+  delete u->t->s;
+  delete u->t;
+  delete u;
+}
+
+void
+bar (void)
+{
+  U *u = new U;
+  U **up = &u;
+  u->t = new T;
+  u->t->s = new S;
+  for (int i = 0; i < 10; i++)
+    (*up)->t->s->x[i] = 0;
+#pragma omp target map(*up, (*up)->t, *(*up)->t, (*up)->t->s, *(*up)->t->s)
+  for (int i = 0; i < 10; i++)
+    (*up)->t->s->x[i] = i * 3;
+  for (int i = 0; i < 10; i++)
+    assert ((*up)->t->s->x[i] == i * 3);
+  delete u->t->s;
+  delete u->t;
+  delete u;
+}
+
+void
+bar_pp (void)
+{
+  U *u = new U;
+  U **up = &u;
+  u->t = new T;
+  u->t->s = new S;
+  for (int i = 0; i < 10; i++)
+    (*up)->t->s->x[i] = 0;
+#pragma omp target map(*up, **up, (*up)->t, *(*up)->t, (*up)->t->s, *(*up)->t->s)
+  for (int i = 0; i < 10; i++)
+    (*up)->t->s->x[i] = i * 3;
+  for (int i = 0; i < 10; i++)
+    assert ((*up)->t->s->x[i] == i * 3);
+  delete u->t->s;
+  delete u->t;
+  delete u;
+}
+
+void
+bar_tofrom (void)
+{
+  U *u = new U;
+  U **up = &u;
+  u->t = new T;
+  u->t->s = new S;
+  for (int i = 0; i < 10; i++)
+    (*up)->t->s->x[i] = 0;
+#pragma omp target map(*up, up, (*up)->t, *(*up)->t, (*up)->t->s, *(*up)->t->s)
+  for (int i = 0; i < 10; i++)
+    (*up)->t->s->x[i] = i * 3;
+  for (int i = 0; i < 10; i++)
+    assert ((*up)->t->s->x[i] == i * 3);
+  delete u->t->s;
+  delete u->t;
+  delete u;
+}
+
+void
+bar_tofrom_pp (void)
+{
+  U *u = new U;
+  U **up = &u;
+  u->t = new T;
+  u->t->s = new S;
+  for (int i = 0; i < 10; i++)
+    (*up)->t->s->x[i] = 0;
+#pragma omp target map(**up, *up, up, (*up)->t, *(*up)->t, (*up)->t->s, \
+		       *(*up)->t->s)
+  for (int i = 0; i < 10; i++)
+    (*up)->t->s->x[i] = i * 3;
+  for (int i = 0; i < 10; i++)
+    assert ((*up)->t->s->x[i] == i * 3);
+  delete u->t->s;
+  delete u->t;
+  delete u;
+}
+
+int main (int argc, char *argv[])
+{
+  foo_siblist ();
+  foo ();
+  foo_tofrom ();
+  bar ();
+  bar_pp ();
+  bar_tofrom ();
+  bar_tofrom_pp ();
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c++/ind-base-2.C b/libgomp/testsuite/libgomp.c++/ind-base-2.C
new file mode 100644
index 00000000000..4c05c2ef8f6
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/ind-base-2.C
@@ -0,0 +1,93 @@ 
+// { dg-do run }
+// { dg-options "-fopenmp" }
+
+#include <cassert>
+
+struct S
+{
+  int x[10];
+};
+
+struct T
+{
+  struct S ***s;
+};
+
+struct U
+{
+  struct T **t;
+};
+
+void
+foo (void)
+{
+  U *u = new U;
+  T *real_t = new T;
+  S *real_s = new S;
+  T **t_pp = &real_t;
+  S **s_pp = &real_s;
+  S ***s_ppp = &s_pp;
+  u->t = t_pp;
+  (*u->t)->s = s_ppp;
+  for (int i = 0; i < 10; i++)
+    (**((*u->t)->s))->x[i] = 0;
+#pragma omp target map(u->t, *u->t, (*u->t)->s, *(*u->t)->s, **(*u->t)->s, \
+		       (**(*u->t)->s)->x[0:10])
+  for (int i = 0; i < 10; i++)
+    (**((*u->t)->s))->x[i] = i * 3;
+  for (int i = 0; i < 10; i++)
+    assert ((**((*u->t)->s))->x[i] == i * 3);
+  delete real_s;
+  delete real_t;
+  delete u;
+}
+
+template<typename X>
+struct St
+{
+  X x[10];
+};
+
+template<typename X>
+struct Tt
+{
+  X ***s;
+};
+
+template<typename X>
+struct Ut
+{
+  X **t;
+};
+
+template<typename I>
+void
+tfoo (void)
+{
+  Ut<Tt<St<I> > > *u = new Ut<Tt<St<I> > >;
+  Tt<St<I> > *real_t = new Tt<St<int> >;
+  St<I> *real_s = new St<int>;
+  Tt<St<I> > **t_pp = &real_t;
+  St<I> **s_pp = &real_s;
+  St<I> ***s_ppp = &s_pp;
+  u->t = t_pp;
+  (*u->t)->s = s_ppp;
+  for (int i = 0; i < 10; i++)
+    (**((*u->t)->s))->x[i] = 0;
+#pragma omp target map(u->t, *u->t, (*u->t)->s, *(*u->t)->s, **(*u->t)->s, \
+		       (**(*u->t)->s)->x[0:10])
+  for (int i = 0; i < 10; i++)
+    (**((*u->t)->s))->x[i] = i * 3;
+  for (int i = 0; i < 10; i++)
+    assert ((**((*u->t)->s))->x[i] == i * 3);
+  delete real_s;
+  delete real_t;
+  delete u;
+}
+
+int main (int argc, char *argv[])
+{
+  foo ();
+  tfoo<int> ();
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c++/lvalue-tofrom-1.C b/libgomp/testsuite/libgomp.c++/lvalue-tofrom-1.C
new file mode 100644
index 00000000000..643cfdb6e28
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/lvalue-tofrom-1.C
@@ -0,0 +1,75 @@ 
+#include <cstring>
+#include <cassert>
+
+static int lo()
+{
+  return 30;
+}
+
+static int len()
+{
+  return 10;
+}
+
+template<typename T>
+void foo ()
+{
+  T arr[100];
+  T *ptr;
+
+  memset (arr, '\0', sizeof arr);
+
+#pragma omp target enter data map(to: arr[0:100])
+
+  for (int i = 0; i < 100; i++)
+    arr[i] = i;
+
+  ptr = &arr[10];
+
+#pragma omp target update to(*ptr)
+
+  for (int i = lo (); i < lo () + len (); i++)
+    arr[i] = i * 2;
+
+#pragma omp target update to(arr[lo():len()])
+
+#pragma omp target exit data map(from: arr[0:100])
+
+  assert (arr[10] == 10);
+  for (int i = lo (); i < lo () + len (); i++)
+    assert (arr[i] == i * 2);
+}
+
+int
+main ()
+{
+  char arr[100];
+  char *ptr;
+
+  memset (arr, '\0', sizeof arr);
+
+#pragma omp target enter data map(to: arr[0:100])
+
+  for (int i = 0; i < 100; i++)
+    arr[i] = i;
+
+  ptr = &arr[10];
+
+#pragma omp target update to(*ptr)
+
+  for (int i = lo (); i < lo () + len (); i++)
+    arr[i] = i * 2;
+
+#pragma omp target update to(arr[lo():len()])
+
+#pragma omp target exit data map(from: arr[0:100])
+
+  assert (arr[10] == 10);
+  for (int i = lo (); i < lo () + len (); i++)
+    assert (arr[i] == i * 2);
+
+  foo<short> ();
+
+  return 0;
+}
+
diff --git a/libgomp/testsuite/libgomp.c++/lvalue-tofrom-2.C b/libgomp/testsuite/libgomp.c++/lvalue-tofrom-2.C
new file mode 100644
index 00000000000..adc493b1315
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/lvalue-tofrom-2.C
@@ -0,0 +1,71 @@ 
+#include <cstring>
+#include <cassert>
+#include <cstdlib>
+
+template<typename T>
+struct t_array_wrapper {
+  T *data;
+  unsigned int length;
+};
+
+template<typename T>
+void foo()
+{
+  struct t_array_wrapper<T> aw;
+
+  aw.data = new T[100];
+  aw.length = 100;
+
+#pragma omp target enter data map(to: aw.data, aw.length) \
+			      map(alloc: aw.data[0:aw.length])
+
+#pragma omp target
+  for (int i = 0; i < aw.length; i++)
+    aw.data[i] = i;
+
+#pragma omp target update from(aw.data[:aw.length])
+
+#pragma omp target exit data map(delete: aw.data, aw.length, \
+				 aw.data[0:aw.length])
+
+  for (int i = 0; i < aw.length; i++)
+    assert (aw.data[i] == i);
+
+  delete[] aw.data;
+}
+
+struct array_wrapper {
+  int *data;
+  unsigned int length;
+};
+
+int
+main ()
+{
+  struct array_wrapper aw;
+
+  aw.data = new int[100];
+  aw.length = 100;
+
+#pragma omp target enter data map(to: aw.data, aw.length) \
+			      map(alloc: aw.data[0:aw.length])
+
+#pragma omp target
+  for (int i = 0; i < aw.length; i++)
+    aw.data[i] = i;
+
+#pragma omp target update from(aw.data[:aw.length])
+
+#pragma omp target exit data map(delete: aw.data, aw.length, \
+				 aw.data[0:aw.length])
+
+  for (int i = 0; i < aw.length; i++)
+    assert (aw.data[i] == i);
+
+  delete[] aw.data;
+
+  foo<unsigned long> ();
+
+  return 0;
+}
+
diff --git a/libgomp/testsuite/libgomp.c++/map-comma-1.C b/libgomp/testsuite/libgomp.c++/map-comma-1.C
new file mode 100644
index 00000000000..ee03c5ac1aa
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/map-comma-1.C
@@ -0,0 +1,15 @@ 
+/* { dg-do run } */
+
+#include <cassert>
+
+int main (int argc, char *argv[])
+{
+  int a = 5, b = 7;
+#pragma omp target map((a, b))
+  {
+    a++;
+    b++;
+  }
+  assert (a == 5 && b == 8);
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c++/map-rvalue-ref-1.C b/libgomp/testsuite/libgomp.c++/map-rvalue-ref-1.C
new file mode 100644
index 00000000000..93811da4000
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/map-rvalue-ref-1.C
@@ -0,0 +1,22 @@ 
+/* { dg-do run } */
+
+#include <cassert>
+
+int foo (int &&x)
+{
+  int y;
+#pragma omp target map(x, y)
+  {
+    x++;
+    y = x;
+  }
+  return y;
+}
+
+int main (int argc, char *argv[])
+{
+  int y = 5;
+  y = foo (y + 3);
+  assert (y == 9);
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c++/struct-ref-1.C b/libgomp/testsuite/libgomp.c++/struct-ref-1.C
new file mode 100644
index 00000000000..d3874650017
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/struct-ref-1.C
@@ -0,0 +1,97 @@ 
+// { dg-do run }
+// { dg-options "-fopenmp" }
+
+#include <cassert>
+
+struct S
+{
+  int x[10];
+};
+
+void
+foo (S *s, int x)
+{
+  S *&r = s;
+  for (int i = 0; i < x; i++)
+    s[0].x[i] = s[1].x[i] = 0;
+  #pragma omp target map (s, x)
+    ;
+  #pragma omp target map (s[0], x)
+  for (int i = 0; i < x; i++)
+    s[0].x[i] = i;
+  #pragma omp target map (s[1], x)
+  for (int i = 0; i < x; i++)
+    s[1].x[i] = i * 2;
+  for (int i = 0; i < x; i++)
+    {
+      assert (s[0].x[i] == i);
+      assert (s[1].x[i] == i * 2);
+      s[0].x[i] = 0;
+      s[1].x[i] = 0;
+    }
+  #pragma omp target map (r, x)
+    ;
+  #pragma omp target map (r[0], x)
+  for (int i = 0; i < x; i++)
+    r[0].x[i] = i;
+  #pragma omp target map (r[1], x)
+  for (int i = 0; i < x; i++)
+    r[1].x[i] = i * 2;
+  for (int i = 0; i < x; i++)
+    {
+      assert (r[0].x[i] == i);
+      assert (r[1].x[i] == i * 2);
+    }
+}
+
+template <int N>
+struct T
+{
+  int x[N];
+};
+
+template <int N>
+void
+bar (T<N> *t, int x)
+{
+  T<N> *&r = t;
+  for (int i = 0; i < x; i++)
+    t[0].x[i] = t[1].x[i] = 0;
+  #pragma omp target map (t, x)
+    ;
+  #pragma omp target map (t[0], x)
+  for (int i = 0; i < x; i++)
+    t[0].x[i] = i;
+  #pragma omp target map (t[1], x)
+  for (int i = 0; i < x; i++)
+    t[1].x[i] = i * 2;
+  for (int i = 0; i < x; i++)
+    {
+      assert (t[0].x[i] == i);
+      assert (t[1].x[i] == i * 2);
+      t[0].x[i] = 0;
+      t[1].x[i] = 0;
+    }
+  #pragma omp target map (r, x)
+    ;
+  #pragma omp target map (r[0], x)
+  for (int i = 0; i < x; i++)
+    r[0].x[i] = i;
+  #pragma omp target map (r[1], x)
+  for (int i = 0; i < x; i++)
+    r[1].x[i] = i * 2;
+  for (int i = 0; i < x; i++)
+    {
+      assert (r[0].x[i] == i);
+      assert (r[1].x[i] == i * 2);
+    }
+}
+
+int main (int argc, char *argv[])
+{
+  S s[2];
+  foo (s, 10);
+  T<10> t[2];
+  bar (t, 10);
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c-c++-common/array-field-1.c b/libgomp/testsuite/libgomp.c-c++-common/array-field-1.c
new file mode 100644
index 00000000000..6dd8b5c48e1
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/array-field-1.c
@@ -0,0 +1,35 @@ 
+/* { dg-do run } */
+
+#include <stdlib.h>
+#include <string.h>
+#include <assert.h>
+
+#define N 16
+
+struct Z {
+  int *ptr;
+  int arr[N];
+  int c;
+};
+
+int main (int argc, char *argv[])
+{
+  struct Z *myz;
+  myz = (struct Z *) calloc (1, sizeof *myz);
+
+#pragma omp target map(tofrom:myz->arr[0:N], myz->c)
+  {
+    for (int i = 0; i < N; i++)
+      myz->arr[i]++;
+    myz->c++;
+  }
+
+  for (int i = 0; i < N; i++)
+    assert (myz->arr[i] == 1);
+  assert (myz->c == 1);
+
+  free (myz);
+
+  return 0;
+}
+
diff --git a/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-1.c b/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-1.c
new file mode 100644
index 00000000000..11215b1df7a
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-1.c
@@ -0,0 +1,65 @@ 
+/* { dg-do run } */
+
+#include <stdlib.h>
+#include <string.h>
+#include <assert.h>
+
+#define N 16
+
+/* NOTE: This test is the same as array-of-struct-2.c, except the fields of
+   this struct are in a different order.  */
+
+struct Z {
+  int arr[N];
+  int *ptr;
+  int c;
+};
+
+void
+foo (struct Z *zarr, int len)
+{
+#pragma omp target map(to:zarr, zarr[5].ptr) map(tofrom:zarr[5].ptr[0:len])
+  {
+    for (int i = 0; i < len; i++)
+      zarr[5].ptr[i]++;
+  }
+
+#pragma omp target map(to:zarr) map(tofrom:zarr[4].arr[0:len])
+  {
+    for (int i = 0; i < len; i++)
+      zarr[4].arr[i]++;
+  }
+
+#pragma omp target map (to:zarr[3].ptr) map(tofrom:zarr[3].ptr[0:len])
+  {
+    for (int i = 0; i < len; i++)
+      zarr[3].ptr[i]++;
+  }
+
+#pragma omp target map(tofrom:zarr[2].arr[0:len])
+  {
+    for (int i = 0; i < len; i++)
+      zarr[2].arr[i]++;
+  }
+}
+
+int main (int argc, char *argv[])
+{
+  struct Z zs[10];
+  memset (zs, 0, sizeof zs);
+
+  for (int i = 0; i < 10; i++)
+    zs[i].ptr = (int *) calloc (N, sizeof (int));
+
+  foo (zs, N);
+
+  for (int i = 0; i < N; i++)
+    {
+      assert (zs[2].arr[i] == 1);
+      assert (zs[4].arr[i] == 1);
+      assert (zs[3].ptr[i] == 1);
+      assert (zs[5].ptr[i] == 1);
+    }
+
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-2.c b/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-2.c
new file mode 100644
index 00000000000..d5d74b8c07d
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-2.c
@@ -0,0 +1,65 @@ 
+/* { dg-do run } */
+
+#include <stdlib.h>
+#include <string.h>
+#include <assert.h>
+
+#define N 16
+
+/* NOTE: This test is the same as array-of-struct-1.c, except the fields of
+   this struct are in a different order.  */
+
+struct Z {
+  int *ptr;
+  int arr[N];
+  int c;
+};
+
+void
+foo (struct Z *zarr, int len)
+{
+#pragma omp target map(to:zarr, zarr[5].ptr) map(tofrom:zarr[5].ptr[0:len])
+  {
+    for (int i = 0; i < len; i++)
+      zarr[5].ptr[i]++;
+  }
+
+#pragma omp target map(to:zarr) map(tofrom:zarr[4].arr[0:len])
+  {
+    for (int i = 0; i < len; i++)
+      zarr[4].arr[i]++;
+  }
+
+#pragma omp target map (to:zarr[3].ptr) map(tofrom:zarr[3].ptr[0:len])
+  {
+    for (int i = 0; i < len; i++)
+      zarr[3].ptr[i]++;
+  }
+
+#pragma omp target map(tofrom:zarr[2].arr[0:len])
+  {
+    for (int i = 0; i < len; i++)
+      zarr[2].arr[i]++;
+  }
+}
+
+int main (int argc, char *argv[])
+{
+  struct Z zs[10];
+  memset (zs, 0, sizeof zs);
+
+  for (int i = 0; i < 10; i++)
+    zs[i].ptr = (int *) calloc (N, sizeof (int));
+
+  foo (zs, N);
+
+  for (int i = 0; i < N; i++)
+    {
+      assert (zs[2].arr[i] == 1);
+      assert (zs[4].arr[i] == 1);
+      assert (zs[3].ptr[i] == 1);
+      assert (zs[5].ptr[i] == 1);
+    }
+
+  return 0;
+}