[gomp4.1] depend(sink) and depend(source) parsing for C
diff mbox

Message ID 55A008FF.10609@redhat.com
State New
Headers show

Commit Message

Aldy Hernandez July 10, 2015, 6:03 p.m. UTC
On 07/09/2015 11:53 AM, Jakub Jelinek wrote:
> Hi!
>
> On Thu, Jul 09, 2015 at 11:24:44AM -0700, Aldy Hernandez wrote:
>
> Thanks for working on it.
>
>> +	      wide_int offset = wi::neg (addend, &overflow);
>> +	      addend = wide_int_to_tree (TREE_TYPE (addend), offset);
>> +	      if (overflow)
>> +		warning_at (c_parser_peek_token (parser)->location,
>> +			    OPT_Woverflow,
>> +			    "possible overflow in %<depend(sink)%> offset");
>
> possible overflow looks weird.  Shouldn't it complain the same
> as it does if you do:
> int c = - (-2147483648);

Done.

> ?
>
>> --- a/gcc/c/c-typeck.c
>> +++ b/gcc/c/c-typeck.c
>> @@ -12489,6 +12489,11 @@ c_finish_omp_clauses (tree clauses, bool declare_simd)
>>   			  == OMP_CLAUSE_DEPEND_SOURCE);
>>   	      break;
>>   	    }
>> +	  if (OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SINK)
>> +	    {
>> +	      gcc_assert (TREE_CODE (t) == TREE_LIST);
>> +	      break;
>> +	    }
>>   	  if (TREE_CODE (t) == TREE_LIST)
>>   	    {
>>   	      if (handle_omp_array_sections (c))
>
> Won't this ICE if somebody uses depend(sink:) ? or depend(sink:.::) or
> similar garbage?  Make sure you don't create OMP_CLAUSE_DEPEND in that
> case.

I've fixed the parser to avoid creating such clause.

>
>> diff --git a/gcc/gimple-walk.c b/gcc/gimple-walk.c
>> index f0e2c67..ba79977 100644
>> --- a/gcc/gimple-walk.c
>> +++ b/gcc/gimple-walk.c
>> @@ -327,6 +327,10 @@ walk_gimple_op (gimple stmt, walk_tree_fn callback_op,
>>         }
>>         break;
>>
>> +    case GIMPLE_OMP_ORDERED:
>> +      /* Ignore clauses.  */
>> +      break;
>> +
>
> I'm not convinced you don't want to walk the clauses.

Ok, I've done so.

Note that the OMP_CLAUSE_DECL will contain a TREE_LIST whose 
TREE_PURPOSE had the variable.  I noticed that walking TREE_LIST's just 
walks the TREE_VALUE, not the TREE_PURPOSE:

     case TREE_LIST:
       WALK_SUBTREE (TREE_VALUE (*tp));
       WALK_SUBTREE_TAIL (TREE_CHAIN (*tp));
       break;


So, I changed the layout of the OMP_CLAUSE_DECL TREE_LIST to have the 
variable in the TREE_VALUE.  The TREE_PURPOSE will contain the lone 
integer, which shouldn't need to be walked.  However, if later (C++ 
iterators??) we have a TREE_PURPOSE that needs to be walked we will have 
to change the walker or the layout.

>
>> diff --git a/gcc/gimple.h b/gcc/gimple.h
>> index 6057ea0..e33fe1e 100644
>> --- a/gcc/gimple.h
>> +++ b/gcc/gimple.h
>> @@ -527,6 +527,17 @@ struct GTY((tag("GSS_OMP_CRITICAL")))
>>     tree name;
>>   };
>>
>> +/* GIMPLE_OMP_ORDERED */
>> +
>> +struct GTY((tag("GSS_OMP_ORDERED")))
>> +  gomp_ordered : public gimple_statement_omp
>> +{
>> +  /* [ WORD 1-7 ] : base class */
>> +
>> +  /* [ WORD 8 ]  */
>> +  tree clauses;
>> +};
>
> I would have expected to use
> struct GTY((tag("GSS_OMP_SINGLE_LAYOUT")))
>    gomp_ordered : public gimple_statement_omp_single_layout
> {
>      /* No extra fields; adds invariant:
>           stmt->code == GIMPLE_OMP_ORDERED.  */
> };
> instead (like gomp_single, gomp_teams, ...).

Oh, neat.  I missed that.  Fixed.

>
>> @@ -149,6 +149,9 @@ struct gimplify_omp_ctx
>>     struct gimplify_omp_ctx *outer_context;
>>     splay_tree variables;
>>     hash_set<tree> *privatized_types;
>> +  /* Iteration variables in an OMP_FOR.  */
>> +  tree *iter_vars;
>> +  int niter_vars;
>
> Wonder if it wouldn't be better to use a vec<tree> instead.
> Then the size would be there as vec_length.

Done.

