blob: 275fa9ae256b60b2138998ca9646be14a1f865f6 (
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
|
#include <openacc.h>
#include <stdlib.h>
/* Test attach/detach operation with chained dereferences. */
typedef struct mystruct {
int *a;
struct mystruct *next;
} mystruct;
int
main (int argc, char* argv[])
{
const int N = 1024;
mystruct *m = (mystruct *) malloc (sizeof (*m));
int i;
m->a = (int *) malloc (N * sizeof (int));
m->next = (mystruct *) malloc (sizeof (*m));
m->next->a = (int *) malloc (N * sizeof (int));
m->next->next = NULL;
for (i = 0; i < N; i++)
{
m->a[i] = 0;
m->next->a[i] = 0;
}
#pragma acc enter data copyin(m[0:1])
acc_copyin (m->next, sizeof (*m));
for (int i = 0; i < 99; i++)
{
int j;
acc_copyin (m->next->a, N * sizeof (int));
acc_attach ((void **) &m->next);
/* This will attach only the innermost pointer, i.e. "a[0:N]". That's
why we have to attach the "m->next" pointer manually above. */
#pragma acc parallel loop copy(m->next->a[0:N])
for (j = 0; j < N; j++)
m->next->a[j]++;
acc_detach ((void **) &m->next);
acc_copyout (m->next->a, N * sizeof (int));
}
acc_copyout (m->next, sizeof (*m));
#pragma acc exit data copyout(m[0:1])
for (i = 0; i < N; i++)
{
if (m->a[i] != 0)
abort ();
if (m->next->a[i] != 99)
abort ();
}
free (m->next->a);
free (m->next);
free (m->a);
free (m);
return 0;
}
|
