aboutsummaryrefslogtreecommitdiff
path: root/offload/test/offloading/target_update_strided_struct_from.c
blob: 7a8805ab35e98164bcaebec5fdbd57265e543f33 (plain)
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
// RUN: %libomptarget-compile-run-and-check-generic
// XFAIL: intelgpu
// This test checks that "update from" with user-defined mapper supports strided
// sections using fixed-size arrays in structs.

#include <omp.h>
#include <stdio.h>
#include <stdlib.h>

#define N 16

typedef struct {
  double data[N];
  size_t len;
} T;

#pragma omp declare mapper(custom : T v) map(to : v, v.len, v.data[0 : v.len])

int main() {
  T s;
  s.len = N;

  for (int i = 0; i < N; i++) {
    s.data[i] = i;
  }

  printf("original host array values:\n");
  for (int i = 0; i < N; i++)
    printf("%f\n", s.data[i]);
  printf("\n");

#pragma omp target data map(mapper(custom), to : s)
  {
// Execute on device with explicit mapper
#pragma omp target map(mapper(custom), tofrom : s)
    {
      for (int i = 0; i < s.len; i++) {
        s.data[i] += i;
      }
    }

// Update strided elements from device: indices 0,2,4,6,8,10,12,14
#pragma omp target update from(s.data[0 : 8 : 2])
  }

  printf("from target array results:\n");
  for (int i = 0; i < N; i++)
    printf("%f\n", s.data[i]);

  // CHECK: original host array values:
  // CHECK-NEXT: 0.000000
  // CHECK-NEXT: 1.000000
  // CHECK-NEXT: 2.000000
  // CHECK-NEXT: 3.000000
  // CHECK-NEXT: 4.000000
  // CHECK-NEXT: 5.000000
  // CHECK-NEXT: 6.000000
  // CHECK-NEXT: 7.000000
  // CHECK-NEXT: 8.000000
  // CHECK-NEXT: 9.000000
  // CHECK-NEXT: 10.000000
  // CHECK-NEXT: 11.000000
  // CHECK-NEXT: 12.000000
  // CHECK-NEXT: 13.000000
  // CHECK-NEXT: 14.000000
  // CHECK-NEXT: 15.000000

  // CHECK: from target array results:
  // CHECK-NEXT: 0.000000
  // CHECK-NEXT: 1.000000
  // CHECK-NEXT: 4.000000
  // CHECK-NEXT: 3.000000
  // CHECK-NEXT: 8.000000
  // CHECK-NEXT: 5.000000
  // CHECK-NEXT: 12.000000
  // CHECK-NEXT: 7.000000
  // CHECK-NEXT: 16.000000
  // CHECK-NEXT: 9.000000
  // CHECK-NEXT: 20.000000
  // CHECK-NEXT: 11.000000
  // CHECK-NEXT: 24.000000
  // CHECK-NEXT: 13.000000
  // CHECK-NEXT: 28.000000
  // CHECK-NEXT: 15.000000

  return 0;
}