>
>> @@ -8169,6 +8185,19 @@ gimplify_transaction (tree *expr_p, gimple_seq *pre_p)
>>     return GS_ALL_DONE;
>>   }
>>
>> +/* Verify the validity of the depend(sink:...) variable VAR.
>> +   Return TRUE if everything is OK, otherwise return FALSE.  */
>> +
>> +static bool
>> +verify_sink_var (location_t loc, tree var)
>> +{
>> +  for (int i = 0; i < gimplify_omp_ctxp->niter_vars; ++i)
>> +    if (var == gimplify_omp_ctxp->iter_vars[i])
>> +      return true;
>> +  error_at (loc, "variable %qE is not an iteration variable", var);
>> +  return false;
>
> I believe what we want to verify is that ith variable in the OMP_CLAUSE_DECL
> vector is iter_vars[i], so not just some random permutation etc.

Fixed.

>
>> @@ -3216,7 +3218,51 @@ check_omp_nesting_restrictions (gimple stmt, omp_context *ctx)
>>   	    break;
>>   	  }
>>         break;
>> +    case GIMPLE_OMP_TASK:
>> +      for (c = gimple_omp_task_clauses (stmt); c; c = OMP_CLAUSE_CHAIN (c))
>> +	if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEPEND
>> +	    && (OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SOURCE
>> +		|| OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SINK))
>> +	  {
>> +	    error_at (OMP_CLAUSE_LOCATION (c),
>> +		      "depend(%s) is only available in 'omp ordered'",
>
> Please avoid using ' in diagnostics, it should be %<omp ordered%> instead.

Fixed.

>
>> +		      OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SOURCE
>> +		      ? "source" : "sink");
>> +	    return false;
>> +	  }
>> +      break;
>
> This will eventually be needed also for GIMPLE_OMP_TARGET and
> GIMPLE_OMP_ENTER/EXIT_DATA.  But as that isn't really supported right now,
> can wait.

I added an assert so we don't forget.

