OpenE2K/gcc
2
0
Fork 0
Code Issues Pull Requests Projects Releases Wiki Activity

Missing changes from "Adjust copy/copyin/copyout/create for OpenACC 2.5"

Browse Source 
Most of that patch's changes were already committed as part of r261813 "Update
OpenACC data clause semantics to the 2.5 behavior", but not all of them.

	libgomp/
	* oacc-mem.c (acc_present_or_create): Remove definition and change
	to alias of acc_create.
	(acc_present_or_copyin): Remove definition and change to alias of
	acc_copyin.
	* oacc-parallel.c (GOACC_enter_exit_data): Call acc_create instead
	of acc_present_or_create.
	* testsuite/libgomp.oacc-c-c++-common/data-already-1.c: Remove.
	* testsuite/libgomp.oacc-c-c++-common/data-already-2.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-3.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-4.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-5.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-6.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-7.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-8.c: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-1.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-2.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-3.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-4.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-5.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-6.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-7.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-8.f: Likewise.

Co-Authored-By: Chung-Lin Tang <cltang@codesourcery.com>

From-SVN: r267153
This commit is contained in:
Thomas Schwinge 2018-12-14 21:43:12 +01:00 committed by Thomas Schwinge
parent f847198ec3
commit c759830b29
19 changed files with 51 additions and 297 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

26
libgomp/ChangeLog
View File

@ -1,3 +1,29 @@
2018-12-14 Thomas Schwinge <thomas@codesourcery.com>
Chung-Lin Tang <cltang@codesourcery.com>
* oacc-mem.c (acc_present_or_create): Remove definition and change
to alias of acc_create.
(acc_present_or_copyin): Remove definition and change to alias of
acc_copyin.
* oacc-parallel.c (GOACC_enter_exit_data): Call acc_create instead
of acc_present_or_create.
* testsuite/libgomp.oacc-c-c++-common/data-already-1.c: Remove.
* testsuite/libgomp.oacc-c-c++-common/data-already-2.c: Likewise.
* testsuite/libgomp.oacc-c-c++-common/data-already-3.c: Likewise.
* testsuite/libgomp.oacc-c-c++-common/data-already-4.c: Likewise.
* testsuite/libgomp.oacc-c-c++-common/data-already-5.c: Likewise.
* testsuite/libgomp.oacc-c-c++-common/data-already-6.c: Likewise.
* testsuite/libgomp.oacc-c-c++-common/data-already-7.c: Likewise.
* testsuite/libgomp.oacc-c-c++-common/data-already-8.c: Likewise.
* testsuite/libgomp.oacc-fortran/data-already-1.f: Likewise.
* testsuite/libgomp.oacc-fortran/data-already-2.f: Likewise.
* testsuite/libgomp.oacc-fortran/data-already-3.f: Likewise.
* testsuite/libgomp.oacc-fortran/data-already-4.f: Likewise.
* testsuite/libgomp.oacc-fortran/data-already-5.f: Likewise.
* testsuite/libgomp.oacc-fortran/data-already-6.f: Likewise.
* testsuite/libgomp.oacc-fortran/data-already-7.f: Likewise.
* testsuite/libgomp.oacc-fortran/data-already-8.f: Likewise.
2018-12-14 Thomas Schwinge <thomas@codesourcery.com>
PR libgomp/88495

47
libgomp/oacc-mem.c
View File

@ -544,6 +544,25 @@ acc_create_async (void *h, size_t s, int async)
present_create_copy (FLAG_PRESENT | FLAG_CREATE, h, s, async);
}
/* acc_present_or_create used to be what acc_create is now. */
/* acc_pcreate is acc_present_or_create by a different name. */
#ifdef HAVE_ATTRIBUTE_ALIAS
strong_alias (acc_create, acc_present_or_create)
strong_alias (acc_create, acc_pcreate)
#else
void *
acc_present_or_create (void *h, size_t s)
{
return acc_create (h, s);
}
void *
acc_pcreate (void *h, size_t s)
{
return acc_create (h, s);
}
#endif
void *
acc_copyin (void *h, size_t s)
{
@ -557,38 +576,22 @@ acc_copyin_async (void *h, size_t s, int async)
present_create_copy (FLAG_PRESENT | FLAG_CREATE | FLAG_COPY, h, s, async);
}
void *
acc_present_or_create (void *h, size_t s)
{
return present_create_copy (FLAG_PRESENT | FLAG_CREATE, h, s, acc_async_sync);
}
/* acc_pcreate is acc_present_or_create by a different name. */
/* acc_present_or_copyin used to be what acc_copyin is now. */
/* acc_pcopyin is acc_present_or_copyin by a different name. */
#ifdef HAVE_ATTRIBUTE_ALIAS
strong_alias (acc_present_or_create, acc_pcreate)
strong_alias (acc_copyin, acc_present_or_copyin)
strong_alias (acc_copyin, acc_pcopyin)
#else
void *
acc_pcreate (void *h, size_t s)
{
return acc_present_or_create (h, s);
}
#endif
void *
acc_present_or_copyin (void *h, size_t s)
{
return present_create_copy (FLAG_PRESENT | FLAG_CREATE | FLAG_COPY, h, s,
acc_async_sync);
return acc_copyin (h, s);
}
/* acc_pcopyin is acc_present_or_copyin by a different name. */
#ifdef HAVE_ATTRIBUTE_ALIAS
strong_alias (acc_present_or_copyin, acc_pcopyin)
#else
void *
acc_pcopyin (void *h, size_t s)
{
return acc_present_or_copyin (h, s);
return acc_copyin (h, s);
}
#endif

