[v2,04/12] AArch64: Diagnose OpenMP offloading when SVE types involved.

diff --git a/gcc/config/aarch64/aarch64-sve-builtins.cc b/gcc/config/aarch64/aarch64-sve-builtins.cc
index e7c703c987e..2c169ea3806 100644
--- a/gcc/config/aarch64/aarch64-sve-builtins.cc
+++ b/gcc/config/aarch64/aarch64-sve-builtins.cc
@@ -4956,12 +4956,35 @@  handle_arm_sve_vector_bits_attribute (tree *node, tree, tree args, int,
   return NULL_TREE;
 }
 
+
+/* Return true if OpenMP context types.  */
+
+static bool
+omp_type_context (type_context_kind context)
+{
+  switch (context)
+    {
+    case TCTX_OMP_MAP:
+    case TCTX_OMP_MAP_IMP_REF:
+    case TCTX_OMP_PRIVATE:
+    case TCTX_OMP_FIRSTPRIVATE:
+    case TCTX_OMP_DEVICE_ADDR:
+      return true;
+    default:
+      return false;;
+    }
+}
+
 /* Implement TARGET_VERIFY_TYPE_CONTEXT for SVE types.  */
 bool
 verify_type_context (location_t loc, type_context_kind context,
 		     const_tree type, bool silent_p)
 {
-  if (!sizeless_type_p (type))
+  const_tree tmp = type;
+  if (omp_type_context (context) && POINTER_TYPE_P (type))
+    tmp = strip_pointer_types (tmp);
+
+  if (!sizeless_type_p (tmp))
     return true;
 
   switch (context)
@@ -5021,6 +5044,33 @@  verify_type_context (location_t loc, type_context_kind context,
       if (!silent_p)
 	error_at (loc, "capture by copy of SVE type %qT", type);
       return false;
+
+    case TCTX_OMP_MAP:
+      if (!silent_p)
+	error_at (loc, "SVE type %qT not allowed in map clause", type);
+      return false;
+
+    case TCTX_OMP_MAP_IMP_REF:
+      /* The diagnosis is done in the caller.  */
+      return false;
+
+    case TCTX_OMP_PRIVATE:
+      if (!silent_p)
+	error_at (loc, "SVE type %qT not allowed in target private clause", type);
+      return false;
+
+    case TCTX_OMP_FIRSTPRIVATE:
+      if (!silent_p)
+	error_at (loc, "SVE type %qT not allowed in target firstprivate clause", type);
+      return false;
+
+    case TCTX_OMP_DEVICE_ADDR:
+      if (!silent_p)
+	error_at (loc, "SVE type %qT not allowed in target device clauses", type);
+      return false;
+
+    default:
+      break;
     }
   gcc_unreachable ();
 }