>
>>       case GIMPLE_OMP_ORDERED:
>> +      for (c = gimple_omp_ordered_clauses (as_a <gomp_ordered *> (stmt));
>> +	   c; c = OMP_CLAUSE_CHAIN (c))
>> +	{
>> +	  enum omp_clause_depend_kind kind = OMP_CLAUSE_DEPEND_KIND (c);
>> +	  if (kind == OMP_CLAUSE_DEPEND_SOURCE
>> +	      || kind == OMP_CLAUSE_DEPEND_SINK)
>> +	    {
>> +	      bool have_ordered = false;
>> +	      /* Look for containing ordered(N) loop.  */
>> +	      for (omp_context *ctx_ = ctx; ctx_; ctx_ = ctx_->outer)
>
> Please use octx or something similar, I don't like the trailing _ ;)

I hate it too, but check_omp_nesting_restrictions() already had a use of 
ctx_ so I followed suit.  Fixed in my code nevertheless.

>
>> +	      if (!have_ordered)
>> +		{
>> +		  error_at (OMP_CLAUSE_LOCATION (c),
>> +			    "depend clause is not within an ordered loop");
>
> Not within is not the right OpenMP term, the requirement is that it must
> be closely nested in ordered loop.

Done.

>
>> +/* depend(sink...) is allowed without an offset.  */
>> +#pragma omp ordered depend(sink:i,j+1)
>
> Can you write depend(sink:i,j-1) at least?  The iteration to depend
> on must be lexicographically earlier in the loop.

Sure.  Neither j+99 or j-HUGE are checked.  We allow anything 
INTEGER_CST.  Perhaps at expansion we can check the sanity of this? 
(Later, when we figure out what we're going to emit for the runtime).

>
>> +#pragma omp ordered depend(sink:i+2,j-2,k+2) /* { dg-error "is not an iteration var" } */
>
> Similarly.  i-2 will be enough.
>
>> --- a/gcc/tree-inline.c
>> +++ b/gcc/tree-inline.c
>> @@ -1479,7 +1479,7 @@ remap_gimple_stmt (gimple stmt, copy_body_data *id)
>>
>>   	case GIMPLE_OMP_ORDERED:
>>   	  s1 = remap_gimple_seq (gimple_omp_body (stmt), id);
>> -	  copy = gimple_build_omp_ordered (s1);
>> +	  copy = gimple_build_omp_ordered (s1, NULL);
>
> You surely don't want to pass NULL here, I bet you want
> gimple_omp_ordered_clauses (stmt) instead.

Fixed.

>
>> --- a/gcc/tree-pretty-print.c
>> +++ b/gcc/tree-pretty-print.c
>> @@ -533,6 +533,9 @@ dump_omp_clause (pretty_printer *pp, tree clause, int spc, int flags)
>>   	case OMP_CLAUSE_DEPEND_SOURCE:
>>   	  pp_string (pp, "source)");
>>   	  return;
>> +	case OMP_CLAUSE_DEPEND_SINK:
>> +	  pp_string (pp, "sink");
>> +	  break;
>
> And here you surely don't want to emit
> #pragma omp ordered(sink
> (note even the missing closing paren).
> It should dump the TREE_LIST (the var and if non-0 addend, the addend after
> it).

Notice this case had a break, not a return, so we would fall down to 
code that printed the TREE_LIST and added a closing parenthesis.  The 
TREE_LIST was in the form of "i 3", which I thought was obvious enough. 
  Be that as it may, I have added code to beautify it as "i+3" as suggested.

OK for branch?

Patch
diff mbox

diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c
index cd3bd5a..50edaf6 100644
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -11701,6 +11701,95 @@  c_parser_omp_clause_simdlen (c_parser *parser, tree list)
   return c;
 }
 
+/* OpenMP 4.1:
+   vec:
+     identifier [+/- integer]
+     vec , identifier [+/- integer]
+*/
+
+static tree
+c_parser_omp_clause_depend_sink (c_parser *parser, location_t clause_loc,
+				 tree list)
+{
+  tree vec = NULL;
+  if (c_parser_next_token_is_not (parser, CPP_NAME)
+      || c_parser_peek_token (parser)->id_kind != C_ID_ID)
+    {
+      c_parser_error (parser, "expected identifier");
+      return list;
+    }
+
+  while (c_parser_next_token_is (parser, CPP_NAME)
+	 && c_parser_peek_token (parser)->id_kind == C_ID_ID)
+    {
+      tree t = lookup_name (c_parser_peek_token (parser)->value);
+      tree addend = NULL;
+
+      if (t == NULL_TREE)
+	{
+	  undeclared_variable (c_parser_peek_token (parser)->location,
+			       c_parser_peek_token (parser)->value);
+	  t = error_mark_node;
+	}
+
+      c_parser_consume_token (parser);
+
+      if (t != error_mark_node)
+	{
+	  bool neg;
+
+	  if (c_parser_next_token_is (parser, CPP_MINUS))
+	    neg = true;
+	  else if (c_parser_next_token_is (parser, CPP_PLUS))
+	    neg = false;
+	  else
+	    {
+	      addend = integer_zero_node;
+	      goto add_to_vector;
+	    }
+	  c_parser_consume_token (parser);
+
+	  if (c_parser_next_token_is_not (parser, CPP_NUMBER))
+	    {
+	      c_parser_error (parser, "expected %<integer%>");
+	      return list;
+	    }
+
+	  addend = c_parser_peek_token (parser)->value;
+	  if (TREE_CODE (addend) != INTEGER_CST)
+	    {
+	      c_parser_error (parser, "expected %<integer%>");
+	      return list;
+	    }
+	  if (neg)
+	    {
+	      bool overflow;
+	      wide_int offset = wi::neg (addend, &overflow);
+	      addend = wide_int_to_tree (TREE_TYPE (addend), offset);
+	      if (overflow)
+		warning_at (c_parser_peek_token (parser)->location,
+			    OPT_Woverflow,
+			    "overflow in implicit constant conversion");
+	    }
+	  c_parser_consume_token (parser);
+
+	add_to_vector:
+	  vec = tree_cons (addend, t, vec);
+
+	  if (c_parser_next_token_is_not (parser, CPP_COMMA))
+	    break;
+
+	  c_parser_consume_token (parser);
+	}
+    }
+
+  tree u = build_omp_clause (clause_loc, OMP_CLAUSE_DEPEND);
+  OMP_CLAUSE_DEPEND_KIND (u) = OMP_CLAUSE_DEPEND_SINK;
+  OMP_CLAUSE_DECL (u) = nreverse (vec);
+  OMP_CLAUSE_CHAIN (u) = list;
+  return u;
+}
+
 /* OpenMP 4.0:
    depend ( depend-kind: variable-list )
 
@@ -11708,10 +11797,9 @@  c_parser_omp_clause_simdlen (c_parser *parser, tree list)
      in | out | inout
 
    OpenMP 4.1:
-   depend ( depend-loop-kind [ : vec ] )
+   depend ( source )
 
-   depend-loop-kind:
-     source | sink  */
+   depend ( sink  : vec )  */
 
 static tree
 c_parser_omp_clause_depend (c_parser *parser, tree list)
@@ -11754,16 +11842,19 @@  c_parser_omp_clause_depend (c_parser *parser, tree list)
       return c;
     }
 
-  /* FIXME: Handle OMP_CLAUSE_DEPEND_SINK.  */
-
   if (!c_parser_require (parser, CPP_COLON, "expected %<:%>"))
     goto resync_fail;
 
-  nl = c_parser_omp_variable_list (parser, clause_loc,
-				   OMP_CLAUSE_DEPEND, list);
+  if (kind == OMP_CLAUSE_DEPEND_SINK)
+    nl = c_parser_omp_clause_depend_sink (parser, clause_loc, list);
+  else
+    {
+      nl = c_parser_omp_variable_list (parser, clause_loc,
+				       OMP_CLAUSE_DEPEND, list);
 
-  for (c = nl; c != list; c = OMP_CLAUSE_CHAIN (c))
-    OMP_CLAUSE_DEPEND_KIND (c) = kind;
+      for (c = nl; c != list; c = OMP_CLAUSE_CHAIN (c))
+	OMP_CLAUSE_DEPEND_KIND (c) = kind;
+    }
 
   c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, "expected %<)%>");
   return nl;
diff --git a/gcc/c/c-typeck.c b/gcc/c/c-typeck.c
index 469cd88..0b332e8 100644
--- a/gcc/c/c-typeck.c
+++ b/gcc/c/c-typeck.c
@@ -12489,6 +12489,11 @@  c_finish_omp_clauses (tree clauses, bool declare_simd)
 			  == OMP_CLAUSE_DEPEND_SOURCE);
 	      break;
 	    }