4
libgomp/oacc-parallel.c
View File

@ -425,14 +425,10 @@ GOACC_enter_exit_data (int device, size_t mapnum,
switch (kind)
{
case GOMP_MAP_ALLOC:
acc_present_or_create (hostaddrs[i], sizes[i]);
break;
case GOMP_MAP_FORCE_ALLOC:
acc_create (hostaddrs[i], sizes[i]);
break;
case GOMP_MAP_TO:
acc_present_or_copyin (hostaddrs[i], sizes[i]);
break;
case GOMP_MAP_FORCE_TO:
acc_copyin (hostaddrs[i], sizes[i]);
break;

20
libgomp/testsuite/libgomp.oacc-c-c++-common/data-already-1.c
View File

@ -1,20 +0,0 @@
/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */
#include <stdio.h>
#include <openacc.h>
int
main (int argc, char *argv[])
{
int i;
acc_copyin (&i, sizeof i);
fprintf (stderr, "CheCKpOInT\n");
#pragma acc data copy (i)
++i;
return 0;
}
/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */

20
libgomp/testsuite/libgomp.oacc-c-c++-common/data-already-2.c
View File

@ -1,20 +0,0 @@
/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */
#include <stdio.h>
int
main (int argc, char *argv[])
{
int i;
#pragma acc data present_or_copy (i)
{
fprintf (stderr, "CheCKpOInT\n");
#pragma acc data copyout (i)
++i;
}
return 0;
}
/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */

20
libgomp/testsuite/libgomp.oacc-c-c++-common/data-already-3.c
View File

@ -1,20 +0,0 @@
/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */
#include <stdio.h>
#include <openacc.h>
int
main (int argc, char *argv[])
{
int i;
#pragma acc data present_or_copy (i)
{
fprintf (stderr, "CheCKpOInT\n");
acc_copyin (&i, sizeof i);
}
return 0;
}
/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */

18
libgomp/testsuite/libgomp.oacc-c-c++-common/data-already-4.c
View File

@ -1,18 +0,0 @@
/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */
#include <stdio.h>
#include <openacc.h>
int
main (int argc, char *argv[])
{
int i;
acc_present_or_copyin (&i, sizeof i);
fprintf (stderr, "CheCKpOInT\n");
acc_copyin (&i, sizeof i);
return 0;
}
/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */

18
libgomp/testsuite/libgomp.oacc-c-c++-common/data-already-5.c
View File

@ -1,18 +0,0 @@
/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */
#include <stdio.h>
#include <openacc.h>
int
main (int argc, char *argv[])
{
int i;
#pragma acc enter data create (i)
fprintf (stderr, "CheCKpOInT\n");
acc_copyin (&i, sizeof i);
return 0;
}
/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */

18
libgomp/testsuite/libgomp.oacc-c-c++-common/data-already-6.c
View File

@ -1,18 +0,0 @@
/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */
#include <stdio.h>
#include <openacc.h>
int
main (int argc, char *argv[])
{
int i;
acc_present_or_copyin (&i, sizeof i);
fprintf (stderr, "CheCKpOInT\n");
#pragma acc enter data create (i)
return 0;
}
/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */

18
libgomp/testsuite/libgomp.oacc-c-c++-common/data-already-7.c
View File

@ -1,18 +0,0 @@
/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */
#include <stdio.h>
#include <openacc.h>
int
main (int argc, char *argv[])
{
int i;
#pragma acc enter data create (i)
fprintf (stderr, "CheCKpOInT\n");
acc_create (&i, sizeof i);
return 0;
}
/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */

20
libgomp/testsuite/libgomp.oacc-c-c++-common/data-already-8.c
View File

@ -1,20 +0,0 @@
/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */
#include <stdio.h>
int
main (int argc, char *argv[])
{
int i;
#pragma acc data create (i)
{
fprintf (stderr, "CheCKpOInT\n");
#pragma acc parallel copyin (i)
++i;
}
return 0;
}
/* { dg-output "CheCKpOInT(\n|\r\n|\r).*" } */

16
libgomp/testsuite/libgomp.oacc-fortran/data-already-1.f
View File

@ -1,16 +0,0 @@
! { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } }
IMPLICIT NONE
INCLUDE "openacc_lib.h"
INTEGER I
CALL ACC_COPYIN (I)
WRITE(0, *) "CheCKpOInT"
!$ACC DATA COPY (I)
I = 0
!$ACC END DATA
END
! { dg-output "CheCKpOInT(\n|\r\n|\r).*" }

16
libgomp/testsuite/libgomp.oacc-fortran/data-already-2.f
View File

@ -1,16 +0,0 @@
! { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } }
IMPLICIT NONE
INTEGER I
!$ACC DATA PRESENT_OR_COPY (I)
WRITE(0, *) "CheCKpOInT"
!$ACC DATA COPYOUT (I)
I = 0
!$ACC END DATA
!$ACC END DATA
END
! { dg-output "CheCKpOInT(\n|\r\n|\r).*" }

