openmp: Add C/C++ support for "omp tile"

This commit adds the C and C++ front end support for the "omp tile"
directive.  The middle end support for the transformation is
implemented in a previous commit.

gcc/c-family/ChangeLog:

	* c-omp.cc (c_omp_directives): Add PRAGMA_OMP_TILE.
	* c-pragma.cc (omp_pragmas_simd): Likewise.
	* c-pragma.h (enum pragma_kind): Add PRAGMA_OMP_TILE.
	(enum pragma_omp_clause): Add PRAGMA_OMP_CLAUSE_TILE

gcc/c/ChangeLog:

	* c-parser.cc (c_parser_nested_omp_unroll_clauses): Rename and
	generalize ...
	(c_parser_omp_nested_loop_transform_clauses): ... to this.
	(c_parser_omp_for_loop): Handle "omp tile" parsing in loop nests.
	(c_parser_omp_tile_sizes): Parse single "sizes" clause.
	(c_parser_omp_loop_transform_clause): New function.
	(c_parser_omp_tile): New function for parsing "omp tile"
	(c_parser_omp_unroll): Adjust to renaming.
	(c_parser_omp_construct): Handle PRAGMA_OMP_TILE.

gcc/cp/ChangeLog:

	* parser.cc (cp_parser_omp_clause_unroll_partial): Adjust.
	(cp_parser_nested_omp_unroll_clauses): Rename ...
	(cp_parser_omp_nested_loop_transform_clauses): ... to this.
	(cp_parser_omp_for_loop): Handle "omp tile" parsing in loop nests.
	(cp_parser_omp_tile_sizes): New function, parses single "sizes" clause
	(cp_parser_omp_tile): New function for parsing "omp tile".
	(cp_parser_omp_loop_transform_clause): New  function.
	(cp_parser_omp_unroll): Adjust to renaming.
	(cp_parser_omp_construct): Handle PRAGMA_OMP_TILE.
	(cp_parser_pragma): Likewise.
	* pt.cc (tsubst_omp_clauses): Handle OMP_CLAUSE_TILE.
	* semantics.cc (finish_omp_clauses): Likewise.

gcc/ChangeLog:

	* gimplify.cc (omp_for_drop_tile_clauses): New function, ...
	(gimplify_omp_for): ... used here.

libgomp/ChangeLog:

	* testsuite/libgomp.c++/loop-transforms/tile-1.C: New test.
	* testsuite/libgomp.c++/loop-transforms/tile-2.C: New test.
	* testsuite/libgomp.c++/loop-transforms/tile-3.C: New test.

gcc/testsuite/ChangeLog:

	* c-c++-common/gomp/loop-transforms/tile-1.c: New test.
	* c-c++-common/gomp/loop-transforms/tile-2.c: New test.
	* c-c++-common/gomp/loop-transforms/tile-3.c: New test.
	* c-c++-common/gomp/loop-transforms/tile-4.c: New test.
	* c-c++-common/gomp/loop-transforms/tile-5.c: New test.
	* c-c++-common/gomp/loop-transforms/tile-6.c: New test.
	* c-c++-common/gomp/loop-transforms/tile-7.c: New test.
	* c-c++-common/gomp/loop-transforms/tile-8.c: New test.
	* c-c++-common/gomp/loop-transforms/unroll-2.c: Adapt
	to changed diagnostic messages.
	* g++.dg/gomp/loop-transforms/tile-1.h: New test.
	* g++.dg/gomp/loop-transforms/tile-1a.C: New test.
	* g++.dg/gomp/loop-transforms/tile-1b.C: New test.

23 files changed, 1823 insertions, 102 deletions

diff --git a/gcc/c-family/c-omp.cc b/gcc/c-family/c-omp.cc
index 9232f5af10e..875fbd2de01 100644
--- a/gcc/c-family/c-omp.cc
+++ b/gcc/c-family/c-omp.cc
@@ -4005,8 +4005,8 @@ const struct c_omp_directive c_omp_directives[] = {
     C_OMP_DIR_STANDALONE, false },
   { "taskyield", nullptr, nullptr, PRAGMA_OMP_TASKYIELD,
     C_OMP_DIR_STANDALONE, false },
-  /* { "tile", nullptr, nullptr, PRAGMA_OMP_TILE,
-    C_OMP_DIR_CONSTRUCT, false },  */
+  { "tile", nullptr, nullptr, PRAGMA_OMP_TILE,
+    C_OMP_DIR_CONSTRUCT, false },
   { "teams", nullptr, nullptr, PRAGMA_OMP_TEAMS,
     C_OMP_DIR_CONSTRUCT, true },
   { "threadprivate", nullptr, nullptr, PRAGMA_OMP_THREADPRIVATE,
diff --git a/gcc/c-family/c-pragma.cc b/gcc/c-family/c-pragma.cc
index 542b4389532..4a72645c5bd 100644
--- a/gcc/c-family/c-pragma.cc
+++ b/gcc/c-family/c-pragma.cc
@@ -1399,6 +1399,7 @@ static const struct omp_pragma_def omp_pragmas_simd[] = {
   { "target", PRAGMA_OMP_TARGET },
   { "taskloop", PRAGMA_OMP_TASKLOOP },
   { "teams", PRAGMA_OMP_TEAMS },
+  { "tile", PRAGMA_OMP_TILE },
   { "unroll", PRAGMA_OMP_UNROLL },
 };
 
diff --git a/gcc/c-family/c-pragma.h b/gcc/c-family/c-pragma.h
index 8e26feffd40..af6aaa192d4 100644
--- a/gcc/c-family/c-pragma.h
+++ b/gcc/c-family/c-pragma.h
@@ -82,6 +82,7 @@ enum pragma_kind {
   PRAGMA_OMP_TASKYIELD,
   PRAGMA_OMP_THREADPRIVATE,
   PRAGMA_OMP_TEAMS,
+  PRAGMA_OMP_TILE,
   PRAGMA_OMP_UNROLL,
   /* PRAGMA_OMP__LAST_ should be equal to the last PRAGMA_OMP_* code.  */
   PRAGMA_OMP__LAST_ = PRAGMA_OMP_UNROLL,
@@ -158,6 +159,7 @@ enum pragma_omp_clause {
   PRAGMA_OMP_CLAUSE_TASKGROUP,
   PRAGMA_OMP_CLAUSE_THREAD_LIMIT,
   PRAGMA_OMP_CLAUSE_THREADS,
+  PRAGMA_OMP_CLAUSE_TILE,
   PRAGMA_OMP_CLAUSE_TO,
   PRAGMA_OMP_CLAUSE_UNIFORM,
   PRAGMA_OMP_CLAUSE_UNTIED,
diff --git a/gcc/c/c-parser.cc b/gcc/c/c-parser.cc
index 80dc9a25e2d..1e1a08a8caf 100644
--- a/gcc/c/c-parser.cc
+++ b/gcc/c/c-parser.cc
@@ -19856,7 +19856,8 @@ c_parser_omp_scan_loop_body (c_parser *parser, bool open_brace_parsed)
 			     "expected %<}%>");
 }
 
-static bool c_parser_nested_omp_unroll_clauses (c_parser *, tree &);
+static int c_parser_omp_nested_loop_transform_clauses (c_parser *, tree &, int,
+						       const char *);
 
 /* Parse the restricted form of loop statements allowed by OpenACC and OpenMP.
    The real trick here is to determine the loop control variable early
@@ -19876,17 +19877,19 @@ c_parser_omp_for_loop (location_t loc, c_parser *parser, enum tree_code code,
   bool fail = false, open_brace_parsed = false;
   int i, collapse = 1, ordered = 0, count, nbraces = 0;
   location_t for_loc;
-  bool tiling = false;
+  bool oacc_tiling = false;
   bool inscan = false;
 
   vec<tree, va_gc> *for_block = make_tree_vector ();
 
   for (cl = clauses; cl; cl = OMP_CLAUSE_CHAIN (cl))
     if (OMP_CLAUSE_CODE (cl) == OMP_CLAUSE_COLLAPSE)
-      collapse = tree_to_shwi (OMP_CLAUSE_COLLAPSE_EXPR (cl));
+      {
+	collapse = tree_to_shwi (OMP_CLAUSE_COLLAPSE_EXPR (cl));
+      }
     else if (OMP_CLAUSE_CODE (cl) == OMP_CLAUSE_OACC_TILE)
       {
-	tiling = true;
+	oacc_tiling = true;
 	collapse = list_length (OMP_CLAUSE_OACC_TILE_LIST (cl));
       }
     else if (OMP_CLAUSE_CODE (cl) == OMP_CLAUSE_ORDERED
@@ -19909,21 +19912,31 @@ c_parser_omp_for_loop (location_t loc, c_parser *parser, enum tree_code code,
       ordered = collapse;
     }
 
-  gcc_assert (tiling || (collapse >= 1 && ordered >= 0));
+  c_parser_omp_nested_loop_transform_clauses (parser, clauses, collapse,
+					      "loop collapse");
+
+  /* Find the depth of the loop nest affected by "omp tile"
+     directives. There can be several such directives, but the tiling
+     depth of the outer ones may not be larger than the depth of the
+     innermost directive. */
+  int omp_tile_depth = 0;
+  for (tree c = clauses; c; c = TREE_CHAIN (c))
+    {
+      if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_TILE)
+	continue;
+
+      omp_tile_depth = list_length (OMP_CLAUSE_TILE_SIZES (c));
+    }
+
+  gcc_assert (oacc_tiling || (collapse >= 1 && ordered >= 0));
   count = ordered ? ordered : collapse;
+  count = MAX (count, omp_tile_depth);
 
   declv = make_tree_vec (count);
   initv = make_tree_vec (count);
   condv = make_tree_vec (count);
   incrv = make_tree_vec (count);
 
