Unverified Commit ff219ea9 authored by Erich Keane's avatar Erich Keane Committed by GitHub
Browse files

[OpenACC] Initial commits to support OpenACC (#70234) 

Initial commits to support OpenACC.  This patchset:

adds a clang-command line argument '-fopenacc', and starts
 to define _OPENACC, albeit to '1' instead of the standardized
value (since we don't properly implement OpenACC yet).

The OpenACC spec defines `_OPENACC` to be equal to the latest standard
implemented. However, since we're not done implementing any standard,
we've defined this by default to be `1`. As it is useful to run our
compiler against existing OpenACC workloads, we're providing a
temporary override flag to change the `_OPENACC` value to be any
entirely digit value, permitting testing against any existing OpenACC
project.

Exactly like the OpenMP parser, the OpenACC pragma parser needs to
consume and reprocess the tokens. This patch sets up the infrastructure
to do so by refactoring the OpenMP version of this into a more general
version that works for OpenACC as well.

Additionally, this adds a few diagnostics and token kinds to get us
started.
parent 9c0e6499

clang/docs/ReleaseNotes.rst

+13 −0
Original line number Diff line number Diff line
@@ -241,6 +241,8 @@ New Compiler Flags 
  handlers will be smaller. A throw expression of a type with a

  potentially-throwing destructor will lead to an error.



* ``-fopenacc`` was added as a part of the effort to support OpenACC in clang.



Deprecated Compiler Flags

-------------------------
@@ -731,6 +733,17 @@ Miscellaneous Clang Crashes Fixed 
- Fixed a crash when ``-ast-dump=json`` was used for code using class

  template deduction guides.



OpenACC Specific Changes

------------------------

- OpenACC Implementation effort is beginning with semantic analysis and parsing

  of OpenACC pragmas. The ``-fopenacc`` flag was added to enable these new,

  albeit incomplete changes. The ``_OPENACC`` macro is currently defined to

  ``1``, as support is too incomplete to update to a standards-required value.

- Added ``-fexperimental-openacc-macro-override``, a command line option to

  permit overriding the ``_OPENACC`` macro to be any digit-only value specified

  by the user, which permits testing the compiler against existing OpenACC

  workloads in order to evaluate implementation progress.



Target Specific Changes

-----------------------

clang/include/clang/Basic/DiagnosticGroups.td

+4 −0
Original line number Diff line number Diff line
@@ -1315,6 +1315,10 @@ def OpenMP : DiagGroup<"openmp", [ 
    OpenMPMapping, OpenMP51Ext, OpenMPExtensions, OpenMPTargetException

  ]>;



// OpenACC warnings.

def SourceUsesOpenACC : DiagGroup<"source-uses-openacc">;

def OpenACC : DiagGroup<"openacc", [SourceUsesOpenACC]>;



// Backend warnings.

def BackendInlineAsm : DiagGroup<"inline-asm">;

def BackendSourceMgr : DiagGroup<"source-mgr">;

clang/include/clang/Basic/DiagnosticParseKinds.td

+9 −0
Original line number Diff line number Diff line
@@ -1342,6 +1342,15 @@ def err_opencl_logical_exclusive_or : Error< 
def err_openclcxx_virtual_function : Error<

  "virtual functions are not supported in C++ for OpenCL">;



// OpenACC Support.

def warn_pragma_acc_ignored : Warning<

  "unexpected '#pragma acc ...' in program">, InGroup<SourceUsesOpenACC>, DefaultIgnore;

def err_acc_unexpected_directive : Error<

  "unexpected OpenACC directive %select{|'#pragma acc %1'}0">;

def warn_pragma_acc_unimplemented

    : Warning<"OpenACC directives not yet implemented, pragma ignored">,

      InGroup<SourceUsesOpenACC>;



// OpenMP support.

def warn_pragma_omp_ignored : Warning<

  "unexpected '#pragma omp ...' in program">, InGroup<SourceUsesOpenMP>, DefaultIgnore;

clang/include/clang/Basic/LangOptions.def

+2 −0
Original line number Diff line number Diff line
@@ -285,6 +285,8 @@ LANGOPT(OffloadUniformBlock, 1, 0, "Assume that kernels are launched with unifor 
LANGOPT(HIPStdPar, 1, 0, "Enable Standard Parallel Algorithm Acceleration for HIP (experimental)")

LANGOPT(HIPStdParInterposeAlloc, 1, 0, "Replace allocations / deallocations with HIP RT calls when Standard Parallel Algorithm Acceleration for HIP is enabled (Experimental)")



LANGOPT(OpenACC           , 1, 0, "OpenACC Enabled")



LANGOPT(SizedDeallocation , 1, 0, "sized deallocation")

LANGOPT(AlignedAllocation , 1, 0, "aligned allocation")

LANGOPT(AlignedAllocationUnavailable, 1, 0, "aligned allocation functions are unavailable")

clang/include/clang/Basic/LangOptions.h

+5 −0
Original line number Diff line number Diff line
@@ -502,6 +502,11 @@ public: 
  // received as a result of a standard operator new (-fcheck-new)

  bool CheckNew = false;



  // In OpenACC mode, contains a user provided override for the _OPENACC macro.

  // This exists so that we can override the macro value and test our incomplete

  // implementation on real-world examples.

  std::string OpenACCMacroOverride;



  LangOptions();



  /// Set language defaults for the given input language and