[gomp4.1] calculate pointer offsets for depend(sink)

diff --git a/gcc/c/c-typeck.c b/gcc/c/c-typeck.c
index b7714e3..597973b 100644
--- a/gcc/c/c-typeck.c
+++ b/gcc/c/c-typeck.c
@@ -12492,6 +12492,24 @@  c_finish_omp_clauses (tree clauses, bool declare_simd)
 	  if (OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SINK)
 	    {
 	      gcc_assert (TREE_CODE (t) == TREE_LIST);
+	      for (; t; t = TREE_CHAIN (t))
+		{
+		  tree decl = TREE_VALUE (t);
+		  if (TREE_CODE (TREE_TYPE (decl)) == POINTER_TYPE)
+		    {
+		      tree offset = TREE_PURPOSE (t);
+		      bool neg = wi::neg_p ((wide_int) offset);
+		      offset = fold_unary (ABS_EXPR, TREE_TYPE (offset), offset);
+		      tree t2 = pointer_int_sum (OMP_CLAUSE_LOCATION (c),
+						 neg ? MINUS_EXPR : PLUS_EXPR,
+						 decl, offset);
+		      t2 = fold_build2_loc (OMP_CLAUSE_LOCATION (c), MINUS_EXPR,
+					    sizetype, t2, decl);
+		      if (t2 == error_mark_node)
+			t2 = size_one_node; // ??
+		      TREE_PURPOSE (t) = t2;
+		    }
+		}
 	      break;
 	    }
 	  if (TREE_CODE (t) == TREE_LIST)
diff --git a/gcc/cp/semantics.c b/gcc/cp/semantics.c
index 8c05936..2598433 100644
--- a/gcc/cp/semantics.c
+++ b/gcc/cp/semantics.c
@@ -5532,6 +5532,41 @@  finish_omp_declare_simd_methods (tree t)
     }
 }
 
+/* Adjust sink depend clause to take into account pointer offsets.  */
+
+static void
+cp_finish_omp_clause_depend_sink (tree sink_clause)
+{
+  tree t = OMP_CLAUSE_DECL (sink_clause);
+  gcc_assert (TREE_CODE (t) == TREE_LIST);
+
+  /* Make sure we don't adjust things twice for templates.  */
+  if (processing_template_decl)
+    return;
+
+  for (; t; t = TREE_CHAIN (t))
+    {
+      tree decl = TREE_VALUE (t);
+      if (TREE_CODE (TREE_TYPE (decl)) == POINTER_TYPE)
+	{
+	  tree offset = TREE_PURPOSE (t);
+	  bool neg = wi::neg_p ((wide_int) offset);
+	  offset = fold_unary (ABS_EXPR, TREE_TYPE (offset), offset);
+	  decl = mark_rvalue_use (decl);
+	  decl = convert_from_reference (decl);
+	  tree t2 = pointer_int_sum (OMP_CLAUSE_LOCATION (sink_clause),
+				     neg ? MINUS_EXPR : PLUS_EXPR,
+				     decl, offset);
+	  t2 = fold_build2_loc (OMP_CLAUSE_LOCATION (sink_clause),
+				MINUS_EXPR, sizetype, t2,
+				decl);
+	  if (t2 == error_mark_node)
+	    t2 = size_one_node; // ??
+	  TREE_PURPOSE (t) = t2;
+	}
+    }
+}
+
 /* For all elements of CLAUSES, validate them vs OpenMP constraints.
    Remove any elements from the list that are invalid.  */
 
@@ -6150,7 +6185,7 @@  finish_omp_clauses (tree clauses, bool allow_fields, bool declare_simd)
 	    }
 	  if (OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SINK)
 	    {
-	      gcc_assert (TREE_CODE (t) == TREE_LIST);
+	      cp_finish_omp_clause_depend_sink (c);
 	      break;
 	    }
 	  if (TREE_CODE (t) == TREE_LIST)
diff --git a/gcc/testsuite/c-c++-common/gomp/sink-4.c b/gcc/testsuite/c-c++-common/gomp/sink-4.c
new file mode 100644
index 0000000..2872404
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/sink-4.c
@@ -0,0 +1,27 @@ 
+/* { dg-do compile } */
+/* { dg-options "-fopenmp -fdump-tree-gimple" } */
+
+/* Test that we adjust pointer offsets for sink variables
+   correctly.  */
+
+typedef struct {
+    char stuff[400];
+} foo;
+
+foo *p, *q;
+
+void
+funk ()
+{
+  int i,j;
+#pragma omp parallel for ordered(1)
+  for (p=q; p < q; p--)
+    {
+#pragma omp ordered depend(sink:p+1)
+      void bar ();
+        bar();
+#pragma omp ordered depend(source)
+    }
+}
+
+/* { dg-final { scan-tree-dump-times "depend\\(sink:p\\+400\\)" 1 "gimple" } } */
diff --git a/gcc/testsuite/g++.dg/gomp/sink-3.C b/gcc/testsuite/g++.dg/gomp/sink-3.C
new file mode 100644
index 0000000..83a742e
--- /dev/null
+++ b/gcc/testsuite/g++.dg/gomp/sink-3.C
@@ -0,0 +1,33 @@ 
+/* { dg-do compile } */
+/* { dg-options "-fopenmp -fdump-tree-gimple" } */
+
+/* Test that we adjust pointer offsets for sink variables
+   correctly.  */
+
+typedef struct {
+    char stuff[400];
+} foo;
+
+foo *p, *q, *r;
+
+template<int N>
+void
+funk ()
+{
+  int i,j;
+#pragma omp parallel for ordered(1)
+  for (p=q; p < q; p--)
+    {
+#pragma omp ordered depend(sink:p+1)
+      void bar ();
+        bar();
+#pragma omp ordered depend(source)
+    }
+}
+
+void foobar()
+{
+  funk<3>();
+}
+
+/* { dg-final { scan-tree-dump-times "depend\\(sink:p\\+400\\)" 1 "gimple" } } */
diff --git a/gcc/tree-pretty-print.c b/gcc/tree-pretty-print.c
index d3cc245..aab2bfc 100644
--- a/gcc/tree-pretty-print.c
+++ b/gcc/tree-pretty-print.c
@@ -567,7 +567,8 @@  dump_omp_clause (pretty_printer *pp, tree clause, int spc, int flags)
 		dump_generic_node (pp, TREE_VALUE (t), spc, flags, false);
 		if (TREE_PURPOSE (t) != integer_zero_node)
 		  {
-		    if (!wi::neg_p (TREE_PURPOSE (t)))
+		    tree p = TREE_PURPOSE (t);
+		    if (!wi::neg_p (p, TYPE_SIGN (TREE_TYPE (p))))
 		      pp_plus (pp);
 		    dump_generic_node (pp, TREE_PURPOSE (t), spc, flags,
 				       false);