diff --git a/gcc/gimplify.cc b/gcc/gimplify.cc
index 3f602469d57..ace43cf78a0 100644
--- a/gcc/gimplify.cc
+++ b/gcc/gimplify.cc
@@ -8430,11 +8430,13 @@  omp_notice_variable (struct gimplify_omp_ctx *ctx, tree decl, bool in_code)
 			  | GOVD_MAP_ALLOC_ONLY)) == flags)
 	    {
 	      tree type = TREE_TYPE (decl);
+	      location_t dummy = UNKNOWN_LOCATION;
 
 	      if (gimplify_omp_ctxp->target_firstprivatize_array_bases
 		  && omp_privatize_by_reference (decl))
 		type = TREE_TYPE (type);
-	      if (!omp_mappable_type (type))
+	      if (!omp_mappable_type (type)
+		  || !verify_type_context (dummy, TCTX_OMP_MAP_IMP_REF, type))
 		{
 		  error ("%qD referenced in target region does not have "
 			 "a mappable type", decl);
@@ -12165,6 +12167,8 @@  gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
       unsigned int flags;
       tree decl;
       auto_vec<omp_addr_token *, 10> addr_tokens;
+      tree op = NULL_TREE;
+      location_t loc = OMP_CLAUSE_LOCATION (c);
 
       if (grp_end && c == OMP_CLAUSE_CHAIN (grp_end))
 	{
@@ -12172,6 +12176,34 @@  gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
 	  grp_end = NULL_TREE;
 	}
 
+      if (code == OMP_TARGET || code == OMP_TARGET_DATA
+	  || code == OMP_TARGET_ENTER_DATA || code == OMP_TARGET_EXIT_DATA)
+	/* Do some target-specific type checks for map operands.  */
+	switch (OMP_CLAUSE_CODE (c))
+	  {
+	  case OMP_CLAUSE_MAP:
+	    op = OMP_CLAUSE_OPERAND (c, 0);
+	    verify_type_context (loc, TCTX_OMP_MAP, TREE_TYPE (op));
+	    break;
+	  case OMP_CLAUSE_PRIVATE:
+	    op = OMP_CLAUSE_OPERAND (c, 0);
+	    verify_type_context (loc, TCTX_OMP_PRIVATE, TREE_TYPE (op));
+	    break;
+	  case OMP_CLAUSE_FIRSTPRIVATE:
+	    op = OMP_CLAUSE_OPERAND (c, 0);
+	    verify_type_context (loc, TCTX_OMP_FIRSTPRIVATE, TREE_TYPE (op));
+	    break;
+	  case OMP_CLAUSE_IS_DEVICE_PTR:
+	  case OMP_CLAUSE_USE_DEVICE_ADDR:
+	  case OMP_CLAUSE_USE_DEVICE_PTR:
+	  case OMP_CLAUSE_HAS_DEVICE_ADDR:
+	    op = OMP_CLAUSE_OPERAND (c, 0);
+	    verify_type_context (loc, TCTX_OMP_DEVICE_ADDR, TREE_TYPE (op));
+	    break;
+	  default:
+	    break;
+	  }
+
       switch (OMP_CLAUSE_CODE (c))
 	{
 	case OMP_CLAUSE_PRIVATE:
diff --git a/gcc/target.h b/gcc/target.h
index 837651d273a..7791daf6315 100644
--- a/gcc/target.h
+++ b/gcc/target.h
@@ -271,7 +271,24 @@  enum type_context_kind {
   TCTX_EXCEPTIONS,
 
   /* Capturing objects of type T by value in a closure.  */
-  TCTX_CAPTURE_BY_COPY
+  TCTX_CAPTURE_BY_COPY,
+
+  /* Objects of type T appearing in OpenMP map clause.  */
+  TCTX_OMP_MAP,
+
+  /* Objects of type T appearing in OpenMP target region
+     without explicit map.  */
+  TCTX_OMP_MAP_IMP_REF,
+
+  /* Objects of type T appearing in OpenMP private clause.  */
+  TCTX_OMP_PRIVATE,
+
+  /* Objects of type T appearing in OpenMP firstprivate clause.  */
+  TCTX_OMP_FIRSTPRIVATE,
+
+  /* Objects of type T appearing in OpenMP device clauses.  */
+  TCTX_OMP_DEVICE_ADDR
+
 };
 
 enum poly_value_estimate_kind
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel-loop.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel-loop.c
new file mode 100644
index 00000000000..b8e078fc816
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel-loop.c
@@ -0,0 +1,442 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define N __ARM_FEATURE_SVE_BITS
+
+svint32_t
+omp_target_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel loop
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_1_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel loop map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_2_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel loop map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target parallel loop map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_enter_exit_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target parallel loop map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target parallel loop map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_alloc_update_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+{
+#pragma omp target parallel loop
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target parallel loop map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+}
+  return va;
+}
+
+int64_t
+omp_target_private_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+/* Combined construct scenario: here private applies to the parallel loop
+   construct, so no error.  */
+#pragma omp target parallel loop private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vla (svbool_t vp)
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel loop firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N)))
+typedef svint32_t v8si FIXED_ATTR;
+typedef svbool_t v8bi FIXED_ATTR;
+
+v8si
+omp_target_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel loop
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_1_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel loop map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_2_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel loop map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target parallel loop map(to: b, c) map(tofrom: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_map_data_enter_exit_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target parallel loop map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target parallel loop map(to: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va)
+
+  return va;
+}
+
+v8si
+omp_target_map_data_alloc_update_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va)
+{
+#pragma omp target parallel loop
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target parallel loop map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+}
+  return va;
+}
+
+int64_t
+omp_target_private_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel loop private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vls (v8bi vp)
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel loop firstprivate (vp) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b);
+      vc = svld1_s32 (vp, c);
+      va = svadd_s32_z (vp, vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel.c
new file mode 100644
index 00000000000..b8edaff6755
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-parallel.c
@@ -0,0 +1,376 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define parallel parallel
+#define N __ARM_FEATURE_SVE_BITS
+
+svint32_t
+omp_target_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+  }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_1_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_2_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+
+#pragma omp target parallel map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+#pragma omp target parallel map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_enter_exit_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target parallel map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+#pragma omp target parallel map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_alloc_update_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+{
+#pragma omp target parallel
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+  }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target parallel map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+}
+  return va;
+}
+
+#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N)))
+typedef svint32_t v8si FIXED_ATTR;
+
+v8si
+omp_target_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_1_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target parallel map(to: b, c) map(from: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_2_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+
+#pragma omp target parallel map(to: b, c) map(from: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+#pragma omp target parallel map(to: b, c) map(tofrom: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+  return va;
+}
+
+v8si
+omp_target_map_data_enter_exit_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target parallel map(from: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+#pragma omp target parallel map(to: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+#pragma omp target exit data map(from: va)
+
+  return va;
+}
+
+v8si
+omp_target_map_data_alloc_update_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va)
+{
+#pragma omp target parallel
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target parallel map(from: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+}
+  return va;
+}
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-simd.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-simd.c
new file mode 100644
index 00000000000..a09aa5399f4
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-simd.c
@@ -0,0 +1,442 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define N __ARM_FEATURE_SVE_BITS
+
+svint32_t
+omp_target_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target simd
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_1_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target simd map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_2_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target simd map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target simd map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_enter_exit_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target simd map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target simd map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_alloc_update_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+{
+#pragma omp target simd
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target simd map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+}
+  return va;
+}
+
+int64_t
+omp_target_private_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+/* Combined construct scenario: here private applies to the simd construct so
+   no error.  */
+#pragma omp target simd private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vla (svbool_t vp)
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target simd firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N)))
+typedef svint32_t v8si FIXED_ATTR;
+typedef svbool_t v8bi FIXED_ATTR;
+
+v8si
+omp_target_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target simd
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_1_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target simd map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_2_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target simd map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target simd map(to: b, c) map(tofrom: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_map_data_enter_exit_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target simd map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target simd map(to: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va)
+
+  return va;
+}
+
+v8si
+omp_target_map_data_alloc_update_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va)
+{
+#pragma omp target simd
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target simd map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+}
+  return va;
+}
+
+int64_t
+omp_target_private_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target simd private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vls (v8bi vp)
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target simd firstprivate (vp) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b);
+      vc = svld1_s32 (vp, c);
+      va = svadd_s32_z (vp, vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute-simd.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute-simd.c
new file mode 100644
index 00000000000..3a998caeefd
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute-simd.c
@@ -0,0 +1,442 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define N __ARM_FEATURE_SVE_BITS
+
+svint32_t
+omp_target_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute simd
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_1_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute simd map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_2_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute simd map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target teams distribute simd map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_enter_exit_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target teams distribute simd map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target teams distribute simd map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_alloc_update_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+#pragma omp target teams distribute simd
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target teams distribute simd map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+  return va;
+}
+
+int64_t
+omp_target_private_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+/* Combined construct scenario: here private applies to the distribute simd
+   construct, so no error.  */
+#pragma omp target teams distribute simd private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vla (svbool_t vp)
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute simd firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N)))
+typedef svint32_t v8si FIXED_ATTR;
+typedef svbool_t v8bi FIXED_ATTR;
+
+v8si
+omp_target_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute simd
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_1_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute simd map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_2_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute simd map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target teams distribute simd map(to: b, c) map(tofrom: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_map_data_enter_exit_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target teams distribute simd map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target teams distribute simd map(to: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va)
+
+  return va;
+}
+
+v8si
+omp_target_map_data_alloc_update_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va)
+  {
+#pragma omp target teams distribute simd
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target teams distribute simd map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+  return va;
+}
+
+int64_t
+omp_target_private_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute simd private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vls (v8bi vp)
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute simd firstprivate (vp) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b);
+      vc = svld1_s32 (vp, c);
+      va = svadd_s32_z (vp, vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute.c
new file mode 100644
index 00000000000..dfb78ef69ee
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-distribute.c
@@ -0,0 +1,442 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define N __ARM_FEATURE_SVE_BITS
+
+svint32_t
+omp_target_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+    }
+  return va;
+}
+
+svint32_t
+omp_target_data_map_1_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_2_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target teams distribute map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_enter_exit_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target teams distribute map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target teams distribute map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_alloc_update_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+#pragma omp target teams distribute
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target teams distribute map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+  return va;
+}
+
+int64_t
+omp_target_private_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+/* Combined construct scenario: here private applies to the teams distribute
+   construct, so no error.  */
+#pragma omp target teams distribute private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate (svbool_t vp)
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N)))
+typedef svint32_t v8si FIXED_ATTR;
+typedef svbool_t v8bi FIXED_ATTR;
+
+v8si
+omp_target_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+  return va;
+}
+
+v8si
+omp_target_data_map_1_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_2_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target teams distribute map(to: b, c) map(tofrom: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_map_data_enter_exit_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target teams distribute map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target teams distribute map(to: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va)
+
+  return va;
+}
+
+v8si
+omp_target_map_data_alloc_update_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va)
+  {
+#pragma omp target teams distribute
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target teams distribute map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+  return va;
+}
+
+int64_t
+omp_target_private_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vls (v8bi vp)
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams distribute firstprivate (vp) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b);
+      vc = svld1_s32 (vp, c);
+      va = svadd_s32_z (vp, vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-loop.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-loop.c
new file mode 100644
index 00000000000..4c96f5a0fc8
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams-loop.c
@@ -0,0 +1,442 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define N __ARM_FEATURE_SVE_BITS
+
+svint32_t
+omp_target_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams loop
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_1_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams loop map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_2_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams loop map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target teams loop map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_enter_exit_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target teams loop map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target teams loop map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_alloc_update_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+{
+#pragma omp target teams loop
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target teams loop map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+}
+  return va;
+}
+
+int64_t
+omp_target_private_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+/* Combined construct scenario: here private applies to the teams loop
+   construct, so no error.  */
+#pragma omp target teams loop private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vla (svbool_t vp)
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams loop firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N)))
+typedef svint32_t v8si FIXED_ATTR;
+typedef svbool_t v8bi FIXED_ATTR;
+
+v8si
+omp_target_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams loop
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_1_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams loop map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_2_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams loop map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target teams loop map(to: b, c) map(tofrom: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_map_data_enter_exit_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target teams loop map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target teams loop map(to: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va)
+
+  return va;
+}
+
+v8si
+omp_target_map_data_alloc_update_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va)
+{
+#pragma omp target teams loop
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target teams loop map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+}
+  return va;
+}
+
+int64_t
+omp_target_private_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams loop private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vls (v8bi vp)
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams loop firstprivate (vp) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b);
+      vc = svld1_s32 (vp, c);
+      va = svadd_s32_z (vp, vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams.c
new file mode 100644
index 00000000000..2c5bf7e8926
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload-teams.c
@@ -0,0 +1,365 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define N __ARM_FEATURE_SVE_BITS
+
+svint32_t
+omp_target_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+  }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_1_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_2_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+
+#pragma omp target teams map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+#pragma omp target teams map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_enter_exit_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target teams map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+#pragma omp target teams map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_alloc_update_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+{
+#pragma omp target teams
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+  }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target teams map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+}
+  return va;
+}
+
+#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N)))
+typedef svint32_t v8si FIXED_ATTR;
+
+v8si
+omp_target_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_1_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target teams map(to: b, c) map(from: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_2_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+
+#pragma omp target teams map(to: b, c) map(from: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+#pragma omp target teams map(to: b, c) map(tofrom: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+  return va;
+}
+
+v8si
+omp_target_map_data_enter_exit_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target teams map(from: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+#pragma omp target teams map(to: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+
+#pragma omp target exit data map(from: va)
+
+  return va;
+}
+
+v8si
+omp_target_map_data_alloc_update_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va)
+{
+#pragma omp target teams
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+  }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target teams map(from: va)
+  {
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+  }
+}
+  return va;
+}
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/offload.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload.c
new file mode 100644
index 00000000000..b2f6e543531
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/offload.c
@@ -0,0 +1,452 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define N __ARM_FEATURE_SVE_BITS
+#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N)))
+
+svint32_t
+omp_target_vla ()
+{
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_1_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_data_map_2_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target map(to: b, c) map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target map(to: b, c) map(tofrom: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_enter_exit_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target map(to: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+
+  return va;
+}
+
+svint32_t
+omp_target_map_data_alloc_update_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+{
+#pragma omp target
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target map(from: va) /* { dg-error {SVE type 'svint32_t' not allowed in map clause} } */
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+	vc = svld1_s32 (svptrue_b32 (), c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+}
+  return va;
+}
+
+int64_t
+omp_target_private_vla ()
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target private (va, vb, vc) map (to: b, c) map (from: res) /* { dg-error {SVE type 'svint32_t' not allowed in target private clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vla (svbool_t vp)
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target firstprivate (vp) map (to: b, c) map (from: res)/* { dg-error {SVE type 'svbool_t' not allowed in target firstprivate clause} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b); /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+typedef svint32_t v8si FIXED_ATTR;
+typedef svbool_t v8bi FIXED_ATTR;
+
+v8si
+omp_target_vls ()
+{
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_1_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_data_map_2_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target map(to: b, c) map(from: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+    }
+
+#pragma omp target map(to: b, c) map(tofrom: va)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, va);
+      va = svadd_s32_z (svptrue_b32 (), vc, va);
+    }
+
+  return va;
+}
+
+v8si
+omp_target_map_data_enter_exit_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target enter data map(to: b, c)
+
+#pragma omp target map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+#pragma omp target map(to: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+
+#pragma omp target exit data map(from: va)
+
+  return va;
+}
+
+v8si
+omp_target_map_data_alloc_update_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data map(to: b, c) map(alloc: va)
+{
+#pragma omp target
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      }
+
+/* Update va on the host from target.  */
+#pragma omp target update from(va)
+
+#pragma omp target map(from: va)
+    for (i = 0; i < 8; i++)
+      {
+	vb = svld1_s32 (svptrue_b32 (), b);
+	vc = svld1_s32 (svptrue_b32 (), c);
+	va = svadd_s32_z (svptrue_b32 (), vb, va);
+	va = svadd_s32_z (svptrue_b32 (), vc, va);
+      }
+}
+  return va;
+}
+
+int64_t
+omp_target_private_vls ()
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target private (va, vb, vc) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (svptrue_b32 (), b);
+      vc = svld1_s32 (svptrue_b32 (), c);
+      va = svadd_s32_z (svptrue_b32 (), vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t
+omp_target_firstprivate_vls (v8bi vp)
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target firstprivate (vp) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b);
+      vc = svld1_s32 (vp, c);
+      va = svadd_s32_z (vp, vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/target-device.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/target-device.c
new file mode 100644
index 00000000000..a20129cb42b
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/target-device.c
@@ -0,0 +1,186 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define N __ARM_FEATURE_SVE_BITS
+#define FIXED_ATTR __attribute__ ((arm_sve_vector_bits (N)))
+
+typedef __SVInt32_t v8si FIXED_ATTR;
+typedef svbool_t v8bi FIXED_ATTR;
+
+int64_t __attribute__ ((noinline))
+omp_target_device_ptr_vls (v8bi vp, v8si *vptr)
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data use_device_ptr (vptr) map (to: b, c)
+#pragma omp target is_device_ptr (vptr) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = *vptr;
+      vc = svld1_s32 (vp, c);
+      va = svadd_s32_z (vp, vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t __attribute__ ((noinline))
+omp_target_device_addr_vls (v8bi vp, v8si *vptr)
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data use_device_addr (vb) map (to: b, c)
+#pragma omp target is_device_ptr (vptr) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = *vptr;
+      vc = svld1_s32 (vp, c);
+      va = svadd_s32_z (vp, vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t __attribute__ ((noinline))
+omp_target_has_device_addr_vls (v8bi vp, v8si *vptr)
+{
+
+  int a[N], b[N], c[N];
+  v8si va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data use_device_addr (vb) map (to: b, c)
+#pragma omp target has_device_addr (vb) map (to: b, c) map (from: res)
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b);
+      vc = svld1_s32 (vp, c);
+      va = svadd_s32_z (vp, vb, vc);
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t __attribute__ ((noinline))
+omp_target_device_ptr_vla (svbool_t vp, svint32_t *vptr)
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data use_device_ptr (vptr) map (to: b, c) /* { dg-error {SVE type 'svint32_t \*' not allowed in target device clauses} } */
+#pragma omp target is_device_ptr (vptr) map (to: b, c) map (from: res) /* { dg-error {SVE type 'svint32_t \*' not allowed in target device clauses} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = *vptr; /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+			      /* { dg-error {'vp' referenced in target region does not have a mappable type} "" { target *-*-* } .-1 } */
+      va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t __attribute__ ((noinline))
+omp_target_device_addr_vla (svbool_t vp, svint32_t *vptr)
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data use_device_addr (vb) map (to: b, c) /* { dg-error {SVE type 'svint32_t' not allowed in target device clauses} } */
+#pragma omp target is_device_ptr (vptr) map (to: b, c) map (from: res) /* { dg-error {SVE type 'svint32_t \*' not allowed in target device clauses} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = *vptr; /* { dg-error {'vb' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+			      /* { dg-error {'vp' referenced in target region does not have a mappable type} "" { target *-*-* } .-1 } */
+      va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
+
+int64_t __attribute__ ((noinline))
+omp_target_has_device_addr_vla (svbool_t vp, svint32_t *vptr)
+{
+
+  int a[N], b[N], c[N];
+  svint32_t va, vb, vc;
+  int64_t res;
+  int i;
+
+#pragma omp parallel for
+  for (i = 0; i < N; i++)
+    {
+      b[i] = i;
+      c[i] = i + 1;
+    }
+
+#pragma omp target data use_device_addr (vb) map (to: b, c) /* { dg-error {SVE type 'svint32_t' not allowed in target device clauses} } */
+#pragma omp target has_device_addr (vb) map (to: b, c) map (from: res) /* { dg-error {SVE type 'svint32_t' not allowed in target device clauses} } */
+  for (i = 0; i < 8; i++)
+    {
+      vb = svld1_s32 (vp, b); /* { dg-error {'vp' referenced in target region does not have a mappable type} } */
+      vc = svld1_s32 (vp, c); /* { dg-error {'vc' referenced in target region does not have a mappable type} } */
+      va = svadd_s32_z (vp, vb, vc); /* { dg-error {'va' referenced in target region does not have a mappable type} } */
+      res = svaddv_s32 (svptrue_b32 (), va);
+    }
+
+  return res;
+}
diff --git a/gcc/testsuite/gcc.target/aarch64/sve/omp/target-link.c b/gcc/testsuite/gcc.target/aarch64/sve/omp/target-link.c
new file mode 100644
index 00000000000..afd9cf4fb05
--- /dev/null
+++ b/gcc/testsuite/gcc.target/aarch64/sve/omp/target-link.c
@@ -0,0 +1,54 @@ 
+/* { dg-do compile } */
+/* { dg-options "-msve-vector-bits=256 -std=gnu99 -fopenmp -O2" } */
+
+#include <arm_sve.h>
+
+#define N __ARM_FEATURE_SVE_BITS
+#define FIXED_ATTR __attribute__((arm_sve_vector_bits (N)))
+
+typedef __SVInt32_t v8si FIXED_ATTR;
+
+static v8si local_vec;
+#pragma omp declare target link(local_vec)
+
+v8si global_vec;
+#pragma omp declare target link(global_vec)
+
+static svint32_t slocal_vec; /* { dg-error {SVE type 'svint32_t' does not have a fixed size} }  */
+#pragma omp declare target link(slocal_vec) /* { dg-error {'slocal_vec' does not have a mappable type in 'link' clause} }  */
+
+void
+one_get_inc2_local_vec_vls ()
+{
+  v8si res, res2, tmp;
+
+#pragma omp target map(from: res, res2)
+  {
+    res = local_vec;
+    local_vec = svadd_s32_z (svptrue_b32 (), local_vec, local_vec);
+    res2 = local_vec;
+  }
+
+  tmp = svadd_s32_z (svptrue_b32 (), res, res);
+  svbool_t p = svcmpne_s32 (svptrue_b32 (), tmp, res2);
+  if (svptest_any (svptrue_b32 (), p))
+    __builtin_abort ();
+}
+
+void
+one_get_inc3_global_vec_vls ()
+{
+  v8si res, res2, tmp;
+
+#pragma omp target map(from: res, res2)
+  {
+    res = global_vec;
+    global_vec = svadd_s32_z (svptrue_b32 (), global_vec, global_vec);
+    res2 = global_vec;
+  }
+
+  tmp = svadd_s32_z (svptrue_b32 (), res, res);
+  svbool_t p = svcmpne_s32 (svptrue_b32 (), tmp, res2);
+  if (svptest_any (svptrue_b32 (), p))
+    __builtin_abort ();
+}