+	  if (OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SINK)
+	    {
+	      gcc_assert (TREE_CODE (t) == TREE_LIST);
+	      break;
+	    }
 	  if (TREE_CODE (t) == TREE_LIST)
 	    {
 	      if (handle_omp_array_sections (c))
diff --git a/gcc/coretypes.h b/gcc/coretypes.h
index e3cec01..a708896 100644
--- a/gcc/coretypes.h
+++ b/gcc/coretypes.h
@@ -113,6 +113,7 @@  struct gomp_atomic_load;
 struct gomp_atomic_store;
 struct gomp_continue;
 struct gomp_critical;
+struct gomp_ordered;
 struct gomp_for;
 struct gomp_parallel;
 struct gomp_task;
diff --git a/gcc/gimple-pretty-print.c b/gcc/gimple-pretty-print.c
index cae8883..83ce3ec 100644
--- a/gcc/gimple-pretty-print.c
+++ b/gcc/gimple-pretty-print.c
@@ -1488,9 +1488,6 @@  dump_gimple_omp_block (pretty_printer *buffer, gimple gs, int spc, int flags)
 	case GIMPLE_OMP_TASKGROUP:
 	  pp_string (buffer, "#pragma omp taskgroup");
 	  break;
-	case GIMPLE_OMP_ORDERED:
-	  pp_string (buffer, "#pragma omp ordered");
-	  break;
 	case GIMPLE_OMP_SECTION:
 	  pp_string (buffer, "#pragma omp section");
 	  break;
@@ -1541,6 +1538,31 @@  dump_gimple_omp_critical (pretty_printer *buffer, gomp_critical *gs,
     }
 }
 
+/* Dump a GIMPLE_OMP_ORDERED tuple on the pretty_printer BUFFER.  */
+
+static void
+dump_gimple_omp_ordered (pretty_printer *buffer, gomp_ordered *gs,
+			 int spc, int flags)
+{
+  if (flags & TDF_RAW)
+    dump_gimple_fmt (buffer, spc, flags, "%G <%+BODY <%S> >", gs,
+		     gimple_omp_body (gs));
+  else
+    {
+      pp_string (buffer, "#pragma omp ordered");
+      dump_omp_clauses (buffer, gimple_omp_ordered_clauses (gs), spc, flags);
+      if (!gimple_seq_empty_p (gimple_omp_body (gs)))
+	{
+	  newline_and_indent (buffer, spc + 2);
+	  pp_left_brace (buffer);
+	  pp_newline (buffer);
+	  dump_gimple_seq (buffer, gimple_omp_body (gs), spc + 4, flags);
+	  newline_and_indent (buffer, spc + 2);
+	  pp_right_brace (buffer);
+	}
+    }
+}
+
 /* Dump a GIMPLE_OMP_RETURN tuple on the pretty_printer BUFFER.  */
 
 static void
@@ -2250,11 +2272,15 @@  pp_gimple_stmt_1 (pretty_printer *buffer, gimple gs, int spc, int flags)
 
     case GIMPLE_OMP_MASTER:
     case GIMPLE_OMP_TASKGROUP:
-    case GIMPLE_OMP_ORDERED:
     case GIMPLE_OMP_SECTION:
       dump_gimple_omp_block (buffer, gs, spc, flags);
       break;
 
+    case GIMPLE_OMP_ORDERED:
+      dump_gimple_omp_ordered (buffer, as_a <gomp_ordered *> (gs), spc,
+			       flags);
+      break;
+
     case GIMPLE_OMP_CRITICAL:
       dump_gimple_omp_critical (buffer, as_a <gomp_critical *> (gs), spc,
 				flags);
diff --git a/gcc/gimple-walk.c b/gcc/gimple-walk.c
index f0e2c67..1fe4365 100644
--- a/gcc/gimple-walk.c
+++ b/gcc/gimple-walk.c
@@ -327,6 +327,16 @@  walk_gimple_op (gimple stmt, walk_tree_fn callback_op,
       }
       break;
 
+    case GIMPLE_OMP_ORDERED:
+      {
+	gomp_ordered *omp_stmt = as_a <gomp_ordered *> (stmt);
+	ret = walk_tree (gimple_omp_ordered_clauses_ptr (omp_stmt),
+			 callback_op, wi, pset);
+	if (ret)
+	  return ret;
+      }
+      break;
+
     case GIMPLE_OMP_FOR:
       ret = walk_tree (gimple_omp_for_clauses_ptr (stmt), callback_op, wi,
 		       pset);
diff --git a/gcc/gimple.c b/gcc/gimple.c
index d8eb77a..b87d44d 100644
--- a/gcc/gimple.c
+++ b/gcc/gimple.c
@@ -991,12 +991,15 @@  gimple_build_omp_continue (tree control_def, tree control_use)
 /* Build a GIMPLE_OMP_ORDERED statement.
 
    BODY is the sequence of statements inside a loop that will executed in
-   sequence.  */
+   sequence.
+   CLAUSES are clauses for this statement.  */
 
