On 1/9/25 05:17, Tobias Burnus wrote:
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.)

Hmmm.  Here the 6.0 spec says:

"If action-time is compilation, the action is performed during compilation if the directive appears in a declarative context or in an executable context that is reachable at runtime. If action-time is compilation and the directive appears in an executable context that is not reachable at runtime, the action may or may not be performed."

IOW, the spec doesn't require implementations to solve the halting problem ;-) but we could theoretically defer the error to suppress cases where the directive is optimized away. How about I just file a PR for this for now? It's a problem with #pragma omp error, not metadirective, anyway, so independent of this patch series.

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

This diagnostic is coming from omp-low.cc rather than any of the front ends. By that time the metadirective has been gimplified into a conditional for late resolution. I'm not sure what the right thing to do here is; is there an actual problem with code generation if there is some expression to test a condition between the "target" and "teams"? This case clearly shows that pushing the outer "target" into each of the metadirective alternatives wouldn't work (because the test expression depends on it). File another PR for now?

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
[snip]

This part has already been addressed in the patches committed this morning.

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


This is my bad. It's coming from omp_context_selector_matches() in omp-general.cc, where it just continues processing after the "sorry" error instead of returning 0. Do you think it's correct for this to be a "sorry" instead of a warning, or just silently always failing to match? It's a trivial fix, anyway.

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."

I'll fix this before reposting a new version of the remaining patches in the series. The C++ code is very similar and needs the same fix.

Also, from offline discussion with Tobias:

BTW: The following seems to be valid but is rejected as the metadirective after 'for' is 
not handled: "error: loop nest expected before ‘#pragma’"
void f() {
  #pragma omp metadirective when(construct={target} : parallel)
    #pragma omp metadirective when(construct={target} : for)
      #pragma omp metadirective when(construct={target} : simd)
        for (int i =0 ; i < 1; i++)
          __builtin_printf("Hello world\n");
}
Unless that's simple to fix, I suggest to just queue it for a new PR.

It's not clear to me that this is supposed to be valid; metadirective is not included in the syntax for "canonical loop nest", nor is it a directive with the loop-transforming property. Anyway, I think this would require changes beyond just suppressing the error message, so yeah, some low-priority PR would be appropriate here.

-Sandra

Reply via email to