[OpenACC] Ensure decl OpenACC constructs don't crash

I initially implemented codegen to be a 'no-op' for these declarations,
  which I thought was properly implemented.  However, when they are a
  top-level decl, we have a separate switch.  This patch makes sure they
  are properly emitted at top-level as a no-op, and adds a test for both
  top-level and not top-level.

Author: erichkeane

clang/lib/CodeGen/CodeGenModule.cpp

@@ -7300,6 +7300,13 @@ void CodeGenModule::EmitTopLevelDecl(Decl *D) {
     getHLSLRuntime().addBuffer(cast<HLSLBufferDecl>(D));
     break;
 
+  case Decl::OpenACCDeclare:
+    EmitOpenACCDeclare(cast<OpenACCDeclareDecl>(D));
+    break;
+  case Decl::OpenACCRoutine:
+    EmitOpenACCRoutine(cast<OpenACCRoutineDecl>(D));
+    break;
+
   default:
     // Make sure we handled everything we should, every other kind is a
     // non-top-level decl.  FIXME: Would be nice to have an isTopLevelDeclKind

clang/lib/CodeGen/CodeGenModule.h

@@ -1574,9 +1574,11 @@ class CodeGenModule : public CodeGenTypeCache {
                             CodeGenFunction *CGF = nullptr);
 
   // Emit code for the OpenACC Declare declaration.
-  void EmitOpenACCDeclare(const OpenACCDeclareDecl *D, CodeGenFunction *CGF);
+  void EmitOpenACCDeclare(const OpenACCDeclareDecl *D,
+                          CodeGenFunction *CGF = nullptr);
   // Emit code for the OpenACC Routine declaration.
-  void EmitOpenACCRoutine(const OpenACCRoutineDecl *D, CodeGenFunction *CGF);
+  void EmitOpenACCRoutine(const OpenACCRoutineDecl *D,
+                          CodeGenFunction *CGF = nullptr);
 
   /// Emit a code for requires directive.
   /// \param D Requires declaration

clang/test/CodeGen/openacc-noop-decl.c

@@ -0,0 +1,15 @@
+// RUN: %clang_cc1 %s -emit-llvm -o - | FileCheck %s
+
+// Check to ensure these are no-ops/don't really do anything.
+void name();
+
+void foo() {
+#pragma acc declare
+  // CHECK-NOT: declare
+#pragma acc routine(name) worker
+  // CHECK-NOT: routine
+}
+#pragma acc declare
+  // CHECK-NOT: declare
+#pragma acc routine(name) worker
+  // CHECK-NOT: routine