-gimple
-gimple_build_omp_ordered (gimple_seq body)
+gomp_ordered *
+gimple_build_omp_ordered (gimple_seq body, tree clauses)
 {
-  gimple p = gimple_alloc (GIMPLE_OMP_ORDERED, 0);
+  gomp_ordered *p
+    = as_a <gomp_ordered *> (gimple_alloc (GIMPLE_OMP_ORDERED, 0));
+  gimple_omp_ordered_set_clauses (p, clauses);
   if (body)
     gimple_omp_set_body (p, body);
 
diff --git a/gcc/gimple.def b/gcc/gimple.def
index 96602df..d3ca402 100644
--- a/gcc/gimple.def
+++ b/gcc/gimple.def
@@ -283,10 +283,6 @@  DEFGSCODE(GIMPLE_OMP_MASTER, "gimple_omp_master", GSS_OMP)
    BODY is the sequence of statements to execute in the taskgroup section.  */
 DEFGSCODE(GIMPLE_OMP_TASKGROUP, "gimple_omp_taskgroup", GSS_OMP)
 
-/* GIMPLE_OMP_ORDERED <BODY> represents #pragma omp ordered.
-   BODY is the sequence of statements to execute in the ordered section.  */
-DEFGSCODE(GIMPLE_OMP_ORDERED, "gimple_omp_ordered", GSS_OMP)
-
 /* GIMPLE_OMP_PARALLEL <BODY, CLAUSES, CHILD_FN, DATA_ARG> represents
 
    #pragma omp parallel [CLAUSES]
@@ -375,6 +371,11 @@  DEFGSCODE(GIMPLE_OMP_TARGET, "gimple_omp_target", GSS_OMP_PARALLEL_LAYOUT)
    CLAUSES is an OMP_CLAUSE chain holding the associated clauses.  */
 DEFGSCODE(GIMPLE_OMP_TEAMS, "gimple_omp_teams", GSS_OMP_SINGLE_LAYOUT)
 
+/* GIMPLE_OMP_ORDERED <BODY, CLAUSES> represents #pragma omp ordered.
+   BODY is the sequence of statements to execute in the ordered section.
+   CLAUSES is an OMP_CLAUSE chain holding the associated clauses.  */
+DEFGSCODE(GIMPLE_OMP_ORDERED, "gimple_omp_ordered", GSS_OMP_SINGLE_LAYOUT)
+
 /* GIMPLE_PREDICT <PREDICT, OUTCOME> specifies a hint for branch prediction.
 
    PREDICT is one of the predictors from predict.def.
diff --git a/gcc/gimple.h b/gcc/gimple.h
index 6057ea0..65f662f 100644
--- a/gcc/gimple.h
+++ b/gcc/gimple.h
@@ -668,7 +668,7 @@  struct GTY((tag("GSS_OMP_CONTINUE")))
   tree control_use;
 };
 
-/* GIMPLE_OMP_SINGLE, GIMPLE_OMP_TEAMS */
+/* GIMPLE_OMP_SINGLE, GIMPLE_OMP_TEAMS, GIMPLE_OMP_ORDERED */
 
 struct GTY((tag("GSS_OMP_SINGLE_LAYOUT")))
   gimple_statement_omp_single_layout : public gimple_statement_omp
@@ -693,6 +693,13 @@  struct GTY((tag("GSS_OMP_SINGLE_LAYOUT")))
          stmt->code == GIMPLE_OMP_TEAMS.  */
 };
 
