[OpenACC] Implement 'routine' construct parsing

clang/include/clang/Basic/DiagnosticParseKinds.td

@@ -1364,6 +1364,8 @@ def warn_pragma_acc_unimplemented_clause_parsing
 def err_acc_invalid_directive
     : Error<"invalid OpenACC directive '%select{%1|%1 %2}0'">;
 def err_acc_missing_directive : Error<"expected OpenACC directive">;
+def err_acc_invalid_open_paren
+    : Error<"expected clause-list or newline in OpenACC directive">;
 
 // OpenMP support.
 def warn_pragma_omp_ignored : Warning<

clang/include/clang/Basic/OpenACCKinds.h

@@ -55,7 +55,8 @@ enum class OpenACCDirectiveKind {
   Update,
   // FIXME: wait construct.
 
-  // FIXME: routine construct.
+  // Procedure Calls in Compute Regions.
+  Routine,
 
   // Invalid.
   Invalid,

clang/include/clang/Parse/Parser.h

@@ -3532,9 +3532,14 @@ class Parser : public CodeCompletionHandler {
   /// Placeholder for now, should just ignore the directives after emitting a
   /// diagnostic. Eventually will be split into a few functions to parse
   /// different situations.
+public:
   DeclGroupPtrTy ParseOpenACCDirectiveDecl();
   StmtResult ParseOpenACCDirectiveStmt();
 
+private:
+  void ParseOpenACCDirective();
+  ExprResult ParseOpenACCRoutineName();
+
 private:
   //===--------------------------------------------------------------------===//
   // C++ 14: Templates [temp]

clang/lib/Parse/ParseOpenACC.cpp

@@ -46,6 +46,7 @@ OpenACCDirectiveKindEx getOpenACCDirectiveKind(StringRef Name) {
           .Case("host_data", OpenACCDirectiveKind::HostData)
           .Case("loop", OpenACCDirectiveKind::Loop)
           .Case("atomic", OpenACCDirectiveKind::Atomic)
+          .Case("routine", OpenACCDirectiveKind::Routine)
           .Case("declare", OpenACCDirectiveKind::Declare)
           .Case("init", OpenACCDirectiveKind::Init)
           .Case("shutdown", OpenACCDirectiveKind::Shutdown)
@@ -97,6 +98,8 @@ bool isOpenACCDirectiveKind(OpenACCDirectiveKind Kind, StringRef Tok) {
 
   case OpenACCDirectiveKind::Atomic:
     return Tok == "atomic";
+  case OpenACCDirectiveKind::Routine:
+    return Tok == "routine";
   case OpenACCDirectiveKind::Declare:
     return Tok == "declare";
   case OpenACCDirectiveKind::Init:
@@ -232,32 +235,87 @@ void ParseOpenACCClauseList(Parser &P) {
     P.Diag(P.getCurToken(), diag::warn_pragma_acc_unimplemented_clause_parsing);
 }
 
-void ParseOpenACCDirective(Parser &P) {
-  OpenACCDirectiveKind DirKind = ParseOpenACCDirectiveKind(P);
+} // namespace
+
+// Routine has an optional paren-wrapped name of a function in the local scope.
+// We parse the name, emitting any diagnostics
+ExprResult Parser::ParseOpenACCRoutineName() {
+
+  ExprResult Res;
+  if (getLangOpts().CPlusPlus) {
+    Res = ParseCXXIdExpression(/*isAddressOfOperand=*/false);
+  } else {
+    // There isn't anything quite the same as ParseCXXIdExpression for C, so we
+    // need to get the identifier, then call into Sema ourselves.
+
+    if (expectIdentifier())
+      return ExprError();
+
+    Token FuncName = getCurToken();
+    UnqualifiedId Name;
+    CXXScopeSpec ScopeSpec;
+    SourceLocation TemplateKWLoc;
+    Name.setIdentifier(FuncName.getIdentifierInfo(), ConsumeToken());
+
+    // Ensure this is a valid identifier. We don't accept causing implicit
+    // function declarations per the spec, so always claim to not have trailing
+    // L Paren.
+    Res = Actions.ActOnIdExpression(getCurScope(), ScopeSpec, TemplateKWLoc,
+                                    Name, /*HasTrailingLParen=*/false,
+                                    /*isAddressOfOperand=*/false);
+  }
+
+  return getActions().CorrectDelayedTyposInExpr(Res);
+}
+
+void Parser::ParseOpenACCDirective() {
+  OpenACCDirectiveKind DirKind = ParseOpenACCDirectiveKind(*this);
 
   // Once we've parsed the construct/directive name, some have additional
   // specifiers that need to be taken care of. Atomic has an 'atomic-clause'
   // that needs to be parsed.
   if (DirKind == OpenACCDirectiveKind::Atomic)
-    ParseOpenACCAtomicKind(P);
+    ParseOpenACCAtomicKind(*this);
+
+  // We've successfully parsed the construct/directive name, however a few of
+  // the constructs have optional parens that contain further details.
+  BalancedDelimiterTracker T(*this, tok::l_paren,
+                             tok::annot_pragma_openacc_end);
+
+  if (!T.consumeOpen()) {
+    switch (DirKind) {
+    default:
+      Diag(T.getOpenLocation(), diag::err_acc_invalid_open_paren);
+      T.skipToEnd();
+      break;
+    case OpenACCDirectiveKind::Routine: {
+      ExprResult RoutineName = ParseOpenACCRoutineName();
+      // If the routine name is invalid, just skip until the closing paren to
+      // recover more gracefully.
+      if (RoutineName.isInvalid())
+        T.skipToEnd();
+      else
+        T.consumeClose();
+      break;
+    }
+    }
+  }
 
   // Parses the list of clauses, if present.
-  ParseOpenACCClauseList(P);
+  ParseOpenACCClauseList(*this);
 
-  P.Diag(P.getCurToken(), diag::warn_pragma_acc_unimplemented);
-  P.SkipUntil(tok::annot_pragma_openacc_end);
+  Diag(getCurToken(), diag::warn_pragma_acc_unimplemented);
+  SkipUntil(tok::annot_pragma_openacc_end);
 }
 
-} // namespace
-
 // Parse OpenACC directive on a declaration.
 Parser::DeclGroupPtrTy Parser::ParseOpenACCDirectiveDecl() {
   assert(Tok.is(tok::annot_pragma_openacc) && "expected OpenACC Start Token");
 
   ParsingOpenACCDirectiveRAII DirScope(*this);
   ConsumeAnnotationToken();
 
-  ParseOpenACCDirective(*this);
+  ParseOpenACCDirective();
 
   return nullptr;
 }
@@ -269,7 +327,7 @@ StmtResult Parser::ParseOpenACCDirectiveStmt() {
   ParsingOpenACCDirectiveRAII DirScope(*this);
   ConsumeAnnotationToken();
 
-  ParseOpenACCDirective(*this);
+  ParseOpenACCDirective();
 
   return StmtEmpty();
 }

clang/test/ParserOpenACC/parse-constructs.c

@@ -16,9 +16,16 @@ void func() {
   // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
 #pragma acc parallel clause list
   for(;;){}
+  // expected-error@+3{{expected clause-list or newline in OpenACC directive}}
   // expected-warning@+2{{OpenACC clause parsing not yet implemented}}
   // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
 #pragma acc parallel() clause list
+  for(;;){}
+  // expected-error@+4{{expected clause-list or newline in OpenACC directive}}
+  // expected-error@+3{{expected ')'}}
+  // expected-note@+2{{to match this '('}}
+  // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc parallel( clause list
   for(;;){}
   // expected-warning@+2{{OpenACC clause parsing not yet implemented}}
   // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
@@ -144,3 +151,38 @@ void func() {
 #pragma acc update clause list
   for(;;){}
 }
+
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine
+void routine_func();
+// expected-warning@+2{{OpenACC clause parsing not yet implemented}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine clause list
+void routine_func();
+
+// expected-error@+2{{use of undeclared identifier 'func_name'}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine (func_name)
+// expected-error@+3{{use of undeclared identifier 'func_name'}}
+// expected-warning@+2{{OpenACC clause parsing not yet implemented}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine (func_name) clause list
+
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine (routine_func)
+// expected-warning@+2{{OpenACC clause parsing not yet implemented}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine (routine_func) clause list
+
+// expected-error@+3{{expected ')'}}
+// expected-note@+2{{to match this '('}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine (routine_func())
+
+// expected-error@+2{{expected identifier}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine()
+
+// expected-error@+2{{expected identifier}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(int)

clang/test/ParserOpenACC/parse-constructs.cpp

@@ -0,0 +1,46 @@
+// RUN: %clang_cc1 %s -verify -fopenacc
+
+namespace NS {
+  void foo(); // expected-note{{declared here}}
+
+  template<typename T>
+  void templ(); // expected-note 2{{declared here}}
+}
+
+// expected-error@+2{{use of undeclared identifier 'foo'; did you mean 'NS::foo'?}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(foo)
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(NS::foo)
+
+// expected-error@+2{{use of undeclared identifier 'templ'; did you mean 'NS::templ'?}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(templ)
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(NS::templ)
+
+// expected-error@+2{{use of undeclared identifier 'templ'; did you mean 'NS::templ'?}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(templ<int>)
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(NS::templ<int>)
+
+// expected-error@+2{{use of undeclared identifier 'T'}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(templ<T>)
+// expected-error@+2{{use of undeclared identifier 'T'}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(NS::templ<T>)
+
+// expected-error@+3{{expected ')'}}
+// expected-note@+2{{to match this '('}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine (NS::foo())
+
+// expected-error@+2 {{expected unqualified-id}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine()
+
+// expected-error@+2 {{expected unqualified-id}}
+// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc routine(int)