15
libgomp/testsuite/libgomp.oacc-fortran/data-already-3.f
View File

@ -1,15 +0,0 @@
! { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } }
IMPLICIT NONE
INCLUDE "openacc_lib.h"
INTEGER I
!$ACC DATA PRESENT_OR_COPY (I)
WRITE(0, *) "CheCKpOInT"
CALL ACC_COPYIN (I)
!$ACC END DATA
END
! { dg-output "CheCKpOInT(\n|\r\n|\r).*" }

14
libgomp/testsuite/libgomp.oacc-fortran/data-already-4.f
View File

@ -1,14 +0,0 @@
! { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } }
IMPLICIT NONE
INCLUDE "openacc_lib.h"
INTEGER I
CALL ACC_PRESENT_OR_COPYIN (I)
WRITE(0, *) "CheCKpOInT"
CALL ACC_COPYIN (I)
END
! { dg-output "CheCKpOInT(\n|\r\n|\r).*" }

14
libgomp/testsuite/libgomp.oacc-fortran/data-already-5.f
View File

@ -1,14 +0,0 @@
! { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } }
IMPLICIT NONE
INCLUDE "openacc_lib.h"
INTEGER I
!$ACC ENTER DATA CREATE (I)
WRITE(0, *) "CheCKpOInT"
CALL ACC_COPYIN (I)
END
! { dg-output "CheCKpOInT(\n|\r\n|\r).*" }

14
libgomp/testsuite/libgomp.oacc-fortran/data-already-6.f
View File

@ -1,14 +0,0 @@
! { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } }
IMPLICIT NONE
INCLUDE "openacc_lib.h"
INTEGER I
CALL ACC_PRESENT_OR_COPYIN (I)
WRITE(0, *) "CheCKpOInT"
!$ACC ENTER DATA CREATE (I)
END
! { dg-output "CheCKpOInT(\n|\r\n|\r).*" }

14
libgomp/testsuite/libgomp.oacc-fortran/data-already-7.f
View File

@ -1,14 +0,0 @@
! { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } }
IMPLICIT NONE
INCLUDE "openacc_lib.h"
INTEGER I
!$ACC ENTER DATA CREATE (I)
WRITE(0, *) "CheCKpOInT"
CALL ACC_CREATE (I)
END
! { dg-output "CheCKpOInT(\n|\r\n|\r).*" }

16
libgomp/testsuite/libgomp.oacc-fortran/data-already-8.f
View File

@ -1,16 +0,0 @@
! { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } }
IMPLICIT NONE
INTEGER I
!$ACC DATA CREATE (I)
WRITE(0, *) "CheCKpOInT"
!$ACC PARALLEL COPYIN (I)
I = 0
!$ACC END PARALLEL
!$ACC END DATA
END
! { dg-output "CheCKpOInT(\n|\r\n|\r).*" }