+struct GTY((tag("GSS_OMP_SINGLE_LAYOUT")))
+  gomp_ordered : public gimple_statement_omp_single_layout
+{
+    /* No extra fields; adds invariant:
+         stmt->code == GIMPLE_OMP_ORDERED.  */
+};
+
 
 /* GIMPLE_OMP_ATOMIC_LOAD.
    Note: This is based on gimple_statement_base, not g_s_omp, because g_s_omp
@@ -1007,6 +1014,14 @@  is_a_helper <gomp_critical *>::test (gimple gs)
 template <>
 template <>
 inline bool
+is_a_helper <gomp_ordered *>::test (gimple gs)
+{
+  return gs->code == GIMPLE_OMP_ORDERED;
+}
+
+template <>
+template <>
+inline bool
 is_a_helper <gomp_for *>::test (gimple gs)
 {
   return gs->code == GIMPLE_OMP_FOR;
@@ -1215,6 +1230,14 @@  is_a_helper <const gomp_critical *>::test (const_gimple gs)
 template <>
 template <>
 inline bool
+is_a_helper <const gomp_ordered *>::test (const_gimple gs)
+{
+  return gs->code == GIMPLE_OMP_ORDERED;
+}
+
+template <>
+template <>
+inline bool
 is_a_helper <const gomp_for *>::test (const_gimple gs)
 {
   return gs->code == GIMPLE_OMP_FOR;
@@ -1355,7 +1378,7 @@  gimple gimple_build_omp_section (gimple_seq);
 gimple gimple_build_omp_master (gimple_seq);
 gimple gimple_build_omp_taskgroup (gimple_seq);
 gomp_continue *gimple_build_omp_continue (tree, tree);
-gimple gimple_build_omp_ordered (gimple_seq);
+gomp_ordered *gimple_build_omp_ordered (gimple_seq, tree);
 gimple gimple_build_omp_return (bool);
 gomp_sections *gimple_build_omp_sections (gimple_seq, tree);
 gimple gimple_build_omp_sections_switch (void);
@@ -4460,6 +4483,35 @@  gimple_omp_critical_set_clauses (gomp_critical *crit_stmt, tree clauses)
 }
 
 
+/* Return the clauses associated with OMP_ORDERED statement ORD_STMT.  */
+
+static inline tree
+gimple_omp_ordered_clauses (const gomp_ordered *ord_stmt)
+{
+  return ord_stmt->clauses;
+}
+
+
+/* Return a pointer to the clauses associated with OMP ordered statement
+   ORD_STMT.  */
+
+static inline tree *
+gimple_omp_ordered_clauses_ptr (gomp_ordered *ord_stmt)
+{
+  return &ord_stmt->clauses;
+}
+
+
+/* Set CLAUSES to be the clauses associated with OMP ordered statement
+   ORD_STMT.  */
+
+static inline void
+gimple_omp_ordered_set_clauses (gomp_ordered *ord_stmt, tree clauses)
+{
+  ord_stmt->clauses = clauses;
+}
+
+
 /* Return the kind of the OMP_FOR statemement G.  */
 
 static inline int
diff --git a/gcc/gimplify.c b/gcc/gimplify.c
index 21f8223..d079530 100644
--- a/gcc/gimplify.c
+++ b/gcc/gimplify.c
@@ -149,6 +149,8 @@  struct gimplify_omp_ctx
   struct gimplify_omp_ctx *outer_context;
   splay_tree variables;
   hash_set<tree> *privatized_types;
+  /* Iteration variables in an OMP_FOR.  */
+  vec<tree> iter_vars;
   location_t location;
   enum omp_clause_default_kind default_kind;
   enum omp_region_type region_type;
@@ -365,6 +367,8 @@  new_omp_context (enum omp_region_type region_type)
 
   c = XCNEW (struct gimplify_omp_ctx);
   c->outer_context = gimplify_omp_ctxp;
+  c->iter_vars.safe_push(0);
+  c->iter_vars.pop();
   c->variables = splay_tree_new (splay_tree_compare_decl_uid, 0, 0);
   c->privatized_types = new hash_set<tree>;
   c->location = input_location;
@@ -384,6 +388,7 @@  delete_omp_context (struct gimplify_omp_ctx *c)
 {
   splay_tree_delete (c->variables);
   delete c->privatized_types;
+  c->iter_vars.release ();
   XDELETE (c);
 }
 
@@ -6343,6 +6348,13 @@  gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
 	  goto do_add;
 
 	case OMP_CLAUSE_DEPEND:
