Hi Sandra,
I had a first glance at your patch, albeit very superficial. I found one
issue reading the code - and observed some issues when running it against
some existing external tests.
I will have a deeper looker later, will now first do another iteration
on your main/middle-end patch #2, which I still have to finish studying.
But now comments to this patch - or C testing as at least one comment
relates to code in Patch #2.
Sandra Loosemore wrote:
Additional shared C/C++ testcases are included in a subsequent patch in this
series.
I started by playing around (well, just running) some existing testcases.
It seems as if the OpenMP example document found cases which aren't handled.
I think those should eventually work - but it can be deferred to a follow-up
patch; and probably they should be handled as follow up. (We might consider
collecting those in a PR / some PRs, if we won't fix them soon.)
Only tested with C, but C++ and Fortran are likely affected by most of
them alike (even if each FE has its own implementation).
* * *
A case where 'omp error' diagnostic should be delayed - and (here) suppressed:
program_control/sources/error.1.c:15:23: error: ‘pragma omp error’ encountered:
GNU compiler required.
15 | otherwise(error at(compilation) severity(fatal) \
| ^~~~~
which is odd given that we have the GNU compiler:
#pragma omp metadirective \
when(implementation={vendor(gnu)}: nothing ) \
otherwise(error at(compilation) severity(fatal) \
message("GNU compiler required."))
https://github.com/OpenMP/Examples/blob/main/program_control/sources/error.1.c
(Clang-19 handles this: with vendor(llvm) no error is printed.)
* * *
Here, it seems as if the 'target' + 'teams' without intervening code check has
to be updated:
program_control/sources/metadirective.1.c:16:12: error: ‘target’ construct with
nested ‘teams’ construct contains directives outside of the ‘teams’ construct
16 | #pragma omp target map(to:v1,v2) map(from:v3) device(0)
| ^~~
for
#pragma omp target map(to:v1,v2) map(from:v3) device(0)
#pragma omp metadirective \
when( device={arch("nvptx")}: teams loop) \
otherwise( parallel loop)
for (int i= 0; i< N; i++) v3[i] = v1[i] * v2[i];
https://github.com/OpenMP/Examples/blob/main/program_control/sources/metadirective.1.c
Same issue with one example in OpenMP_VV (formerly: sollve_vv):
https://github.com/OpenMP-Validation-and-Verification/OpenMP_VV/blob/master/tests/5.0/metadirective/test_metadirective_arch_is_nvidia.c
* * *
Regarding the following:
program_control/sources/metadirective.3.c:34:7: warning: direct calls to an
offloadable function containing metadirectives with a ‘construct={target}’
selector may produce unexpected results
https://github.com/OpenMP/Examples/blob/main/program_control/sources/metadirective.3.c
First, it seems as if it shouldn't be unconditionally printed but only with
-Wopenmp,
enabled by default. (i.e. change the 0 to OPT_Wopenmp in the warning_at call).
That's for:
#pragma omp begin declare target
void exp_pi_diff(double *d, double my_pi){
#pragma omp metadirective \
when( construct={target}: distribute parallel for ) \
otherwise( parallel for simd
I think the warning perfectly makes sense and the description in the example
document is either wrong or at least misleading, but for GCC, I am wondering
whether we should add the following, which might help the user:
inform (loc, "consider whether the context selector %<device={kind(nohost)}%> should
be used instead");
(needs to be implemented such that -Wno-openmp does not print it, e.g. if
(warning_at ...))
Disclaimer: That's not exactly the same for 'target device(omp_initial_device)'
as running it inside a target region with host fallback, the nteams-var and
nthreads-var ICV are probably different from running it outside of the host,
such that the choice between 'when' and 'otherwise' still might matter but
I bet that in most cases 'kind(nohost)' is what the user actually wants to
have.
On the example document side, I have filed the OpenMP Example Issue #483 as
the testcase seems to be either wrong or misleading, judging from the
testcase and its description.
* * *
Continuing with OpenMP_VV - I think the programmer meant 'device' and not
'target_device', but I think GCC overdoes the diagnostic, given that there
are only two 'target_device':
tests/5.1/metadirective/test_metadirective_target_device.c:30:15: sorry, unimplemented: ‘target_device’ selector set inside of ‘target’ directive
30 | #pragma omp metadirective \
| ^~~
tests/5.1/metadirective/test_metadirective_target_device.c:30:15: sorry,
unimplemented: ‘target_device’ selector set inside of ‘target’ directive
tests/5.1/metadirective/test_metadirective_target_device.c:30:15: sorry,
unimplemented: ‘target_device’ selector set inside of ‘target’ directive
tests/5.1/metadirective/test_metadirective_target_device.c:30:15: sorry,
unimplemented: ‘target_device’ selector set inside of ‘target’ directive
tests/5.1/metadirective/test_metadirective_target_device.c:30:15: warning:
‘target’ construct inside of ‘target’ region [-Wopenmp]
tests/5.1/metadirective/test_metadirective_target_device.c:30:15: warning:
‘target’ construct inside of ‘target’ region [-Wopenmp]
I wonder whether we could get rid of some of them. On the other hand, as they
are nicely follow each other, it is not a Prio 1 issue.
https://github.com/OpenMP-Validation-and-Verification/OpenMP_VV/blob/master/tests/5.1/metadirective/test_metadirective_target_device.c
* * *
But now back to the patch:
gcc/c-family/ChangeLog
PR middle-end/112779
PR middle-end/113904
* c-common.h (enum c_omp_directive_kind): Add C_OMP_DIR_META.
You also need to handle this in for the "absent" / "contains" clause,
which currently reads:
c_parser_omp_assumption_clauses (c_parser *parser, bool is_assume)
...
if (dir == NULL
|| dir->kind == C_OMP_DIR_DECLARATIVE
|| dir->kind == C_OMP_DIR_INFORMATIONAL
|| dir->id == PRAGMA_OMP_END
|| (!dir->second && directive[1])
|| (!dir->third && directive[2]))
error_at (dloc, "unknown OpenMP directive name in "
"%qs clause argument", p);
But I have to admit that I find the gfortran error more helpful:
if (kind == GFC_OMP_DIR_DECLARATIVE
|| kind == GFC_OMP_DIR_INFORMATIONAL
|| kind == GFC_OMP_DIR_META)
{
gfc_error ("Invalid %qs directive at %L in %s clause: declarative, "
"informational and meta directives not permitted",
gfc_ascii_statement (st, true), &old_loc,
is_absent ? "ABSENT" : "CONTAINS");
* * *
In any case, the following is now silently accepted:
void f() {
#pragma omp assume absent(metadirective)
{
int x = 5;
}
}
* * *
OpenMP 6 states in "10.6.1 assumption Clauses":
"A directive-name list item must not specify a directive
that is a declarative directive, an informational directive,
or a metadirective."
* * *
Thanks,
Tobias
