Hi! On Tue, 31 Jul 2018 08:12:51 -0700, Cesar Philippidis <cesar_philippi...@mentor.com> wrote: > This is an old patch which removes the struct map from the nvptx plugin.
(This got committed to trunk in r263212.) > I believe at one point this was supposed to be used to manage async data > mappings, but in practice that never worked out. The original submission seems to be <http://mid.mail-archive.com/566EC830.8030508@codesourcery.com>: | The attached patch fixes an issue in the managing of | the page-locked buffer which holds the kernel launch | mappings. In the process of fixing the issue, I discovered | that 'struct map' was no longer needed, so that has | been removed as well. I can't tell/remember what the "issue in the managing of the page-locked buffer which holds the kernel launch mappings" would've been. > --- /dev/null > +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/mapping-1.c I wanted to note that on all branches down to GCC 5 this test case PASSes already without any libgomp code changes. > @@ -0,0 +1,63 @@ > +/* { dg-do run } */ > + > +#include <stdio.h> > +#include <stdlib.h> > +#include <unistd.h> > + > +/* Exercise the kernel launch argument mapping. */ > + > +int > +main (int argc, char **argv) > +{ > + int a[256], b[256], c[256], d[256], e[256], f[256]; > + int i; > + int n; > + > + /* 48 is the size of the mappings for the first parallel construct. */ > + n = sysconf (_SC_PAGESIZE) / 48 - 1; The rationale for the number 48 doesn't seem right -- the first 'parallel' construct has four mappings, so four times eight bytes is 32 bytes on x86_64? Ha, or maybe this applies to the 'struct map', so it's these four plus the 'int async', 'size_t size'? Anyway, I can't tell what this is trying to achieve? > + > + i = 0; > + > + for (i = 0; i < n; i++) > + { > + #pragma acc parallel copy (a, b, c, d) > + { > + int j; > + > + for (j = 0; j < 256; j++) > + { > + a[j] = j; > + b[j] = j; > + c[j] = j; > + d[j] = j; > + } > + } > + } Maybe filling up some improperly-managed statically-sized data structure (said "page-locked buffer"?), which then... > + > +#pragma acc parallel copy (a, b, c, d, e, f) ... would overflow here? > + { > + int j; > + > + for (j = 0; j < 256; j++) > + { > + a[j] = j; > + b[j] = j; > + c[j] = j; > + d[j] = j; > + e[j] = j; > + f[j] = j; > + } > + } > + > + for (i = 0; i < 256; i++) > + { > + if (a[i] != i) abort(); > + if (b[i] != i) abort(); > + if (c[i] != i) abort(); > + if (d[i] != i) abort(); > + if (e[i] != i) abort(); > + if (f[i] != i) abort(); > + } > + > + exit (0); > +} Anyway -- the libgomp code has been cleaned up; the test case seems like it can be disregarded. Grüße Thomas
signature.asc
Description: PGP signature