+	  if (OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SINK
+	      || OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SOURCE)
+	    {
+	      /* Nothing to do.  OMP_CLAUSE_DECL will be lowered in
+		 omp-low.c.  */
+	      break;
+	    }
 	  if (TREE_CODE (OMP_CLAUSE_DECL (c)) == COMPOUND_EXPR)
 	    {
 	      gimplify_expr (&TREE_OPERAND (OMP_CLAUSE_DECL (c), 0), pre_p,
@@ -7282,6 +7294,8 @@  gimplify_omp_for (tree *expr_p, gimple_seq *pre_p)
 	      == TREE_VEC_LENGTH (OMP_FOR_COND (for_stmt)));
   gcc_assert (TREE_VEC_LENGTH (OMP_FOR_INIT (for_stmt))
 	      == TREE_VEC_LENGTH (OMP_FOR_INCR (for_stmt)));
+  gimplify_omp_ctxp->iter_vars.create (TREE_VEC_LENGTH
+				       (OMP_FOR_INIT (for_stmt)));
   for (i = 0; i < TREE_VEC_LENGTH (OMP_FOR_INIT (for_stmt)); i++)
     {
       t = TREE_VEC_ELT (OMP_FOR_INIT (for_stmt), i);
@@ -7290,6 +7304,7 @@  gimplify_omp_for (tree *expr_p, gimple_seq *pre_p)
       gcc_assert (DECL_P (decl));
       gcc_assert (INTEGRAL_TYPE_P (TREE_TYPE (decl))
 		  || POINTER_TYPE_P (TREE_TYPE (decl)));
+      gimplify_omp_ctxp->iter_vars.quick_push (decl);
 
       /* Make sure the iteration variable is private.  */
       tree c = NULL_TREE;
@@ -8982,7 +8997,36 @@  gimplify_expr (tree *expr_p, gimple_seq *pre_p, gimple_seq *post_p,
 		}
 		break;
 	      case OMP_ORDERED:
-		g = gimple_build_omp_ordered (body);
+		if (gimplify_omp_ctxp)
+		  for (tree c = OMP_ORDERED_CLAUSES (*expr_p);
+		       c; c = OMP_CLAUSE_CHAIN (c))
+		    if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEPEND
+			&& OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SINK)
+		      {
+			unsigned int n = 0;
+			bool fail = false;
+			for (tree decls = OMP_CLAUSE_DECL (c);
+			     decls && TREE_CODE (decls) == TREE_LIST;
+			     decls = TREE_CHAIN (decls), ++n)
+			  if (n < gimplify_omp_ctxp->iter_vars.length ()
+			      && TREE_VALUE (decls)
+			      != gimplify_omp_ctxp->iter_vars[n])
+			    {
+			      error_at (OMP_CLAUSE_LOCATION (c),
+					"variable %qE is not an iteration "
+					"variable", TREE_VALUE (decls));
+			      fail = true;
+			    }
+			/* Avoid being too redundant.  */
+			if (!fail
+			    && n != gimplify_omp_ctxp->iter_vars.length ())
+			  error_at (OMP_CLAUSE_LOCATION (c),
+			     "number of variables in depend(sink) clause "
+			     "does not match number of iteration variables");
+		      }
+
+		g = gimple_build_omp_ordered (body,
+					      OMP_ORDERED_CLAUSES (*expr_p));
 		break;
 	      case OMP_CRITICAL:
 		gimplify_scan_omp_clauses (&OMP_CRITICAL_CLAUSES (*expr_p),
diff --git a/gcc/omp-low.c b/gcc/omp-low.c
index 83677ea..3dec095 100644
--- a/gcc/omp-low.c
+++ b/gcc/omp-low.c
@@ -2996,6 +2996,8 @@  scan_omp_teams (gomp_teams *stmt, omp_context *outer_ctx)
 static bool
 check_omp_nesting_restrictions (gimple stmt, omp_context *ctx)
 {
+  tree c;
+
   /* No nesting of non-OpenACC STMT (that is, an OpenMP one, or a GOMP builtin)
      inside an OpenACC CTX.  */
   if (!(is_gimple_omp (stmt)
@@ -3216,7 +3218,54 @@  check_omp_nesting_restrictions (gimple stmt, omp_context *ctx)
 	    break;
 	  }
       break;
+    case GIMPLE_OMP_TASK:
+      for (c = gimple_omp_task_clauses (stmt); c; c = OMP_CLAUSE_CHAIN (c))
+	if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEPEND
+	    && (OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SOURCE
+		|| OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SINK))
+	  {
+	    enum omp_clause_depend_kind kind = OMP_CLAUSE_DEPEND_KIND (c);
+	    gcc_assert (kind == OMP_CLAUSE_DEPEND_SOURCE
+			|| kind == OMP_CLAUSE_DEPEND_SINK);
+	    error_at (OMP_CLAUSE_LOCATION (c),
+		      "depend(%s) is only available in %<omp ordered%>",
+		      kind == OMP_CLAUSE_DEPEND_SOURCE ? "source" : "sink");
+	    return false;
+	  }
+      break;
     case GIMPLE_OMP_ORDERED:
+      for (c = gimple_omp_ordered_clauses (as_a <gomp_ordered *> (stmt));
+	   c; c = OMP_CLAUSE_CHAIN (c))
+	{
+	  enum omp_clause_depend_kind kind = OMP_CLAUSE_DEPEND_KIND (c);
+	  if (kind == OMP_CLAUSE_DEPEND_SOURCE
+	      || kind == OMP_CLAUSE_DEPEND_SINK)
+	    {
+	      bool have_ordered = false;
+	      /* Look for containing ordered(N) loop.  */
+	      for (omp_context *octx = ctx; octx; octx = octx->outer)
+		if (gimple_code (octx->stmt) == GIMPLE_OMP_FOR
+		    && find_omp_clause (gimple_omp_for_clauses (octx->stmt),
+					OMP_CLAUSE_ORDERED))
+		  {
+		    have_ordered = true;
+		    break;
+		  }
+	      if (!have_ordered)
+		{
+		  error_at (OMP_CLAUSE_LOCATION (c),
+			    "depend clause must be closely nested inside an "
+			    "ordered loop");
+		  return false;
+		}
+	    }
+	  else
+	    {
+	      error_at (OMP_CLAUSE_LOCATION (c),
+			"invalid depend kind in omp ordered depend");
+	      return false;
+	    }
+	}
       for (; ctx != NULL; ctx = ctx->outer)
 	switch (gimple_code (ctx->stmt))
 	  {
@@ -12562,6 +12611,10 @@  lower_depend_clauses (gimple stmt, gimple_seq *iseq, gimple_seq *oseq)
 	case OMP_CLAUSE_DEPEND_INOUT:
 	  n_out++;
 	  break;
+	case OMP_CLAUSE_DEPEND_SOURCE:
+	case OMP_CLAUSE_DEPEND_SINK:
+	  /* FIXME:  */
+	  break;
 	default:
 	  gcc_unreachable ();
 	}
diff --git a/gcc/testsuite/c-c++-common/gomp/sink-1.c b/gcc/testsuite/c-c++-common/gomp/sink-1.c
new file mode 100644
index 0000000..d411d97
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/sink-1.c
@@ -0,0 +1,88 @@ 
+/* { dg-do compile } */
+/* { dg-options "-fopenmp -Wunknown-pragmas -Werror" } */
+
+extern void bark (void);
+int i,j,k;
+int array[555];
+
+int
+main()
+{
+#pragma omp parallel for ordered(2)
+  for (i=0; i < 100; ++i)
+    for (j=0; j < 100; ++j)
+      {
+/* OUT variant does not apply to ORDERED construct.  */
+#pragma omp ordered depend(out:i) /* { dg-error "invalid depend kind" } */
+
+/* depend(sink...) is allowed without an offset.  */
+#pragma omp ordered depend(sink:i,j-1)
+
+#pragma omp ordered depend(sink:i-1,j+2)
+      bark ();
+      }
+
+/* depend(sink...) does not apply to `omp task'.  */
+#pragma omp task depend(sink:i+3) /* { dg-error "only available in 'omp ordered'" } */
+  bark();
+
+#pragma omp ordered depend(source) /* { dg-error "depend clause must be closely nested" } */
+
+#pragma omp parallel for ordered(2)
+  for (i=0; i < 100; ++i)
+    for (j=0; j < 100; ++j)
+      {
+/* Multiple depend(source) allowed.  */
+#pragma omp ordered depend(source)
+#pragma omp ordered depend(source)
+      }
+
+#pragma omp parallel for ordered(2)
+  for (i=0; i < 100; ++i)
+    for (j=0; j < 100; ++j)
+      {
+#pragma omp ordered depend(sink:i-2,j-2,k+2) /* { dg-error "does not match number of iteration var" } */
+	bark();
+      }
+
+#pragma omp parallel for ordered(2)
+  for (i=0; i < 100; ++i)
+    for (j=0; j < 100; ++j)
+      {
+#pragma omp ordered depend(sink:i-2) /* { dg-error "does not match number of iteration variables" } */
+	bark();
+      }
+}
+
+void bar (int, int, int);
+
+void
+foo (int n, int m, int o)
+{
+  int i, j, k;
+  #pragma omp for collapse(2) ordered(2)
+  for (i = 0; i < m; i++)
+    {
+      for (j = 0; j < n; j++)
+        for (k = 0; k < o; k++)
+          {
+#pragma omp ordered depend(sink: i-1,j,k) depend(sink: i,j-1,k-1) depend(sink: i-1,j-1,k+1)
+	    bar (i, j, k);
+#pragma omp ordered depend(source)
+	  }
+    }
+}
+
+int
+baz ()
+{
+  int i, j;
+#pragma omp parallel for ordered(2)
+  for (i=0; i < 100; ++i)
+    for (j=0; j < 100; ++j)
+      {
+#pragma omp ordered depend(sink:i-1,j-3)
+        bar (i, j, 0);
+#pragma omp ordered depend(source)
+      }
+}
diff --git a/gcc/tree-inline.c b/gcc/tree-inline.c
index 06b4077..f4e8cf2 100644
--- a/gcc/tree-inline.c
+++ b/gcc/tree-inline.c
@@ -1479,7 +1479,9 @@  remap_gimple_stmt (gimple stmt, copy_body_data *id)
 
 	case GIMPLE_OMP_ORDERED:
 	  s1 = remap_gimple_seq (gimple_omp_body (stmt), id);
-	  copy = gimple_build_omp_ordered (s1);
+	  copy = gimple_build_omp_ordered
+	           (s1,
+		    gimple_omp_ordered_clauses (as_a <gomp_ordered *> (stmt)));
 	  break;
 
 	case GIMPLE_OMP_SECTION:
diff --git a/gcc/tree-pretty-print.c b/gcc/tree-pretty-print.c
index 7f8e0fe..3c5d0f5 100644
--- a/gcc/tree-pretty-print.c
+++ b/gcc/tree-pretty-print.c
@@ -533,6 +533,22 @@  dump_omp_clause (pretty_printer *pp, tree clause, int spc, int flags)
 	case OMP_CLAUSE_DEPEND_SOURCE:
 	  pp_string (pp, "source)");
 	  return;
+	case OMP_CLAUSE_DEPEND_SINK:
+	  pp_string (pp, "sink:");
+	  for (tree t = OMP_CLAUSE_DECL (clause); t; t = TREE_CHAIN (t))
+	    if (TREE_CODE (t) == TREE_LIST)
+	      {
+		dump_generic_node (pp, TREE_VALUE (t), spc, flags, false);
+		if (TREE_PURPOSE (t) != integer_zero_node)
+		  dump_generic_node (pp, TREE_PURPOSE (t), spc, flags,
+				     false);
+		if (TREE_CHAIN (t))
+		  pp_comma (pp);
+	      }
+	    else
+	      gcc_unreachable ();
+	  pp_right_paren (pp);
+	  return;
 	default:
 	  gcc_unreachable ();
 	}