-  if (c_parser_nested_omp_unroll_clauses (parser, clauses)
-      && count > 1)
-    {
-      error_at (loc, "collapse cannot be larger than 1 on an unrolled loop");
-      return NULL;
-    }
-
   if (!c_parser_next_token_is_keyword (parser, RID_FOR))
     {
       c_parser_error (parser, "for statement expected");
@@ -23646,47 +23659,224 @@ c_parser_omp_taskloop (location_t loc, c_parser *parser,
 	( (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_PARTIAL)	\
 	  | (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_FULL) )
 
-/* Parse zero or more '#pragma omp unroll' that follow
-   another directive that requires a canonical loop nest. */
+/* OpenMP 5.1: Parse sizes list for "omp tile sizes"
+   sizes ( size-expr-list ) */
+static tree
+c_parser_omp_tile_sizes (c_parser *parser, location_t loc)
+{
+  tree sizes = NULL_TREE;
 
-static bool
-c_parser_nested_omp_unroll_clauses (c_parser *parser, tree &clauses)
+  c_token *tok = c_parser_peek_token (parser);
+  if (tok->type != CPP_NAME
+      || strcmp ("sizes", IDENTIFIER_POINTER (tok->value)))
+    {
+      c_parser_error (parser, "expected %<sizes%>");
+      return error_mark_node;
+    }
+  c_parser_consume_token (parser);
+
+  if (!c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>"))
+    return error_mark_node;
+
+  do
+    {
+      if (sizes && !c_parser_require (parser, CPP_COMMA, "expected %<,%>"))
+	return error_mark_node;
+
+      location_t expr_loc = c_parser_peek_token (parser)->location;
+      c_expr cexpr = c_parser_expr_no_commas (parser, NULL);
+      cexpr = convert_lvalue_to_rvalue (expr_loc, cexpr, false, true);
+      tree expr = cexpr.value;
+
+      if (expr == error_mark_node)
+	{
+	  c_parser_skip_until_found (parser, CPP_CLOSE_PAREN,
+				     "expected %<)%>");
+	  return error_mark_node;
+	}
+
+      expr = c_fully_fold (expr, false, NULL);
+
+      if (!INTEGRAL_TYPE_P (TREE_TYPE (expr)) || !tree_fits_shwi_p (expr)
+	  || tree_to_shwi (expr) <= 0)
+	{
+	  c_parser_error (parser, "%<tile sizes%> argument needs positive"
+				  " integral constant");
+	  expr = integer_zero_node;
+	}
+
+      sizes = tree_cons (NULL_TREE, expr, sizes);
+    }
+  while (c_parser_next_token_is_not (parser, CPP_CLOSE_PAREN));
+  c_parser_consume_token (parser);
+
+  gcc_assert (sizes);
+  tree c  = build_omp_clause (loc, OMP_CLAUSE_TILE);
+  OMP_CLAUSE_TILE_SIZES (c) = sizes;
+
+  return c;
+}
+
+/* Parse a single OpenMP loop transformation directive and return the
+   clause that is used internally to represent the directive. */
+
+static tree
+c_parser_omp_loop_transform_clause (c_parser *parser)
 {
-  static const char *p_name = "#pragma omp unroll";
-  c_token *tok;
-  bool found_unroll = false;
-  while (c_parser_next_token_is (parser, CPP_PRAGMA)
-	 && (tok = c_parser_peek_token (parser),
-	     tok->pragma_kind == PRAGMA_OMP_UNROLL))
+  c_token *tok = c_parser_peek_token (parser);
+  if (tok->type != CPP_PRAGMA)
+    return NULL_TREE;
+
+  tree c;
+  switch (tok->pragma_kind)
     {
+    case PRAGMA_OMP_UNROLL:
       c_parser_consume_pragma (parser);
-      tree c = c_parser_omp_all_clauses (parser, OMP_UNROLL_CLAUSE_MASK,
-					 p_name, true);
-      if (c)
+      c = c_parser_omp_all_clauses (parser, OMP_UNROLL_CLAUSE_MASK,
+				    "#pragma omp unroll", false, true);
+      if (!c)
 	{
-	  gcc_assert (!TREE_CHAIN (c));
-	  found_unroll = true;
-	  if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_UNROLL_FULL)
-	    {
-	      error_at (tok->location, "%<full%> clause is invalid here; "
-			"turns loop into non-loop");
-	      continue;
-	    }
+	  if (c_parser_next_token_is (parser, CPP_PRAGMA_EOL))
+	    c = build_omp_clause (tok->location, OMP_CLAUSE_UNROLL_NONE);
+	  else
+	    c = error_mark_node;
 	}
-      else
+      c_parser_skip_to_pragma_eol (parser);
+      break;
+
+    case PRAGMA_OMP_TILE:
+      c_parser_consume_pragma (parser);
+      c = c_parser_omp_tile_sizes (parser, tok->location);
+      c_parser_skip_to_pragma_eol (parser);
+      break;
+
+    default:
+      c = NULL_TREE;
+      break;
+    }
+
+  gcc_assert (!c || !TREE_CHAIN (c));
+  return c;
+}
+
+/* Parse zero or more OpenMP loop transformation directives that
+   follow another directive that requires a canonical loop nest and
+   append all to CLAUSES.  Return the nesting depth
+   of the transformed loop nest.
+
+   REQUIRED_DEPTH is the nesting depth of the loop nest required by
+   the preceding directive.  OUTER_DESCR is a description of the
+   language construct that requires the loop nest depth (e.g. "loop
+   collpase", "outer transformation") that is used for error
+   messages. */
+
+static int
+c_parser_omp_nested_loop_transform_clauses (c_parser *parser, tree &clauses,
+					    int required_depth,
+					    const char *outer_descr)
+{
+  tree c = NULL_TREE;
+  tree last_c = tree_last (clauses);
+
+  /* The depth of the loop nest, counting from LEVEL, after the
+     transformations. That is, the nesting depth left by the outermost
+     transformation which is the first to be parsed, but the last to be
+     executed. */
+  int transformed_depth = 0;
+
+  /* The minimum nesting depth required by the last parsed transformation. */
+  int last_depth = required_depth;
+  while ((c = c_parser_omp_loop_transform_clause (parser)))
+    {
+      /* The nesting depth left after the current transformation */
+      int depth = 1;
+      if (TREE_CODE (c) == ERROR_MARK)
+	goto error;
+
+      gcc_assert (!TREE_CHAIN (c));
+      switch (OMP_CLAUSE_CODE (c))
 	{
-	  error_at (tok->location, "%<#pragma omp unroll%> without "
-				   "%<partial%> clause is invalid here; "
-				   "turns loop into non-loop");
-	  continue;
+	case OMP_CLAUSE_UNROLL_FULL:
+	  error_at (OMP_CLAUSE_LOCATION (c),
+		    "%<full%> clause is invalid here; "
+		    "turns loop into non-loop");
+	  goto error;
+	case OMP_CLAUSE_UNROLL_NONE:
+	  error_at (OMP_CLAUSE_LOCATION (c),
+		    "%<#pragma omp unroll%> without "
+		    "%<partial%> clause is invalid here; "
+		    "turns loop into non-loop");
+	  goto error;
+	case OMP_CLAUSE_UNROLL_PARTIAL:
+	  depth = 1;
+	  break;
+	case OMP_CLAUSE_TILE:
+	  depth = list_length (OMP_CLAUSE_TILE_SIZES (c));
+	  break;
+	default:
+	  gcc_unreachable ();
+	}
+
+      if (depth < last_depth)
+	{
+	  bool is_outermost_clause = !transformed_depth;
+	  error_at (OMP_CLAUSE_LOCATION (c),
+		    "nesting depth left after this transformation too low "
+		    "for %s",
+		    is_outermost_clause ? outer_descr
+					: "outer transformation");
+	  goto error;
 	}
 
-      clauses = chainon (clauses, c);
+      last_depth = depth;
+
+      if (!transformed_depth)
+	transformed_depth = last_depth;
+
+      if (!clauses)
+	clauses = c;
+      else if (last_c)
+	TREE_CHAIN (last_c) = c;
+
+      last_c = c;
     }
 
-  return found_unroll;
+  return transformed_depth;
+
+error:
+  while (c_parser_omp_loop_transform_clause (parser))
+    ;
+  clauses = NULL_TREE;
+  return -1;
 }
 
+/* OpenMP 5.1:
+   tile sizes ( size-expr-list ) */
+
+static tree
+c_parser_omp_tile (location_t loc, c_parser *parser, bool *if_p)
+{
+  tree block;
+  tree ret = error_mark_node;
+
+  tree clauses = c_parser_omp_tile_sizes (parser, loc);
+  c_parser_skip_to_pragma_eol (parser);
+
+  if (!clauses || clauses == error_mark_node)
+    return error_mark_node;
+
+  int required_depth = list_length (OMP_CLAUSE_TILE_SIZES (clauses));
+  c_parser_omp_nested_loop_transform_clauses (parser, clauses, required_depth,
+					      "outer transformation");
+
+  block = c_begin_compound_stmt (true);
+  ret = c_parser_omp_for_loop (loc, parser, OMP_LOOP_TRANS, clauses, NULL, if_p);
+  block = c_end_compound_stmt (loc, block, true);
+  add_stmt (block);
+
+  return ret;
+ }
+
 static tree
 c_parser_omp_unroll (location_t loc, c_parser *parser, bool *if_p)
 {
@@ -23695,7 +23885,9 @@ c_parser_omp_unroll (location_t loc, c_parser *parser, bool *if_p)
   omp_clause_mask mask = OMP_UNROLL_CLAUSE_MASK;
 
   tree clauses = c_parser_omp_all_clauses (parser, mask, p_name, false);
-  c_parser_nested_omp_unroll_clauses (parser, clauses);
+  int required_depth = 1;
+  c_parser_omp_nested_loop_transform_clauses (parser, clauses, required_depth,
+					      "outer transformation");
 
   if (!clauses)
     {
@@ -24558,6 +24750,9 @@ c_parser_omp_construct (c_parser *parser, bool *if_p)
     case PRAGMA_OMP_ASSUME:
       c_parser_omp_assume (parser, if_p);
       return;
+    case PRAGMA_OMP_TILE:
+      stmt = c_parser_omp_tile (loc, parser, if_p);
+      break;
     case PRAGMA_OMP_UNROLL:
       stmt = c_parser_omp_unroll (loc, parser, if_p);
       break;
diff --git a/gcc/cp/parser.cc b/gcc/cp/parser.cc
index 1c4450b337a..230aec51886 100644
--- a/gcc/cp/parser.cc
+++ b/gcc/cp/parser.cc
@@ -43005,7 +43005,8 @@ cp_parser_omp_scan_loop_body (cp_parser *parser)
   braces.require_close (parser);
 }
 
-static bool cp_parser_nested_omp_unroll_clauses (cp_parser *, tree &);
+static int cp_parser_omp_nested_loop_transform_clauses (cp_parser *, tree &,
+							int, const char *);
 
 /* Parse the restricted form of the for statement allowed by OpenMP.  */
 
@@ -43017,20 +43018,20 @@ cp_parser_omp_for_loop (cp_parser *parser, enum tree_code code, tree clauses,
   tree orig_decl;
   tree real_decl, initv, condv, incrv, declv, orig_declv;
   tree this_pre_body, cl, ordered_cl = NULL_TREE;
-  location_t loc_first;
   bool collapse_err = false;
   int i, collapse = 1, ordered = 0, count, nbraces = 0;
   releasing_vec for_block;
   auto_vec<tree, 4> orig_inits;
-  bool tiling = false;
+  bool oacc_tiling = false;
   bool inscan = false;
+  location_t loc_first = cp_lexer_peek_token (parser->lexer)->location;
 
   for (cl = clauses; cl; cl = OMP_CLAUSE_CHAIN (cl))
     if (OMP_CLAUSE_CODE (cl) == OMP_CLAUSE_COLLAPSE)
       collapse = tree_to_shwi (OMP_CLAUSE_COLLAPSE_EXPR (cl));
     else if (OMP_CLAUSE_CODE (cl) == OMP_CLAUSE_OACC_TILE)
       {
-	tiling = true;
+	oacc_tiling = true;
 	collapse = list_length (OMP_CLAUSE_OACC_TILE_LIST (cl));
       }
     else if (OMP_CLAUSE_CODE (cl) == OMP_CLAUSE_ORDERED
@@ -43053,26 +43054,33 @@ cp_parser_omp_for_loop (cp_parser *parser, enum tree_code code, tree clauses,
       ordered = collapse;
     }
 
-  gcc_assert (tiling || (collapse >= 1 && ordered >= 0));
+
+  gcc_assert (oacc_tiling || (collapse >= 1 && ordered >= 0));
   count = ordered ? ordered : collapse;
 
+  cp_parser_omp_nested_loop_transform_clauses (parser, clauses, count,
+					       "loop collapse");
+
+  /* Find the depth of the loop nest affected by "omp tile"
+     directives. There can be several such directives, but the tiling
+     depth of the outer ones may not be larger than the depth of the
+     innermost directive. */
+  int omp_tile_depth = 0;
+  for (tree c = clauses; c; c = TREE_CHAIN (c))
+    {
+      if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_TILE)
+	continue;
+
+      omp_tile_depth = list_length (OMP_CLAUSE_TILE_SIZES (c));
+    }
+  count = MAX (count, omp_tile_depth);
+
   declv = make_tree_vec (count);
   initv = make_tree_vec (count);
   condv = make_tree_vec (count);
   incrv = make_tree_vec (count);
   orig_declv = NULL_TREE;
 
-  loc_first = cp_lexer_peek_token (parser->lexer)->location;
-
-  if (cp_parser_nested_omp_unroll_clauses (parser, clauses)
-      && count > 1)
-    {
-      error_at (loc_first,
-		"collapse cannot be larger than 1 on an unrolled loop");
-      return NULL;
-    }
-
-
   for (i = 0; i < count; i++)
     {
       int bracecount = 0;
@@ -45137,51 +45145,225 @@ cp_parser_omp_target (cp_parser *parser, cp_token *pragma_tok,
   return true;
 }
 
+
+/* OpenMP 5.1: Parse sizes list for "omp tile sizes"
+   sizes ( size-expr-list ) */
+static tree
+cp_parser_omp_tile_sizes (cp_parser *parser, location_t loc)
+{
+  tree sizes = NULL_TREE;
+  cp_lexer *lexer = parser->lexer;
+
+  cp_token *tok = cp_lexer_peek_token (lexer);
+  if (tok->type != CPP_NAME
+      || strcmp ("sizes", IDENTIFIER_POINTER (tok->u.value)))
+    {
+      cp_parser_error (parser, "expected %<sizes%>");
+      return error_mark_node;
+    }
+  cp_lexer_consume_token (lexer);
+
+  if (!cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN))
+    return error_mark_node;
+
+  do
+    {
+      if (sizes && !cp_parser_require (parser, CPP_COMMA, RT_COMMA))
+	return error_mark_node;
+
+      tree expr = cp_parser_constant_expression (parser);
+      if (expr == error_mark_node)
+	{
+	  cp_parser_skip_to_closing_parenthesis (parser,
+						 /*recovering=*/true,
+						 /*or_comma=*/false,
+						 /*consume_paren=*/
+						 true);
+	  return error_mark_node;
+	}
+
+      sizes = tree_cons (NULL_TREE, expr, sizes);
+    }
+  while (cp_lexer_next_token_is_not (lexer, CPP_CLOSE_PAREN));
+  cp_lexer_consume_token (lexer);
+
+  gcc_assert (sizes);
+  tree c  = build_omp_clause (loc, OMP_CLAUSE_TILE);
+  OMP_CLAUSE_TILE_SIZES (c) = sizes;
+
+  return c;
+}
+
+/* OpenMP 5.1:
+   tile sizes ( size-expr-list ) */
+
+static tree
+cp_parser_omp_tile (cp_parser *parser, cp_token *tok, bool *if_p)
+{
+  tree block;
+  tree ret = error_mark_node;
+
+  tree clauses = cp_parser_omp_tile_sizes (parser, tok->location);
+  cp_parser_require_pragma_eol (parser, tok);
+
+  if (!clauses || clauses == error_mark_node)
+    return error_mark_node;
+
+  int required_depth = list_length (OMP_CLAUSE_TILE_SIZES (clauses));
+  cp_parser_omp_nested_loop_transform_clauses (
+      parser, clauses, required_depth, "outer transformation");
+
+  block = begin_omp_structured_block ();
+  clauses = finish_omp_clauses (clauses, C_ORT_OMP);
+
+  ret = cp_parser_omp_for_loop (parser, OMP_LOOP_TRANS, clauses, NULL, if_p);
+  block = finish_omp_structured_block (block);
+  add_stmt (block);
+
+  return ret;
+}
+
 #define OMP_UNROLL_CLAUSE_MASK					\
 	( (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_PARTIAL)	\
 	  | (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_FULL) )
 
-/* Parse zero or more '#pragma omp unroll' that follow
-   another directive that requires a canonical loop nest. */
+/* Parse a single OpenMP loop transformation directive and return the
+   clause that is used internally to represent the directive. */
 
-static bool
-cp_parser_nested_omp_unroll_clauses (cp_parser *parser, tree &clauses)
+static tree
+cp_parser_omp_loop_transform_clause (cp_parser *parser)
 {
-  static const char *p_name = "#pragma omp unroll";
-  cp_token *tok;
-  bool unroll_found = false;
-  while (cp_lexer_next_token_is (parser->lexer, CPP_PRAGMA)
-	 && (tok = cp_lexer_peek_token (parser->lexer),
-	     cp_parser_pragma_kind (tok) == PRAGMA_OMP_UNROLL))
+  cp_lexer *lexer = parser->lexer;
+  cp_token *tok = cp_lexer_peek_token (lexer);
+  if (tok->type != CPP_PRAGMA)
+    return NULL_TREE;
+
+  tree c;
+  switch (cp_parser_pragma_kind (tok))
     {
-      cp_lexer_consume_token (parser->lexer);
-      gcc_assert (tok->type == CPP_PRAGMA);
-      parser->lexer->in_pragma = true;
-      tree c = cp_parser_omp_all_clauses (parser, OMP_UNROLL_CLAUSE_MASK,
-					  p_name, tok);
-      if (c)
-	{
-	  gcc_assert (!TREE_CHAIN (c));
-	  unroll_found = true;
-	  if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_UNROLL_FULL)
-	    {
-	      error_at (tok->location, "%<full%> clause is invalid here; "
-			"turns loop into non-loop");
-	      continue;
-	    }
+    case PRAGMA_OMP_UNROLL:
+      cp_lexer_consume_token (lexer);
+      lexer->in_pragma = true;
+      c = cp_parser_omp_all_clauses (parser, OMP_UNROLL_CLAUSE_MASK,
+				     "#pragma omp unroll", tok,
+				     false, true);
+      if (!c)
+	{
+	  if (cp_lexer_next_token_is (lexer, CPP_PRAGMA_EOL))
+	    c = build_omp_clause (tok->location, OMP_CLAUSE_UNROLL_NONE);
+	  else
+	    c = error_mark_node;
+	}
+      cp_parser_skip_to_pragma_eol (parser, tok);
+      break;
+
+    case PRAGMA_OMP_TILE:
+      cp_lexer_consume_token (lexer);
+      lexer->in_pragma = true;
+      c = cp_parser_omp_tile_sizes (parser, tok->location);
+      cp_parser_require_pragma_eol (parser, tok);
+      break;
+
+    default:
+      c = NULL_TREE;
+      break;
+    }
+
+  gcc_assert (!c || !TREE_CHAIN (c));
+  return c;
+}
+
+/* Parse zero or more OpenMP loop transformation directives that
+   follow another directive that requires a canonical loop nest and
+   append all to CLAUSES.  Return the nesting depth
+   of the transformed loop nest.
+
+   REQUIRED_DEPTH is the nesting depth of the loop nest required by
+   the preceding directive.  OUTER_DESCR is a description of the
+   language construct that requires the loop nest depth (e.g. "loop
+   collpase", "outer transformation") that is used for error
+   messages. */
 
-	  c = finish_omp_clauses (c, C_ORT_OMP);
+static int
+cp_parser_omp_nested_loop_transform_clauses (cp_parser *parser, tree &clauses,
+					     int required_depth,
+					     const char *outer_descr)
+{
+  tree c = NULL_TREE;
+  tree last_c = tree_last (clauses);
+
+  /* The depth of the loop nest after the transformations. That is,
+     the nesting depth left by the outermost transformation which is
+     the first to be parsed, but the last to be executed. */
+  int transformed_depth = 0;
+
+  /* The minimum nesting depth required by the last parsed transformation. */
+  int last_depth = required_depth;
+
+  while ((c = cp_parser_omp_loop_transform_clause (parser)))
+    {
+      /* The nesting depth left after the current transformation */
+      int depth = 1;
+      if (TREE_CODE (c) == ERROR_MARK)
+	goto error;
+
+      gcc_assert (!TREE_CHAIN (c));
+      switch (OMP_CLAUSE_CODE (c))
+	{
+	case OMP_CLAUSE_UNROLL_FULL:
+	  error_at (OMP_CLAUSE_LOCATION (c),
+		    "%<full%> clause is invalid here; "
+		    "turns loop into non-loop");
+	  goto error;
+	case OMP_CLAUSE_UNROLL_NONE:
+	  error_at (OMP_CLAUSE_LOCATION (c),
+		    "%<#pragma omp unroll%> without "
+		    "%<partial%> clause is invalid here; "
+		    "turns loop into non-loop");
+	  goto error;
+	case OMP_CLAUSE_UNROLL_PARTIAL:
+	  depth = 1;
+	  break;
+	case OMP_CLAUSE_TILE:
+	  depth = list_length (OMP_CLAUSE_TILE_SIZES (c));
+	  break;
+	default:
+	  gcc_unreachable ();
 	}
-      else
+
+      if (depth < last_depth)
 	{
-	  error_at (tok->location, "%<#pragma omp unroll%> without "
-				   "%<partial%> clause is invalid here; "
-				   "turns loop into non-loop");
-	  continue;
+	  bool is_outermost_clause = !transformed_depth;
+	  error_at (OMP_CLAUSE_LOCATION (c),
+		    "nesting depth left after this transformation too low "
+		    "for %s",
+		    is_outermost_clause ? outer_descr
+					: "outer transformation");
+	  goto error;
 	}
-      clauses = chainon (clauses, c);
+
+      last_depth = depth;
+
+      if (!transformed_depth)
+	transformed_depth = last_depth;
+
+      c = finish_omp_clauses (c, C_ORT_OMP);
+
+      if (!clauses)
+	clauses = c;
+      else if (last_c)
+	TREE_CHAIN (last_c) = c;
+
+      last_c = c;
     }
-  return unroll_found;
+
+  return transformed_depth;
+
+error:
+  while (cp_parser_omp_loop_transform_clause (parser))
+    ;
+  clauses = NULL_TREE;
+  return -1;
 }
 
 static tree
@@ -45191,7 +45373,7 @@ cp_parser_omp_unroll (cp_parser *parser, cp_token *tok, bool *if_p)
   static const char *p_name = "#pragma omp unroll";
   omp_clause_mask mask = OMP_UNROLL_CLAUSE_MASK;
 
-  tree clauses = cp_parser_omp_all_clauses (parser, mask, p_name, tok, false);
+  tree clauses = cp_parser_omp_all_clauses (parser, mask, p_name, tok, true);
 
   if (!clauses)
     {
@@ -45200,7 +45382,9 @@ cp_parser_omp_unroll (cp_parser *parser, cp_token *tok, bool *if_p)
       clauses = c;
     }
 
-  cp_parser_nested_omp_unroll_clauses (parser, clauses);
+  int required_depth = 1;
+  cp_parser_omp_nested_loop_transform_clauses (
+      parser, clauses, required_depth, "outer transformation");
 
   block = begin_omp_structured_block ();
   ret = cp_parser_omp_for_loop (parser, OMP_LOOP_TRANS, clauses, NULL, if_p);
@@ -48728,6 +48912,9 @@ cp_parser_omp_construct (cp_parser *parser, cp_token *pragma_tok, bool *if_p)
     case PRAGMA_OMP_ASSUME:
       cp_parser_omp_assume (parser, pragma_tok, if_p);
       return;
+    case PRAGMA_OMP_TILE:
+      stmt = cp_parser_omp_tile (parser, pragma_tok, if_p);
+      break;
     case PRAGMA_OMP_UNROLL:
       stmt = cp_parser_omp_unroll (parser, pragma_tok, if_p);
       break;
@@ -49361,6 +49548,7 @@ cp_parser_pragma (cp_parser *parser, enum pragma_context context, bool *if_p)
       cp_parser_omp_construct (parser, pragma_tok, if_p);
       pop_omp_privatization_clauses (stmt);
       return true;
+    case PRAGMA_OMP_TILE:
     case PRAGMA_OMP_UNROLL:
       if (context != pragma_stmt && context != pragma_compound)
 	goto bad_stmt;
diff --git a/gcc/cp/pt.cc b/gcc/cp/pt.cc
index dcd3f6117d2..5dfbcdb154c 100644
--- a/gcc/cp/pt.cc
+++ b/gcc/cp/pt.cc
@@ -17833,6 +17833,7 @@ tsubst_omp_clauses (tree clauses, enum c_omp_region_type ort,
 	case OMP_CLAUSE_WAIT:
 	case OMP_CLAUSE_DETACH:
 	case OMP_CLAUSE_UNROLL_PARTIAL:
+	case OMP_CLAUSE_TILE:
 	  OMP_CLAUSE_OPERAND (nc, 0)
 	    = tsubst_expr (OMP_CLAUSE_OPERAND (oc, 0), args, complain,
 			   in_decl, /*integral_constant_expression_p=*/false);
diff --git a/gcc/cp/semantics.cc b/gcc/cp/semantics.cc
index b08f663c39f..0a568d3a8bc 100644
--- a/gcc/cp/semantics.cc
+++ b/gcc/cp/semantics.cc
@@ -8880,6 +8880,46 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 	    }
 	  break;
 
+	case OMP_CLAUSE_TILE:
+	  for (tree list = OMP_CLAUSE_TILE_SIZES (c); !remove && list;
+	       list = TREE_CHAIN (list))
+	    {
+	      t = TREE_VALUE (list);
+
+	      if (t == error_mark_node)
+		remove = true;
+	      else if (!type_dependent_expression_p (t)
+		       && !INTEGRAL_TYPE_P (TREE_TYPE (t)))
+		{
+		  error_at (OMP_CLAUSE_LOCATION (c),
+			    "%<tile sizes%> argument needs integral type");
+		  remove = true;
+		}
+	      else
+		{
+		  t = mark_rvalue_use (t);
+		  if (!processing_template_decl)
+		    {
+		      t = maybe_constant_value (t);
+		      int n;
+		      if (!tree_fits_shwi_p (t)
+			  || !INTEGRAL_TYPE_P (TREE_TYPE (t))
+			  || (n = tree_to_shwi (t)) <= 0 || (int)n != n)
+			{
+			  error_at (OMP_CLAUSE_LOCATION (c),
+				    "%<tile sizes%> argument needs positive "
+				    "integral constant");
+			  remove = true;
+			}
+		      t = fold_build_cleanup_point_expr (TREE_TYPE (t), t);
+		    }
+		}
+
+	      /* Update list item.  */
+	      TREE_VALUE (list) = t;
+	    }
+	  break;
+
 	case OMP_CLAUSE_ORDERED:
 	  ordered_seen = true;
 	  break;
diff --git a/gcc/gimplify.cc b/gcc/gimplify.cc
index 65df4f0d1f8..6a3bd689b9e 100644
--- a/gcc/gimplify.cc
+++ b/gcc/gimplify.cc
@@ -13227,6 +13227,28 @@ find_standalone_omp_ordered (tree *tp, int *walk_subtrees, void *)
   return NULL_TREE;
 }
 
+static void omp_for_drop_tile_clauses (tree for_stmt)
+{
+  /* Drop erroneous loop transformation clauses to avoid follow up errors
+     in pass-omp_transform_loops. */
+  tree last_c = NULL_TREE;
+  for (tree c = OMP_FOR_CLAUSES (for_stmt); c;
+       c = OMP_CLAUSE_CHAIN (c))
+    {
+
+      if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_TILE)
+	continue;
+
+      if (last_c)
+	TREE_CHAIN (last_c) = TREE_CHAIN (c);
+      else
+	OMP_FOR_CLAUSES (for_stmt) = TREE_CHAIN (c);
+
+      error_at (OMP_CLAUSE_LOCATION (c),
+		"'tile' loop transformation may not appear on "
+		"non-rectangular for");
+    }
+}
 
 /* Gimplify the gross structure of an OMP_FOR statement.  */
 
@@ -13419,6 +13441,8 @@ gimplify_omp_for (tree *expr_p, gimple_seq *pre_p)
     case OMP_FOR:
       if (OMP_FOR_NON_RECTANGULAR (inner_for_stmt ? inner_for_stmt : for_stmt))
 	{
+	  omp_for_drop_tile_clauses (for_stmt);
+
 	  if (omp_find_clause (OMP_FOR_CLAUSES (for_stmt),
 			       OMP_CLAUSE_SCHEDULE))
 	    error_at (EXPR_LOCATION (for_stmt),
@@ -13464,6 +13488,8 @@ gimplify_omp_for (tree *expr_p, gimple_seq *pre_p)
       ort = ORT_SIMD;
       break;
     case OMP_LOOP_TRANS:
+      if (OMP_FOR_NON_RECTANGULAR (inner_for_stmt ? inner_for_stmt : for_stmt))
+	omp_for_drop_tile_clauses (for_stmt);
       break;
     default:
       gcc_unreachable ();
@@ -14365,6 +14391,7 @@ gimplify_omp_for (tree *expr_p, gimple_seq *pre_p)
 	  case OMP_CLAUSE_UNROLL_FULL:
 	  case OMP_CLAUSE_UNROLL_NONE:
 	  case OMP_CLAUSE_UNROLL_PARTIAL:
+	  case OMP_CLAUSE_TILE:
 	    *gfor_clauses_ptr = c;
 	    gfor_clauses_ptr = &OMP_CLAUSE_CHAIN (c);
 	    break;
diff --git a/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-1.c b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-1.c
new file mode 100644
index 00000000000..8a2f2126af4
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-1.c
@@ -0,0 +1,164 @@
+extern void dummy (int);
+
+void
+test ()
+{
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(0) /* { dg-error {'tile sizes' argument needs positive integral constant} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(-1) /* { dg-error {'tile sizes' argument needs positive integral constant} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes() /* { dg-error {expected expression before} "" { target c} } */
+    /* { dg-error {expected primary-expression before} "" { target c++ } .-1 } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(,) /* { dg-error {expected expression before} "" { target c } } */
+    /* { dg-error {expected primary-expression before} "" { target c++ } .-1 } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(1,2 /* { dg-error {expected '\,' before end of line} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes /* { dg-error {expected '\(' before end of line} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(1) sizes(1) /* { dg-error {expected end of line before 'sizes'} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(1)
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(1, 2)
+    #pragma omp tile sizes(1) /* { dg-error {nesting depth left after this transformation too low for outer transformation} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    	dummy (i);
+
+    #pragma omp tile sizes(1, 2)
+    #pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    	dummy (i);
+
+    #pragma omp tile sizes(5, 6)
+    #pragma omp tile sizes(1, 2, 3)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    for (int k = 0; k < 100; ++k)
+    	dummy (i);
+
+    #pragma omp tile sizes(1)
+    #pragma omp unroll partia /* { dg-error {expected '#pragma omp' clause before 'partia'} } */
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(1)
+    #pragma omp unroll /* { dg-error {'#pragma omp unroll' without 'partial' clause is invalid here; turns loop into non-loop} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(1)
+    #pragma omp unroll full /* { dg-error {'full' clause is invalid here; turns loop into non-loop} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(1)
+    #pragma omp unroll partial
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(8,8)
+    #pragma omp unroll partial /* { dg-error {nesting depth left after this transformation too low for outer transformation} } */
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(8,8)
+    #pragma omp unroll partial /* { dg-error {nesting depth left after this transformation too low for outer transformation} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+	dummy (i);
+
+    #pragma omp tile sizes(1, 2) /* { dg-error {'tile' loop transformation may not appear on non-rectangular for} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = i; j < 100; ++j)
+	dummy (i);
+
+    #pragma omp tile sizes(1, 2) /* { dg-error {'tile' loop transformation may not appear on non-rectangular for} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = 2; j < i; ++j)
+	dummy (i);
+
+    #pragma omp tile sizes(1, 2, 3)
+    for (int i = 0; i < 100; ++i)
+      for (int j = 0; j < 100; ++j)
+        dummy (i); /* { dg-error {not enough perfectly nested loops before 'dummy'} "" { target c } } */
+    /* { dg-error {not enough for loops to collapse} "" { target c++ } .-1 } */
+    /* { dg-error {'i' was not declared in this scope} "" { target c++ } .-2 } */
+
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+      {
+	dummy (i);
+        for (int j = 0; j < 100; ++j)
+          dummy (i);
+      }
+
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+      {
+        for (int j = 0; j < 100; ++j)
+	    dummy (j);
+	dummy (i);
+      }
+
+    #pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+      {
+        dummy (i); /* { dg-error {not enough perfectly nested loops before 'dummy'} "" { target c } } */
+	/* { dg-error {not enough for loops to collapse} "" { target c++ } .-1 } */
+        for (int j = 0; j < 100; ++j)
+          dummy (j);
+      }
+
+    #pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+      {
+        for (int j = 0; j < 100; ++j)
+	  dummy (j);
+	dummy (i); /* { dg-error {collapsed loops not perfectly nested before 'dummy'} "" { target c} } */
+	/* { dg-error {collapsed loops not perfectly nested} "" { target c++ } .-1 } */
+      }
+
+    int s;
+    #pragma omp tile sizes(s) /* { dg-error {'tile sizes' argument needs positive integral constant} "" { target { ! c++98_only } } } */
+    /* { dg-error {the value of 's' is not usable in a constant expression} "" { target { c++ && { ! c++98_only } } } .-1 } */
+    /* { dg-error {'s' cannot appear in a constant-expression} "" { target c++98_only } .-2 } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp tile sizes(42.0) /* { dg-error {'tile sizes' argument needs positive integral constant} "" { target c } } */
+    /* { dg-error {'tile sizes' argument needs integral type} "" { target c++ } .-1 } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-2.c b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-2.c
new file mode 100644
index 00000000000..51d62552945
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-2.c
@@ -0,0 +1,183 @@
+extern void dummy (int);
+
+void
+test ()
+{
+    #pragma omp parallel for
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(0) /* { dg-error {'tile sizes' argument needs positive integral constant} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(-1) /* { dg-error {'tile sizes' argument needs positive integral constant} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes() /* { dg-error {expected expression before} "" { target c} } */
+    /* { dg-error {expected primary-expression before} "" { target c++ } .-1 } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(,) /* { dg-error {expected expression before} "" { target c } } */
+    /* { dg-error {expected primary-expression before} "" { target c++ } .-1 } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1,2 /* { dg-error {expected '\,' before end of line} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes /* { dg-error {expected '\(' before end of line} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1) sizes(1) /* { dg-error {expected end of line before 'sizes'} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1)
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1, 2)
+    #pragma omp tile sizes(1) /* { dg-error {nesting depth left after this transformation too low for outer transformation} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1, 2)
+    #pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(5, 6)
+    #pragma omp tile sizes(1, 2, 3)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    for (int k = 0; k < 100; ++k)
+    	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1)
+    #pragma omp unroll partia /* { dg-error {expected '#pragma omp' clause before 'partia'} } */
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1)
+    #pragma omp unroll /* { dg-error {'#pragma omp unroll' without 'partial' clause is invalid here; turns loop into non-loop} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1)
+    #pragma omp unroll full /* { dg-error {'full' clause is invalid here; turns loop into non-loop} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1)
+    #pragma omp unroll partial
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(8,8)
+    #pragma omp unroll partial /* { dg-error {nesting depth left after this transformation too low for outer transformation} } */
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(8,8)
+    #pragma omp unroll partial /* { dg-error {nesting depth left after this transformation too low for outer transformation} } */
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1, 2) /* { dg-error {'tile' loop transformation may not appear on non-rectangular for} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = i; j < 100; ++j)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1, 2) /* { dg-error {'tile' loop transformation may not appear on non-rectangular for} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = 2; j < i; ++j)
+	dummy (i);
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1, 2, 3)
+    for (int i = 0; i < 100; ++i)
+      for (int j = 0; j < 100; ++j)
+        dummy (i); /* { dg-error {not enough perfectly nested loops before 'dummy'} "" { target c } } */
+    /* { dg-error {not enough for loops to collapse} "" { target c++ } .-1 } */
+    /* { dg-error {'i' was not declared in this scope} "" { target c++ } .-2 } */
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+      {
+	dummy (i);
+        for (int j = 0; j < 100; ++j)
+          dummy (i);
+      }
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+      {
+        for (int j = 0; j < 100; ++j)
+	    dummy (j);
+	dummy (i);
+      }
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+      {
+        dummy (i); /* { dg-error {not enough perfectly nested loops before 'dummy'} "" { target c } } */
+	/* { dg-error {not enough for loops to collapse} "" { target c++ } .-1 } */
+        for (int j = 0; j < 100; ++j)
+          dummy (j);
+      }
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+      {
+        for (int j = 0; j < 100; ++j)
+	  dummy (j);
+	dummy (i); /* { dg-error {collapsed loops not perfectly nested before 'dummy'} "" { target c} } */
+	/* { dg-error {collapsed loops not perfectly nested} "" { target c++ } .-1 } */
+      }
+
+    #pragma omp parallel for
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-3.c b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-3.c
new file mode 100644
index 00000000000..7fffc72b335
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-3.c
@@ -0,0 +1,117 @@
+extern void dummy (int);
+
+void
+test ()
+{
+    #pragma omp for
+    #pragma omp tile sizes(1, 2) /* { dg-error {'tile' loop transformation may not appear on non-rectangular for} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = i; j < 100; ++j)
+	dummy (i);
+
+    #pragma omp for
+    #pragma omp tile sizes(1, 2) /* { dg-error {'tile' loop transformation may not appear on non-rectangular for} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < i; ++j)
+	dummy (i);
+
+
+#pragma omp for collapse(1)
+    #pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+	dummy (i);
+
+#pragma omp for collapse(2)
+    #pragma omp tile sizes(1) /* { dg-error {nesting depth left after this transformation too low for loop collapse} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+	dummy (i);
+
+#pragma omp for collapse(2)
+    #pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+	dummy (i);
+
+#pragma omp for collapse(3)
+    #pragma omp tile sizes(1, 2) /* { dg-error {nesting depth left after this transformation too low for loop collapse} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+	dummy (i);
+    /* { dg-error {not enough perfectly nested loops before 'dummy'} "" { target c } .-1 } */
+    /* { dg-error {not enough for loops to collapse} "" { target c++ } .-2 } */
+    /* { dg-error {'i' was not declared in this scope} "" { target c++ } .-3 } */
+
+#pragma omp for collapse(1)
+#pragma omp tile sizes(1)
+#pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+#pragma omp for collapse(2)
+#pragma omp tile sizes(1, 2)
+#pragma omp tile sizes(1) /* { dg-error {nesting depth left after this transformation too low for outer transformation} } */
+    for (int i = 0; i < 100; ++i)
+    	dummy (i); /* { dg-error {not enough perfectly nested loops before 'dummy'} "" { target c } } */
+    /* { dg-error {not enough for loops to collapse} "" { target c++ } .-1 } */
+
+#pragma omp for collapse(2)
+#pragma omp tile sizes(1, 2)
+#pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+    	dummy (i); /* { dg-error {not enough perfectly nested loops before 'dummy'} "" { target c } } */
+    /* { dg-error {not enough for loops to collapse} "" { target c++ } .-1 } */
+
+#pragma omp for collapse(2)
+#pragma omp tile sizes(5, 6)
+#pragma omp tile sizes(1, 2, 3)
+    for (int i = 0; i < 100; ++i)
+    	dummy (i); /* { dg-error {not enough perfectly nested loops before 'dummy'} "" { target c } } */
+    /* { dg-error {not enough for loops to collapse} "" { target c++ } .-1 } */
+
+
+#pragma omp for collapse(1)
+#pragma omp tile sizes(1)
+#pragma omp tile sizes(1)
+    for (int i = 0; i < 100; ++i)
+	dummy (i);
+
+#pragma omp for collapse(2)
+#pragma omp tile sizes(1, 2)
+#pragma omp tile sizes(1) /* { dg-error {nesting depth left after this transformation too low for outer transformation} } */
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    	dummy (i);
+
+#pragma omp for collapse(2)
+#pragma omp tile sizes(1, 2)
+#pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    	dummy (i);
+
+#pragma omp for collapse(2)
+#pragma omp tile sizes(5, 6)
+#pragma omp tile sizes(1, 2, 3)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    for (int k = 0; k < 100; ++k)
+    	dummy (i);
+
+#pragma omp for collapse(3)
+#pragma omp tile sizes(1, 2) /* { dg-error {nesting depth left after this transformation too low for loop collapse} } */
+#pragma omp tile sizes(1, 2)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    	dummy (i); /* { dg-error {not enough perfectly nested loops before 'dummy'} "" { target c } } */
+    /* { dg-error {not enough for loops to collapse} "" { target c++ } .-1 } */
+
+#pragma omp for collapse(3)
+#pragma omp tile sizes(5, 6) /* { dg-error {nesting depth left after this transformation too low for loop collapse} } */
+#pragma omp tile sizes(1, 2, 3)
+    for (int i = 0; i < 100; ++i)
+    for (int j = 0; j < 100; ++j)
+    for (int k = 0; k < 100; ++k)
+    	dummy (i);
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-4.c b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-4.c
new file mode 100644
index 00000000000..d46bb0cb642
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-4.c
@@ -0,0 +1,322 @@
+/* { dg-do run } */
+/* { dg-options "-O0 -fopenmp-simd" } */
+
+#include <stdio.h>
+
+#define ASSERT_EQ(var, val) if (var != val) { fprintf (stderr, "%s:%d: Unexpected value %d, expected %d\n", __FILE__, __LINE__, var, val); \
+    __builtin_abort (); }
+
+int
+test1 ()
+{
+  int iter = 0;
+  int i;
+#pragma omp tile sizes(3)
+  for (i = 0; i < 10; i=i+2)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter = iter + 2;
+	}
+
+  ASSERT_EQ (i, 10)
+  return iter;
+}
+
+int
+test2 ()
+{
+  int iter = 0;
+  int i;
+#pragma omp tile sizes(3)
+  for (i = 0; i < 10; i=i+2)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter = iter + 2;
+	}
+
+  ASSERT_EQ (i, 10)
+  return iter;
+}
+
+int
+test3 ()
+{
+  int iter = 0;
+  int i;
+#pragma omp tile sizes(8)
+  for (i = 0; i < 10; i=i+2)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter = iter + 2;
+	}
+
+  ASSERT_EQ (i, 10)
+  return iter;
+}
+
+int
+test4 ()
+{
+  int iter = 10;
+  int i;
+#pragma omp tile sizes(8)
+  for (i = 10; i > 0; i=i-2)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter = iter - 2;
+	}
+  ASSERT_EQ (i, 0)
+  return iter;
+}
+
+int
+test5 ()
+{
+  int iter = 10;
+  int i;
+#pragma omp tile sizes(71)
+  for (i = 10; i > 0; i=i-2)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter = iter - 2;
+	}
+
+  ASSERT_EQ (i, 0)
+  return iter;
+}
+
+int
+test6 ()
+{
+  int iter = 10;
+  int i;
+#pragma omp tile sizes(1)
+  for (i = 10; i > 0; i=i-2)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter = iter - 2;
+	}
+  ASSERT_EQ (i, 0)
+  return iter;
+}
+
+int
+test7 ()
+{
+  int iter = 5;
+  int i;
+#pragma omp tile sizes(2)
+  for (i = 5; i < -5; i=i-3)
+	{
+	  fprintf (stderr, "%d\n", i);
+	  __builtin_abort ();
+	  iter = iter - 3;
+	}
+
+  ASSERT_EQ (i, 5)
+
+  /* No iteration expected */
+  return iter;
+}
+
+int
+test8 ()
+{
+  int iter = 5;
+  int i;
+#pragma omp tile sizes(2)
+  for (i = 5; i > -5; i=i-3)
+	{
+	  ASSERT_EQ (i, iter)
+	  /* Expect only first iteration of the last tile to execute */
+	  if (iter != -4)
+	    iter = iter - 3;
+	}
+
+  ASSERT_EQ (i, -7)
+  return iter;
+}
+
+
+int
+test9 ()
+{
+  int iter = 5;
+  int i;
+#pragma omp tile sizes(5)
+  for (i = 5; i >= -5; i=i-4)
+	{
+	  ASSERT_EQ (i, iter)
+	  /* Expect only first iteration of the last tile to execute */
+	  if (iter != - 3)
+	    iter = iter - 4;
+	}
+
+  ASSERT_EQ (i, -7)
+  return iter;
+}
+
+int
+test10 ()
+{
+  int iter = 5;
+  int i;
+#pragma omp tile sizes(5)
+  for (i = 5; i >= -5; i--)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter--;
+	}
+
+  ASSERT_EQ (i, -6)
+  return iter;
+}
+
+int
+test11 ()
+{
+  int iter = 5;
+  int i;
+#pragma omp tile sizes(15)
+  for (i = 5; i != -5; i--)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter--;
+	}
+  ASSERT_EQ (i, -5)
+  return iter;
+}
+
+int
+test12 ()
+{
+  int iter = 0;
+  unsigned i;
+#pragma omp tile sizes(3)
+  for (i = 0; i != 5; i++)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter++;
+	}
+
+  ASSERT_EQ (i, 5)
+  return iter;
+}
+
+int
+test13 ()
+{
+  int iter = -5;
+  long long unsigned int i;
+#pragma omp tile sizes(15)
+  for (int i = -5; i < 5; i=i+3)
+	{
+	  ASSERT_EQ (i, iter)
+	  iter++;
+	}
+
+  ASSERT_EQ (i, 5)
+  return iter;
+}
+
+int
+test14 (unsigned init, int step)
+{
+  int iter = init;
+  long long unsigned int i;
+#pragma omp tile sizes(8)
+  for (i = init; i < 2*init; i=i+step)
+    iter++;
+
+  ASSERT_EQ (i, 2*init)
+  return iter;
+}
+
+int
+test15 (unsigned init, int step)
+{
+  int iter = init;
+  int i;
+#pragma omp tile sizes(8)
+  for (unsigned i = init; i > 2* init; i=i+step)
+    iter++;
+
+  return iter;
+}
+
+int
+main ()
+{
+  int last_iter;
+
+  last_iter = test1 ();
+  ASSERT_EQ (last_iter, 10);
+
+  last_iter = test2 ();
+  ASSERT_EQ (last_iter, 10);
+
+  last_iter = test3 ();
+  ASSERT_EQ (last_iter, 10);
+
+  last_iter = test4 ();
+  ASSERT_EQ (last_iter, 0);
+
+  last_iter = test5 ();
+  ASSERT_EQ (last_iter, 0);
+
+  last_iter = test6 ();
+  ASSERT_EQ (last_iter, 0);
+
+  last_iter = test7 ();
+  ASSERT_EQ (last_iter, 5);
+
+  last_iter = test8 ();
+  ASSERT_EQ (last_iter, -4);
+
+  last_iter = test9 ();
+  ASSERT_EQ (last_iter, -3);
+
+  last_iter = test10 ();
+  ASSERT_EQ (last_iter, -6);
+  return 0;
+
+  last_iter = test11 ();
+  ASSERT_EQ (last_iter, -4);
+  return 0;
+
+  last_iter = test12 ();
+  ASSERT_EQ (last_iter, 5);
+  return 0;
+
+  last_iter = test13 ();
+  ASSERT_EQ (last_iter, 4);
+  return 0;
+
+  last_iter = test14 (0, 1);
+  ASSERT_EQ (last_iter, 0);
+  return 0;
+
+  last_iter = test14 (0, -1);
+  ASSERT_EQ (last_iter, 0);
+  return 0;
+
+  last_iter = test14 (8, 2);
+  ASSERT_EQ (last_iter, 16);
+  return 0;
+
+  last_iter = test14 (5, 3);
+  ASSERT_EQ (last_iter, 9);
+  return 0;
+
+  last_iter = test15 (8, -1);
+  ASSERT_EQ (last_iter, 9);
+  return 0;
+
+  last_iter = test15 (8, -2);
+  ASSERT_EQ (last_iter, 10);
+  return 0;
+
+  last_iter = test15 (5, -3);
+  ASSERT_EQ (last_iter, 6);
+  return 0;
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-5.c b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-5.c
new file mode 100644
index 00000000000..815318ab27a
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-5.c
@@ -0,0 +1,150 @@
+/* { dg-do run } */
+/* { dg-options "-O0 -fopenmp-simd" } */
+
+#include <stdio.h>
+
+#define ASSERT_EQ(var, val) if (var != val) { fprintf (stderr, "%s:%d: Unexpected value %d\n", __FILE__, __LINE__, var); \
+    __builtin_abort (); }
+
+#define ASSERT_EQ_PTR(var, ptr) if (var != ptr) { fprintf (stderr, "%s:%d: Unexpected value %p\n", __FILE__, __LINE__, var); \
+    __builtin_abort (); }
+
+int
+test1 (int data[10])
+{
+  int iter = 0;
+  int *i;
+  #pragma omp tile sizes(5)
+  for (i = data; i < data + 10 ; i++)
+	{
+	  ASSERT_EQ (*i, data[iter]);
+	  ASSERT_EQ_PTR (i, data + iter);
+	  iter++;
+	}
+
+  ASSERT_EQ_PTR (i, data + 10)
+  return iter;
+}
+
+int
+test2 (int data[10])
+{
+  int iter = 0;
+  int *i;
+  #pragma omp tile sizes(5)
+  for (i = data; i < data + 10 ; i=i+2)
+	{
+	  ASSERT_EQ_PTR (i, data + 2 * iter);
+	  ASSERT_EQ (*i, data[2 * iter]);
+	  iter++;
+	}
+
+  ASSERT_EQ_PTR (i, data + 10)
+  return iter;
+}
+
+int
+test3 (int data[10])
+{
+  int iter = 0;
+  int *i;
+  #pragma omp tile sizes(5)
+  for (i = data; i <= data + 9 ; i=i+2)
+	{
+	  ASSERT_EQ (*i, data[2 * iter]);
+	  iter++;
+	}
+
+  ASSERT_EQ_PTR (i, data + 10)
+  return iter;
+}
+
+int
+test4 (int data[10])
+{
+  int iter = 0;
+  int *i;
+  #pragma omp tile sizes(5)
+  for (i = data; i != data + 10 ; i=i+1)
+	{
+	  ASSERT_EQ (*i, data[iter]);
+	  iter++;
+	}
+
+  ASSERT_EQ_PTR (i, data + 10)
+  return iter;
+}
+
+int
+test5 (int data[10])
+{
+  int iter = 0;
+  int *i;
+  #pragma omp tile sizes(3)
+  for (i = data + 9; i >= data ; i--)
+	{
+	  ASSERT_EQ (*i, data[9 - iter]);
+	  iter++;
+	}
+
+  ASSERT_EQ_PTR (i, data - 1)
+  return iter;
+}
+
+int
+test6 (int data[10])
+{
+  int iter = 0;
+  int *i;
+  #pragma omp tile sizes(3)
+  for (i = data + 9; i > data - 1 ; i--)
+	{
+	  ASSERT_EQ (*i, data[9 - iter]);
+	  iter++;
+	}
+
+  ASSERT_EQ_PTR (i, data - 1)
+  return iter;
+}
+
+int
+test7 (int data[10])
+{
+  int iter = 0;
+  #pragma omp tile sizes(1)
+  for (int *i = data + 9; i != data - 1 ; i--)
+	{
+	  ASSERT_EQ (*i, data[9 - iter]);
+	  iter++;
+	}
+
+  return iter;
+}
+
+int
+main ()
+{
+  int iter_count;
+  int data[10] = { 0, 1, 2, 3, 4, 5, 6, 7, 8, 9 };
+
+  iter_count = test1 (data);
+  ASSERT_EQ (iter_count, 10);
+
+  iter_count = test2 (data);
+  ASSERT_EQ (iter_count, 5);
+
+  iter_count = test3 (data);
+  ASSERT_EQ (iter_count, 5);
+
+  iter_count = test4 (data);
+  ASSERT_EQ (iter_count, 10);
+
+  iter_count = test5 (data);
+  ASSERT_EQ (iter_count, 10);
+
+  iter_count = test6 (data);
+  ASSERT_EQ (iter_count, 10);
+
+  iter_count = test7 (data);
+  ASSERT_EQ (iter_count, 10);
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-6.c b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-6.c
new file mode 100644
index 00000000000..8132128a5a8
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-6.c
@@ -0,0 +1,34 @@
+/* { dg-do run } */
+/* { dg-options "-O0 -fopenmp-simd" } */
+
+#include <stdio.h>
+
+int
+test1 ()
+{
+   int sum = 0;
+for (int k = 0; k < 10; k++)
+  {
+#pragma omp tile sizes(5,7)
+  for (int i = 0; i < 10; i++)
+  for (int j = 0; j < 10; j=j+2)
+	{
+	  sum = sum + 1;
+	}
+  }
+
+  return sum;
+}
+
+int
+main ()
+{
+  int result = test1 ();
+
+  if (result != 500)
+    {
+      fprintf (stderr, "Wrong result: %d\n", result);
+    __builtin_abort ();
+    }
+  return 0;
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-7.c b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-7.c
new file mode 100644
index 00000000000..cd25a62c5c0
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-7.c
@@ -0,0 +1,31 @@
+/* { dg-do run } */
+/* { dg-options "-O0 -fopenmp-simd" } */
+
+#include <stdio.h>
+#define ASSERT_EQ(var, val) if (var != val) { fprintf (stderr, "%s:%d: Unexpected value %d\n", __FILE__, __LINE__, var); \
+    __builtin_abort (); }
+
+#define ASSERT_EQ_PTR(var, ptr) if (var != ptr) { fprintf (stderr, "%s:%d: Unexpected value %p\n", __FILE__, __LINE__, var); \
+    __builtin_abort (); }
+
+int
+main ()
+{
+  int iter_count;
+  int data[10] = { 0, 1, 2, 3, 4, 5, 6, 7, 8, 9 };
+
+  int iter = 0;
+  int *i;
+  #pragma omp tile sizes(1)
+  for (i = data; i < data + 10; i=i+2)
+	{
+	  ASSERT_EQ_PTR (i, data + 2 * iter);
+	  ASSERT_EQ (*i, data[2 * iter]);
+	  iter++;
+	}
+
+  unsigned long real_iter_count = ((unsigned long)i - (unsigned long)data) / (sizeof (int) * 2);
+  ASSERT_EQ (real_iter_count, 5);
+
+  return 0;
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-8.c b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-8.c
new file mode 100644
index 00000000000..c26e03d7e74
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/loop-transforms/tile-8.c
@@ -0,0 +1,40 @@
+/* { dg-do run } */
+/* { dg-options "-O0 -fopenmp-simd" } */
+
+#include <stdio.h>
+
+#define ASSERT_EQ(var, val) if (var != val) { fprintf (stderr, "%s:%d: Unexpected value %d, expected %d\n", __FILE__, __LINE__, var, val); \
+    __builtin_abort (); }
+
+int
+main ()
+{
+  int iter_j = 0, iter_k = 0;
+  unsigned i, j, k;
+#pragma omp tile sizes(3,5,8)
+  for (i = 0; i < 2; i=i+2)
+  for (j = 0; j < 3; j=j+1)
+  for (k = 0; k < 5; k=k+3)
+	{
+	  /* fprintf (stderr, "i=%d j=%d k=%d\n", i, j, k);
+	   * fprintf (stderr, "iter_j=%d iter_k=%d\n", iter_j, iter_k); */
+	  ASSERT_EQ (i, 0);
+	  if (k == 0)
+	    {
+	      ASSERT_EQ (j, iter_j);
+	      iter_k = 0;
+	    }
+
+	  ASSERT_EQ (k, iter_k);
+
+	  iter_k = iter_k + 3;
+	  if (k == 3)
+	    iter_j++;
+	}
+
+  ASSERT_EQ (i, 2);
+  ASSERT_EQ (j, 3);
+  ASSERT_EQ (k, 6);
+
+  return 0;
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/loop-transforms/unroll-2.c b/gcc/testsuite/c-c++-common/gomp/loop-transforms/unroll-2.c
index 8f7c3088a2e..e4fee72c04d 100644
--- a/gcc/testsuite/c-c++-common/gomp/loop-transforms/unroll-2.c
+++ b/gcc/testsuite/c-c++-common/gomp/loop-transforms/unroll-2.c
@@ -19,7 +19,7 @@ test ()
 
 #pragma omp for
 #pragma omp unroll full /* { dg-error {'full' clause is invalid here; turns loop into non-loop} } */
-#pragma omp unroll full /* { dg-error {'full' clause is invalid here; turns loop into non-loop} } */
+#pragma omp unroll full
   for (int i = -300; i != 100; ++i)
     dummy (i);
 
@@ -45,13 +45,11 @@ test ()
   int i;
 #pragma omp for
 #pragma omp unroll( /* { dg-error {expected '#pragma omp' clause before '\(' token} } */
-  /* { dg-error {'#pragma omp unroll' without 'partial' clause is invalid here; turns loop into non-loop} "" { target *-*-* } .-1 } */
   for (int i = -300; i != 100; ++i)
     dummy (i);
 
 #pragma omp for
 #pragma omp unroll foo /* { dg-error {expected '#pragma omp' clause before 'foo'} } */
-  /* { dg-error {'#pragma omp unroll' without 'partial' clause is invalid here; turns loop into non-loop} "" { target *-*-* } .-1 } */
   for (int i = -300; i != 100; ++i)
     dummy (i);
 
@@ -67,7 +65,7 @@ test ()
 
 #pragma omp unroll partial(i)
  /* { dg-error {the value of 'i' is not usable in a constant expression} "" { target c++ } .-1 } */
- /* { dg-error {partial argument needs positive constant integer expression} "" { target c } .-2 } */
+ /* { dg-error {partial argument needs positive constant integer expression} "" { target *-*-* } .-2 } */
   for (int i = -300; i != 100; ++i)
     dummy (i);
 
@@ -78,20 +76,18 @@ test ()
 #pragma omp for
 #pragma omp unroll partial(1)
 #pragma omp unroll parti /* { dg-error {expected '#pragma omp' clause before 'parti'} } */
-  /* { dg-error {'#pragma omp unroll' without 'partial' clause is invalid here; turns loop into non-loop} "" { target *-*-* } .-1 } */
   for (int i = -300; i != 100; ++i)
     dummy (i);
 
 #pragma omp for
 #pragma omp unroll partial(1)
 #pragma omp unroll parti /* { dg-error {expected '#pragma omp' clause before 'parti'} } */
-  /* { dg-error {'#pragma omp unroll' without 'partial' clause is invalid here; turns loop into non-loop} "" { target *-*-* } .-1 } */
   for (int i = -300; i != 100; ++i)
     dummy (i);
 
 int sum = 0;
-#pragma omp parallel for reduction(+ : sum) collapse(2) /* { dg-error {collapse cannot be larger than 1 on an unrolled loop} "" { target c } } */
-#pragma omp unroll partial(1) /* { dg-error {collapse cannot be larger than 1 on an unrolled loop} "" { target c++ } } */
+#pragma omp parallel for reduction(+ : sum) collapse(2)
+#pragma omp unroll partial(1) /* { dg-error {nesting depth left after this transformation too low for loop collapse} } */
   for (int i = 3; i < 10; ++i)
     for (int j = -2; j < 7; ++j)
       sum++;
diff --git a/gcc/testsuite/g++.dg/gomp/loop-transforms/tile-1.h b/gcc/testsuite/g++.dg/gomp/loop-transforms/tile-1.h
new file mode 100644
index 00000000000..166d1d48677
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/loop-transforms/tile-1.h
@@ -0,0 +1,27 @@
+// { dg-do compile }
+// { dg-additional-options "-std=c++11" }
+
+#include <vector>
+
+extern void dummy (int);
+
+template<class T, int U, unsigned V> void
+test1_template ()
+{
+  std::vector<int> v;
+
+  for (unsigned i = 0; i < 10; i++)
+    v.push_back (i);
+
+#pragma omp for
+  for (int i : v)
+    dummy (i);
+
+#pragma omp tile sizes (U, 10, V)
+  for (T i : v)
+  for (T j : v)
+  for (T k : v)
+    dummy (i);
+}
+
+void test () { test1_template <long, 5, 3> (); };
diff --git a/gcc/testsuite/g++.dg/gomp/loop-transforms/tile-1a.C b/gcc/testsuite/g++.dg/gomp/loop-transforms/tile-1a.C
new file mode 100644
index 00000000000..1ee76da3d4a
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/loop-transforms/tile-1a.C
@@ -0,0 +1,27 @@
+// { dg-do compile }
+// { dg-additional-options "-std=c++11" }
+
+#include <vector>
+
+extern void dummy (int);
+
+template<class T, int U, unsigned V> void
+test1_template ()
+{
+  std::vector<int> v;
+
+  for (unsigned i = 0; i < 10; i++)
+    v.push_back (i);
+
+#pragma omp teams distribute parallel for num_teams(V)
+  for (int i : v)
+    dummy (i);
+
+#pragma omp tile sizes (V, U)
+  for (T i : v)
+  for (T j : v)
+  for (T k : v)
+    dummy (i);
+}
+
+void test () { test1_template <long, 5, 3> (); };
diff --git a/gcc/testsuite/g++.dg/gomp/loop-transforms/tile-1b.C b/gcc/testsuite/g++.dg/gomp/loop-transforms/tile-1b.C
new file mode 100644
index 00000000000..263c9b301c6
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/loop-transforms/tile-1b.C
@@ -0,0 +1,27 @@
+// { dg-do compile }
+// { dg-additional-options "-std=c++11" }
+
+#include <vector>
+
+extern void dummy (int);
+
+template<class T, int U, unsigned V> void
+test1_template ()
+{
+  std::vector<int> v;
+
+  for (unsigned i = 0; i < 10; i++)
+    v.push_back (i);
+
+#pragma omp for
+  for (int i : v)
+    dummy (i);
+
+#pragma omp tile sizes (U, 10, V) // { dg-error {'tile sizes' argument needs positive integral constant} }
+  for (T i : v)
+  for (T j : v)
+  for (T k : v)
+    dummy (i);
+}
+
+void test () { test1_template <long, 5, 0> (); };
diff --git a/libgomp/testsuite/libgomp.c++/loop-transforms/tile-1.C b/libgomp/testsuite/libgomp.c++/loop-transforms/tile-1.C
new file mode 100644
index 00000000000..ac7060ebc2f
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/loop-transforms/tile-1.C
@@ -0,0 +1,52 @@
+#include <string.h>
+#include <stdio.h>
+#include <math.h>
+
+void
+mult (float *matrix1, float *matrix2, float *result, unsigned dim0,
+      unsigned dim1)
+{
+  memset (result, 0, sizeof (float) * dim0 * dim1);
+#pragma omp target parallel for collapse(3)
+#pragma omp tile sizes(8, 16, 4)
+  for (unsigned i = 0; i < dim0; i++)
+    for (unsigned j = 0; j < dim1; j++)
+      for (unsigned k = 0; k < dim1; k++)
+	result[i * dim1 + j] += matrix1[i * dim1 + k] * matrix2[k * dim0 + j];
+}
+
+int
+main ()
+{
+  unsigned dim0 = 20;
+  unsigned dim1 = 20;
+
+  float *result = (float *)malloc (sizeof (float) * dim0 * dim1);
+  float *matrix1 = (float *)malloc (sizeof (float) * dim0 * dim1);
+  float *matrix2 = (float *)malloc (sizeof (float) * dim0 * dim1);
+
+  for (unsigned i = 0; i < dim0; i++)
+    for (unsigned j = 0; j < dim1; j++)
+      matrix1[i * dim1 + j] = j;
+
+  for (unsigned i = 0; i < dim1; i++)
+    for (unsigned j = 0; j < dim0; j++)
+      if (i == j)
+	matrix2[i * dim0 + j] = 1;
+      else
+	matrix2[i * dim0 + j] = 0;
+
+  mult (matrix1, matrix2, result, dim0, dim1);
+
+  for (unsigned i = 0; i < dim0; i++)
+    for (unsigned j = 0; j < dim1; j++)
+      {
+	if (matrix1[i * dim1 + j] != result[i * dim1 + j])
+	  {
+	    printf ("ERROR at %d, %d\n", i, j);
+	    __builtin_abort ();
+	  }
+      }
+
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c++/loop-transforms/tile-2.C b/libgomp/testsuite/libgomp.c++/loop-transforms/tile-2.C
new file mode 100644
index 00000000000..780421fa4c7
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/loop-transforms/tile-2.C
@@ -0,0 +1,69 @@
+// { dg-additional-options "-std=c++11" }
+// { dg-additional-options "-O0" }
+
+#include <vector>
+#include <stdio.h>
+
+constexpr unsigned fib (unsigned n)
+{
+  return n <= 2 ? 1 : fib (n-1) + fib (n-2);
+}
+
+int
+test1 ()
+{
+  std::vector<int> v;
+
+  for (unsigned i = 0; i <= 9; i++)
+    v.push_back (1);
+
+  int sum = 0;
+  for (int k = 0; k < 10; k++)
+    #pragma omp tile sizes(fib(4))
+    for (int i : v) {
+      for (int j = 8; j != -2; --j)
+	sum = sum + i;
+    }
+
+  return sum;
+}
+
+int
+test2 ()
+{
+  std::vector<int> v;
+
+  for (unsigned i = 0; i <= 10; i++)
+    v.push_back (i);
+
+  int sum = 0;
+  for (int k = 0; k < 10; k++)
+#pragma omp parallel for collapse(2) reduction(+:sum)
+#pragma omp tile sizes(fib(4), 1)
+  for (int i : v)
+    for (int j = 8; j > -2; --j)
+	sum = sum + i;
+
+  return sum;
+}
+
+int
+main ()
+{
+  int result = test1 ();
+
+  if (result != 1000)
+    {
+      fprintf (stderr, "%d: Wrong result: %d\n", __LINE__, result);
+      __builtin_abort ();
+    }
+
+  result = test2 ();
+  if (result != 5500)
+    {
+      fprintf (stderr, "%d: Wrong result: %d\n", __LINE__, result);
+    __builtin_abort ();
+    }
+
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c++/loop-transforms/tile-3.C b/libgomp/testsuite/libgomp.c++/loop-transforms/tile-3.C
new file mode 100644
index 00000000000..91ec8f5c137
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c++/loop-transforms/tile-3.C
@@ -0,0 +1,28 @@
+// { dg-additional-options "-std=c++11" }
+// { dg-additional-options "-O0" }
+
+#include <vector>
+
+int
+main ()
+{
+  std::vector<int> v;
+  std::vector<int> w;
+
+  for (unsigned i = 0; i <= 9; i++)
+    v.push_back (i);
+
+  int iter = 0;
+#pragma omp for
+#pragma omp tile sizes(5)
+  for (int i : v)
+    {
+      w.push_back (iter);
+      iter++;
+    }
+
+  for (int i = 0; i < w.size (); i++)
+    if (w[i] != i)
+      __builtin_abort ();
+  return 0;
+}