1	//===--- ParseOpenACC.cpp - OpenACC-specific parsing support --------------===//
2	//
3	// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4	// See https://llvm.org/LICENSE.txt for license information.
5	// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6	//
7	//===----------------------------------------------------------------------===//
8	//
9	// This file implements the parsing logic for OpenACC language features.
10	//
11	//===----------------------------------------------------------------------===//
12
13	#include "clang/AST/OpenACCClause.h"
14	#include "clang/Basic/DiagnosticParse.h"
15	#include "clang/Basic/OpenACCKinds.h"
16	#include "clang/Parse/Parser.h"
17	#include "clang/Parse/RAIIObjectsForParser.h"
18	#include "clang/Sema/ParsedAttr.h"
19	#include "clang/Sema/SemaOpenACC.h"
20	#include "llvm/ADT/StringRef.h"
21	#include "llvm/ADT/StringSwitch.h"
22
23	using namespace clang;
24	using namespace llvm;
25
26	namespace {
27	// An enum that contains the extended 'partial' parsed variants. This type
28	// should never escape the initial parse functionality, but is useful for
29	// simplifying the implementation.
30	enum class OpenACCDirectiveKindEx {
31	Invalid = static_cast<int>(OpenACCDirectiveKind::Invalid),
32	// 'enter data' and 'exit data'
33	Enter,
34	Exit,
35	};
36
37	// Translate single-token string representations to the OpenACC Directive Kind.
38	// This doesn't completely comprehend 'Compound Constructs' (as it just
39	// identifies the first token), and doesn't fully handle 'enter data', 'exit
40	// data', nor any of the 'atomic' variants, just the first token of each. So
41	// this should only be used by `ParseOpenACCDirectiveKind`.
42	OpenACCDirectiveKindEx getOpenACCDirectiveKind(Token Tok) {
43	if (!Tok.is(K: tok::identifier))
44	return OpenACCDirectiveKindEx::Invalid;
45	OpenACCDirectiveKind DirKind =
46	llvm::StringSwitch<OpenACCDirectiveKind>(
47	Tok.getIdentifierInfo()->getName())
48	.Case(S: "parallel", Value: OpenACCDirectiveKind::Parallel)
49	.Case(S: "serial", Value: OpenACCDirectiveKind::Serial)
50	.Case(S: "kernels", Value: OpenACCDirectiveKind::Kernels)
51	.Case(S: "data", Value: OpenACCDirectiveKind::Data)
52	.Case(S: "host_data", Value: OpenACCDirectiveKind::HostData)
53	.Case(S: "loop", Value: OpenACCDirectiveKind::Loop)
54	.Case(S: "cache", Value: OpenACCDirectiveKind::Cache)
55	.Case(S: "atomic", Value: OpenACCDirectiveKind::Atomic)
56	.Case(S: "routine", Value: OpenACCDirectiveKind::Routine)
57	.Case(S: "declare", Value: OpenACCDirectiveKind::Declare)
58	.Case(S: "init", Value: OpenACCDirectiveKind::Init)
59	.Case(S: "shutdown", Value: OpenACCDirectiveKind::Shutdown)
60	.Case(S: "set", Value: OpenACCDirectiveKind::Set)
61	.Case(S: "update", Value: OpenACCDirectiveKind::Update)
62	.Case(S: "wait", Value: OpenACCDirectiveKind::Wait)
63	.Default(Value: OpenACCDirectiveKind::Invalid);
64
65	if (DirKind != OpenACCDirectiveKind::Invalid)
66	return static_cast<OpenACCDirectiveKindEx>(DirKind);
67
68	return llvm::StringSwitch<OpenACCDirectiveKindEx>(
69	Tok.getIdentifierInfo()->getName())
70	.Case(S: "enter", Value: OpenACCDirectiveKindEx::Enter)
71	.Case(S: "exit", Value: OpenACCDirectiveKindEx::Exit)
72	.Default(Value: OpenACCDirectiveKindEx::Invalid);
73	}
74
75	// Translate single-token string representations to the OpenCC Clause Kind.
76	OpenACCClauseKind getOpenACCClauseKind(Token Tok) {
77	// auto is a keyword in some language modes, so make sure we parse it
78	// correctly.
79	if (Tok.is(K: tok::kw_auto))
80	return OpenACCClauseKind::Auto;
81
82	// default is a keyword, so make sure we parse it correctly.
83	if (Tok.is(K: tok::kw_default))
84	return OpenACCClauseKind::Default;
85
86	// if is also a keyword, make sure we parse it correctly.
87	if (Tok.is(K: tok::kw_if))
88	return OpenACCClauseKind::If;
89
90	// 'private' is also a keyword, make sure we parse it correctly.
91	if (Tok.is(K: tok::kw_private))
92	return OpenACCClauseKind::Private;
93
94	// 'delete' is a keyword, make sure we parse it correctly.
95	if (Tok.is(K: tok::kw_delete))
96	return OpenACCClauseKind::Delete;
97
98	if (!Tok.is(K: tok::identifier))
99	return OpenACCClauseKind::Invalid;
100
101	return llvm::StringSwitch<OpenACCClauseKind>(
102	Tok.getIdentifierInfo()->getName())
103	.Case(S: "async", Value: OpenACCClauseKind::Async)
104	.Case(S: "attach", Value: OpenACCClauseKind::Attach)
105	.Case(S: "auto", Value: OpenACCClauseKind::Auto)
106	.Case(S: "bind", Value: OpenACCClauseKind::Bind)
107	.Case(S: "create", Value: OpenACCClauseKind::Create)
108	.Case(S: "pcreate", Value: OpenACCClauseKind::PCreate)
109	.Case(S: "present_or_create", Value: OpenACCClauseKind::PresentOrCreate)
110	.Case(S: "collapse", Value: OpenACCClauseKind::Collapse)
111	.Case(S: "copy", Value: OpenACCClauseKind::Copy)
112	.Case(S: "pcopy", Value: OpenACCClauseKind::PCopy)
113	.Case(S: "present_or_copy", Value: OpenACCClauseKind::PresentOrCopy)
114	.Case(S: "copyin", Value: OpenACCClauseKind::CopyIn)
115	.Case(S: "pcopyin", Value: OpenACCClauseKind::PCopyIn)
116	.Case(S: "present_or_copyin", Value: OpenACCClauseKind::PresentOrCopyIn)
117	.Case(S: "copyout", Value: OpenACCClauseKind::CopyOut)
118	.Case(S: "pcopyout", Value: OpenACCClauseKind::PCopyOut)
119	.Case(S: "present_or_copyout", Value: OpenACCClauseKind::PresentOrCopyOut)
120	.Case(S: "default", Value: OpenACCClauseKind::Default)
121	.Case(S: "default_async", Value: OpenACCClauseKind::DefaultAsync)
122	.Case(S: "delete", Value: OpenACCClauseKind::Delete)
123	.Case(S: "detach", Value: OpenACCClauseKind::Detach)
124	.Case(S: "device", Value: OpenACCClauseKind::Device)
125	.Case(S: "device_num", Value: OpenACCClauseKind::DeviceNum)
126	.Case(S: "device_resident", Value: OpenACCClauseKind::DeviceResident)
127	.Case(S: "device_type", Value: OpenACCClauseKind::DeviceType)
128	.Case(S: "deviceptr", Value: OpenACCClauseKind::DevicePtr)
129	.Case(S: "dtype", Value: OpenACCClauseKind::DType)
130	.Case(S: "finalize", Value: OpenACCClauseKind::Finalize)
131	.Case(S: "firstprivate", Value: OpenACCClauseKind::FirstPrivate)
132	.Case(S: "gang", Value: OpenACCClauseKind::Gang)
133	.Case(S: "host", Value: OpenACCClauseKind::Host)
134	.Case(S: "if", Value: OpenACCClauseKind::If)
135	.Case(S: "if_present", Value: OpenACCClauseKind::IfPresent)
136	.Case(S: "independent", Value: OpenACCClauseKind::Independent)
137	.Case(S: "link", Value: OpenACCClauseKind::Link)
138	.Case(S: "no_create", Value: OpenACCClauseKind::NoCreate)
139	.Case(S: "num_gangs", Value: OpenACCClauseKind::NumGangs)
140	.Case(S: "num_workers", Value: OpenACCClauseKind::NumWorkers)
141	.Case(S: "nohost", Value: OpenACCClauseKind::NoHost)
142	.Case(S: "present", Value: OpenACCClauseKind::Present)
143	.Case(S: "private", Value: OpenACCClauseKind::Private)
144	.Case(S: "reduction", Value: OpenACCClauseKind::Reduction)
145	.Case(S: "self", Value: OpenACCClauseKind::Self)
146	.Case(S: "seq", Value: OpenACCClauseKind::Seq)
147	.Case(S: "tile", Value: OpenACCClauseKind::Tile)
148	.Case(S: "use_device", Value: OpenACCClauseKind::UseDevice)
149	.Case(S: "vector", Value: OpenACCClauseKind::Vector)
150	.Case(S: "vector_length", Value: OpenACCClauseKind::VectorLength)
151	.Case(S: "wait", Value: OpenACCClauseKind::Wait)
152	.Case(S: "worker", Value: OpenACCClauseKind::Worker)
153	.Default(Value: OpenACCClauseKind::Invalid);
154	}
155
156	// Since 'atomic' is effectively a compound directive, this will decode the
157	// second part of the directive.
158	OpenACCAtomicKind getOpenACCAtomicKind(Token Tok) {
159	if (!Tok.is(K: tok::identifier))
160	return OpenACCAtomicKind::None;
161	return llvm::StringSwitch<OpenACCAtomicKind>(
162	Tok.getIdentifierInfo()->getName())
163	.Case(S: "read", Value: OpenACCAtomicKind::Read)
164	.Case(S: "write", Value: OpenACCAtomicKind::Write)
165	.Case(S: "update", Value: OpenACCAtomicKind::Update)
166	.Case(S: "capture", Value: OpenACCAtomicKind::Capture)
167	.Default(Value: OpenACCAtomicKind::None);
168	}
169
170	OpenACCDefaultClauseKind getOpenACCDefaultClauseKind(Token Tok) {
171	if (!Tok.is(K: tok::identifier))
172	return OpenACCDefaultClauseKind::Invalid;
173
174	return llvm::StringSwitch<OpenACCDefaultClauseKind>(
175	Tok.getIdentifierInfo()->getName())
176	.Case(S: "none", Value: OpenACCDefaultClauseKind::None)
177	.Case(S: "present", Value: OpenACCDefaultClauseKind::Present)
178	.Default(Value: OpenACCDefaultClauseKind::Invalid);
179	}
180
181	enum class OpenACCSpecialTokenKind {
182	ReadOnly,
183	DevNum,
184	Queues,
185	Zero,
186	Force,
187	Num,
188	Length,
189	Dim,
190	Static,
191	};
192
193	bool isOpenACCSpecialToken(OpenACCSpecialTokenKind Kind, Token Tok) {
194	if (Tok.is(K: tok::kw_static) && Kind == OpenACCSpecialTokenKind::Static)
195	return true;
196
197	if (!Tok.is(K: tok::identifier))
198	return false;
199
200	switch (Kind) {
201	case OpenACCSpecialTokenKind::ReadOnly:
202	return Tok.getIdentifierInfo()->isStr(Str: "readonly");
203	case OpenACCSpecialTokenKind::DevNum:
204	return Tok.getIdentifierInfo()->isStr(Str: "devnum");
205	case OpenACCSpecialTokenKind::Queues:
206	return Tok.getIdentifierInfo()->isStr(Str: "queues");
207	case OpenACCSpecialTokenKind::Zero:
208	return Tok.getIdentifierInfo()->isStr(Str: "zero");
209	case OpenACCSpecialTokenKind::Force:
210	return Tok.getIdentifierInfo()->isStr(Str: "force");
211	case OpenACCSpecialTokenKind::Num:
212	return Tok.getIdentifierInfo()->isStr(Str: "num");
213	case OpenACCSpecialTokenKind::Length:
214	return Tok.getIdentifierInfo()->isStr(Str: "length");
215	case OpenACCSpecialTokenKind::Dim:
216	return Tok.getIdentifierInfo()->isStr(Str: "dim");
217	case OpenACCSpecialTokenKind::Static:
218	return Tok.getIdentifierInfo()->isStr(Str: "static");
219	}
220	llvm_unreachable("Unknown 'Kind' Passed");
221	}
222
223	/// Used for cases where we have a token we want to check against an
224	/// 'identifier-like' token, but don't want to give awkward error messages in
225	/// cases where it is accidentially a keyword.
226	bool isTokenIdentifierOrKeyword(Parser &P, Token Tok) {
227	if (Tok.is(K: tok::identifier))
228	return true;
229
230	if (!Tok.isAnnotation() && Tok.getIdentifierInfo() &&
231	Tok.getIdentifierInfo()->isKeyword(LangOpts: P.getLangOpts()))
232	return true;
233
234	return false;
235	}
236
237	/// Parses and consumes an identifer followed immediately by a single colon, and
238	/// diagnoses if it is not the 'special token' kind that we require. Used when
239	/// the tag is the only valid value.
240	/// Return 'true' if the special token was matched, false if no special token,
241	/// or an invalid special token was found.
242	template <typename DirOrClauseTy>
243	bool tryParseAndConsumeSpecialTokenKind(Parser &P, OpenACCSpecialTokenKind Kind,
244	DirOrClauseTy DirOrClause) {
245	Token IdentTok = P.getCurToken();
246	// If this is an identifier-like thing followed by ':', it is one of the
247	// OpenACC 'special' name tags, so consume it.
248	if (isTokenIdentifierOrKeyword(P, Tok: IdentTok) && P.NextToken().is(K: tok::colon)) {
249	P.ConsumeToken();
250	P.ConsumeToken();
251
252	if (!isOpenACCSpecialToken(Kind, Tok: IdentTok)) {
253	P.Diag(IdentTok, diag::err_acc_invalid_tag_kind)
254	<< IdentTok.getIdentifierInfo() << DirOrClause
255	<< std::is_same_v<DirOrClauseTy, OpenACCClauseKind>;
256	return false;
257	}
258
259	return true;
260	}
261
262	return false;
263	}
264
265	bool isOpenACCDirectiveKind(OpenACCDirectiveKind Kind, Token Tok) {
266	if (!Tok.is(K: tok::identifier))
267	return false;
268
269	switch (Kind) {
270	case OpenACCDirectiveKind::Parallel:
271	return Tok.getIdentifierInfo()->isStr(Str: "parallel");
272	case OpenACCDirectiveKind::Serial:
273	return Tok.getIdentifierInfo()->isStr(Str: "serial");
274	case OpenACCDirectiveKind::Kernels:
275	return Tok.getIdentifierInfo()->isStr(Str: "kernels");
276	case OpenACCDirectiveKind::Data:
277	return Tok.getIdentifierInfo()->isStr(Str: "data");
278	case OpenACCDirectiveKind::HostData:
279	return Tok.getIdentifierInfo()->isStr(Str: "host_data");
280	case OpenACCDirectiveKind::Loop:
281	return Tok.getIdentifierInfo()->isStr(Str: "loop");
282	case OpenACCDirectiveKind::Cache:
283	return Tok.getIdentifierInfo()->isStr(Str: "cache");
284
285	case OpenACCDirectiveKind::ParallelLoop:
286	case OpenACCDirectiveKind::SerialLoop:
287	case OpenACCDirectiveKind::KernelsLoop:
288	case OpenACCDirectiveKind::EnterData:
289	case OpenACCDirectiveKind::ExitData:
290	return false;
291
292	case OpenACCDirectiveKind::Atomic:
293	return Tok.getIdentifierInfo()->isStr(Str: "atomic");
294	case OpenACCDirectiveKind::Routine:
295	return Tok.getIdentifierInfo()->isStr(Str: "routine");
296	case OpenACCDirectiveKind::Declare:
297	return Tok.getIdentifierInfo()->isStr(Str: "declare");
298	case OpenACCDirectiveKind::Init:
299	return Tok.getIdentifierInfo()->isStr(Str: "init");
300	case OpenACCDirectiveKind::Shutdown:
301	return Tok.getIdentifierInfo()->isStr(Str: "shutdown");
302	case OpenACCDirectiveKind::Set:
303	return Tok.getIdentifierInfo()->isStr(Str: "set");
304	case OpenACCDirectiveKind::Update:
305	return Tok.getIdentifierInfo()->isStr(Str: "update");
306	case OpenACCDirectiveKind::Wait:
307	return Tok.getIdentifierInfo()->isStr(Str: "wait");
308	case OpenACCDirectiveKind::Invalid:
309	return false;
310	}
311	llvm_unreachable("Unknown 'Kind' Passed");
312	}
313
314	OpenACCReductionOperator ParseReductionOperator(Parser &P) {
315	// If there is no colon, treat as if the reduction operator was missing, else
316	// we probably will not recover from it in the case where an expression starts
317	// with one of the operator tokens.
318	if (P.NextToken().isNot(K: tok::colon)) {
319	P.Diag(P.getCurToken(), diag::err_acc_expected_reduction_operator);
320	return OpenACCReductionOperator::Invalid;
321	}
322	Token ReductionKindTok = P.getCurToken();
323	// Consume both the kind and the colon.
324	P.ConsumeToken();
325	P.ConsumeToken();
326
327	switch (ReductionKindTok.getKind()) {
328	case tok::plus:
329	return OpenACCReductionOperator::Addition;
330	case tok::star:
331	return OpenACCReductionOperator::Multiplication;
332	case tok::amp:
333	return OpenACCReductionOperator::BitwiseAnd;
334	case tok::pipe:
335	return OpenACCReductionOperator::BitwiseOr;
336	case tok::caret:
337	return OpenACCReductionOperator::BitwiseXOr;
338	case tok::ampamp:
339	return OpenACCReductionOperator::And;
340	case tok::pipepipe:
341	return OpenACCReductionOperator::Or;
342	case tok::identifier:
343	if (ReductionKindTok.getIdentifierInfo()->isStr(Str: "max"))
344	return OpenACCReductionOperator::Max;
345	if (ReductionKindTok.getIdentifierInfo()->isStr(Str: "min"))
346	return OpenACCReductionOperator::Min;
347	[[fallthrough]];
348	default:
349	P.Diag(ReductionKindTok, diag::err_acc_invalid_reduction_operator);
350	return OpenACCReductionOperator::Invalid;
351	}
352	llvm_unreachable("Reduction op token kind not caught by 'default'?");
353	}
354
355	/// Used for cases where we expect an identifier-like token, but don't want to
356	/// give awkward error messages in cases where it is accidentially a keyword.
357	bool expectIdentifierOrKeyword(Parser &P) {
358	Token Tok = P.getCurToken();
359
360	if (isTokenIdentifierOrKeyword(P, Tok))
361	return false;
362
363	P.Diag(P.getCurToken(), diag::err_expected) << tok::identifier;
364	return true;
365	}
366
367	OpenACCDirectiveKind
368	ParseOpenACCEnterExitDataDirective(Parser &P, Token FirstTok,
369	OpenACCDirectiveKindEx ExtDirKind) {
370	Token SecondTok = P.getCurToken();
371
372	if (SecondTok.isAnnotation()) {
373	P.Diag(FirstTok, diag::err_acc_invalid_directive)
374	<< 0 << FirstTok.getIdentifierInfo();
375	return OpenACCDirectiveKind::Invalid;
376	}
377
378	// Consume the second name anyway, this way we can continue on without making
379	// this oddly look like a clause.
380	P.ConsumeAnyToken();
381
382	if (!isOpenACCDirectiveKind(Kind: OpenACCDirectiveKind::Data, Tok: SecondTok)) {
383	if (!SecondTok.is(tok::identifier))
384	P.Diag(SecondTok, diag::err_expected) << tok::identifier;
385	else
386	P.Diag(FirstTok, diag::err_acc_invalid_directive)
387	<< 1 << FirstTok.getIdentifierInfo()->getName()
388	<< SecondTok.getIdentifierInfo()->getName();
389	return OpenACCDirectiveKind::Invalid;
390	}
391
392	return ExtDirKind == OpenACCDirectiveKindEx::Enter
393	? OpenACCDirectiveKind::EnterData
394	: OpenACCDirectiveKind::ExitData;
395	}
396
397	OpenACCAtomicKind ParseOpenACCAtomicKind(Parser &P) {
398	Token AtomicClauseToken = P.getCurToken();
399
400	// #pragma acc atomic is equivilent to update:
401	if (AtomicClauseToken.isAnnotation())
402	return OpenACCAtomicKind::None;
403
404	OpenACCAtomicKind AtomicKind = getOpenACCAtomicKind(Tok: AtomicClauseToken);
405
406	// If this isn't a valid atomic-kind, don't consume the token, and treat the
407	// rest as a clause list, which despite there being no permissible clauses,
408	// will diagnose as a clause.
409	if (AtomicKind != OpenACCAtomicKind::None)
410	P.ConsumeToken();
411
412	return AtomicKind;
413	}
414
415	// Parse and consume the tokens for OpenACC Directive/Construct kinds.
416	OpenACCDirectiveKind ParseOpenACCDirectiveKind(Parser &P) {
417	Token FirstTok = P.getCurToken();
418
419	// Just #pragma acc can get us immediately to the end, make sure we don't
420	// introspect on the spelling before then.
421	if (FirstTok.isNot(K: tok::identifier)) {
422	P.Diag(FirstTok, diag::err_acc_missing_directive);
423
424	if (P.getCurToken().isNot(K: tok::annot_pragma_openacc_end))
425	P.ConsumeAnyToken();
426
427	return OpenACCDirectiveKind::Invalid;
428	}
429
430	P.ConsumeToken();
431
432	OpenACCDirectiveKindEx ExDirKind = getOpenACCDirectiveKind(Tok: FirstTok);
433
434	// OpenACCDirectiveKindEx is meant to be an extended list
435	// over OpenACCDirectiveKind, so any value below Invalid is one of the
436	// OpenACCDirectiveKind values. This switch takes care of all of the extra
437	// parsing required for the Extended values. At the end of this block,
438	// ExDirKind can be assumed to be a valid OpenACCDirectiveKind, so we can
439	// immediately cast it and use it as that.
440	if (ExDirKind >= OpenACCDirectiveKindEx::Invalid) {
441	switch (ExDirKind) {
442	case OpenACCDirectiveKindEx::Invalid: {
443	P.Diag(FirstTok, diag::err_acc_invalid_directive)
444	<< 0 << FirstTok.getIdentifierInfo();
445	return OpenACCDirectiveKind::Invalid;
446	}
447	case OpenACCDirectiveKindEx::Enter:
448	case OpenACCDirectiveKindEx::Exit:
449	return ParseOpenACCEnterExitDataDirective(P, FirstTok, ExtDirKind: ExDirKind);
450	}
451	}
452
453	OpenACCDirectiveKind DirKind = static_cast<OpenACCDirectiveKind>(ExDirKind);
454
455	// Combined Constructs allows parallel loop, serial loop, or kernels loop. Any
456	// other attempt at a combined construct will be diagnosed as an invalid
457	// clause.
458	Token SecondTok = P.getCurToken();
459	if (!SecondTok.isAnnotation() &&
460	isOpenACCDirectiveKind(Kind: OpenACCDirectiveKind::Loop, Tok: SecondTok)) {
461	switch (DirKind) {
462	default:
463	// Nothing to do except in the below cases, as they should be diagnosed as
464	// a clause.
465	break;
466	case OpenACCDirectiveKind::Parallel:
467	P.ConsumeToken();
468	return OpenACCDirectiveKind::ParallelLoop;
469	case OpenACCDirectiveKind::Serial:
470	P.ConsumeToken();
471	return OpenACCDirectiveKind::SerialLoop;
472	case OpenACCDirectiveKind::Kernels:
473	P.ConsumeToken();
474	return OpenACCDirectiveKind::KernelsLoop;
475	}
476	}
477
478	return DirKind;
479	}
480
481	enum ClauseParensKind {
482	None,
483	Optional,
484	Required
485	};
486
487	ClauseParensKind getClauseParensKind(OpenACCDirectiveKind DirKind,
488	OpenACCClauseKind Kind) {
489	switch (Kind) {
490	case OpenACCClauseKind::Self:
491	return DirKind == OpenACCDirectiveKind::Update ? ClauseParensKind::Required
492	: ClauseParensKind::Optional;
493	case OpenACCClauseKind::Async:
494	case OpenACCClauseKind::Worker:
495	case OpenACCClauseKind::Vector:
496	case OpenACCClauseKind::Gang:
497	case OpenACCClauseKind::Wait:
498	return ClauseParensKind::Optional;
499
500	case OpenACCClauseKind::Default:
501	case OpenACCClauseKind::If:
502	case OpenACCClauseKind::Create:
503	case OpenACCClauseKind::PCreate:
504	case OpenACCClauseKind::PresentOrCreate:
505	case OpenACCClauseKind::Copy:
506	case OpenACCClauseKind::PCopy:
507	case OpenACCClauseKind::PresentOrCopy:
508	case OpenACCClauseKind::CopyIn:
509	case OpenACCClauseKind::PCopyIn:
510	case OpenACCClauseKind::PresentOrCopyIn:
511	case OpenACCClauseKind::CopyOut:
512	case OpenACCClauseKind::PCopyOut:
513	case OpenACCClauseKind::PresentOrCopyOut:
514	case OpenACCClauseKind::UseDevice:
515	case OpenACCClauseKind::NoCreate:
516	case OpenACCClauseKind::Present:
517	case OpenACCClauseKind::DevicePtr:
518	case OpenACCClauseKind::Attach:
519	case OpenACCClauseKind::Detach:
520	case OpenACCClauseKind::Private:
521	case OpenACCClauseKind::FirstPrivate:
522	case OpenACCClauseKind::Delete:
523	case OpenACCClauseKind::DeviceResident:
524	case OpenACCClauseKind::Device:
525	case OpenACCClauseKind::Link:
526	case OpenACCClauseKind::Host:
527	case OpenACCClauseKind::Reduction:
528	case OpenACCClauseKind::Collapse:
529	case OpenACCClauseKind::Bind:
530	case OpenACCClauseKind::VectorLength:
531	case OpenACCClauseKind::NumGangs:
532	case OpenACCClauseKind::NumWorkers:
533	case OpenACCClauseKind::DeviceNum:
534	case OpenACCClauseKind::DefaultAsync:
535	case OpenACCClauseKind::DeviceType:
536	case OpenACCClauseKind::DType:
537	case OpenACCClauseKind::Tile:
538	return ClauseParensKind::Required;
539
540	case OpenACCClauseKind::Shortloop:
541	llvm_unreachable("Shortloop shouldn't be generated in clang");
542	case OpenACCClauseKind::Auto:
543	case OpenACCClauseKind::Finalize:
544	case OpenACCClauseKind::IfPresent:
545	case OpenACCClauseKind::Independent:
546	case OpenACCClauseKind::Invalid:
547	case OpenACCClauseKind::NoHost:
548	case OpenACCClauseKind::Seq:
549	return ClauseParensKind::None;
550	}
551	llvm_unreachable("Unhandled clause kind");
552	}
553
554	bool ClauseHasOptionalParens(OpenACCDirectiveKind DirKind,
555	OpenACCClauseKind Kind) {
556	return getClauseParensKind(DirKind, Kind) == ClauseParensKind::Optional;
557	}
558
559	bool ClauseHasRequiredParens(OpenACCDirectiveKind DirKind,
560	OpenACCClauseKind Kind) {
561	return getClauseParensKind(DirKind, Kind) == ClauseParensKind::Required;
562	}
563
564	// Skip until we see the end of pragma token, but don't consume it. This is us
565	// just giving up on the rest of the pragma so we can continue executing. We
566	// have to do this because 'SkipUntil' considers paren balancing, which isn't
567	// what we want.
568	void SkipUntilEndOfDirective(Parser &P) {
569	while (P.getCurToken().isNot(K: tok::annot_pragma_openacc_end))
570	P.ConsumeAnyToken();
571	}
572
573	bool doesDirectiveHaveAssociatedStmt(OpenACCDirectiveKind DirKind) {
574	switch (DirKind) {
575	case OpenACCDirectiveKind::Routine:
576	// FIXME: Routine MIGHT end up needing to be 'true' here, as it needs a way
577	// to capture a lambda-expression on the next line.
578	case OpenACCDirectiveKind::Cache:
579	case OpenACCDirectiveKind::Declare:
580	case OpenACCDirectiveKind::Set:
581	case OpenACCDirectiveKind::EnterData:
582	case OpenACCDirectiveKind::ExitData:
583	case OpenACCDirectiveKind::Wait:
584	case OpenACCDirectiveKind::Init:
585	case OpenACCDirectiveKind::Shutdown:
586	case OpenACCDirectiveKind::Update:
587	case OpenACCDirectiveKind::Invalid:
588	return false;
589	case OpenACCDirectiveKind::Parallel:
590	case OpenACCDirectiveKind::Serial:
591	case OpenACCDirectiveKind::Kernels:
592	case OpenACCDirectiveKind::ParallelLoop:
593	case OpenACCDirectiveKind::SerialLoop:
594	case OpenACCDirectiveKind::KernelsLoop:
595	case OpenACCDirectiveKind::Loop:
596	case OpenACCDirectiveKind::Data:
597	case OpenACCDirectiveKind::HostData:
598	case OpenACCDirectiveKind::Atomic:
599	return true;
600	}
601	llvm_unreachable("Unhandled directive->assoc stmt");
602	}
603
604	unsigned getOpenACCScopeFlags(OpenACCDirectiveKind DirKind) {
605	switch (DirKind) {
606	case OpenACCDirectiveKind::Parallel:
607	case OpenACCDirectiveKind::Serial:
608	case OpenACCDirectiveKind::Kernels:
609	case OpenACCDirectiveKind::ParallelLoop:
610	case OpenACCDirectiveKind::SerialLoop:
611	case OpenACCDirectiveKind::KernelsLoop:
612	// Mark this as a BreakScope/ContinueScope as well as a compute construct
613	// so that we can diagnose trying to 'break'/'continue' inside of one.
614	return Scope::BreakScope | Scope::ContinueScope |
615	Scope::OpenACCComputeConstructScope;
616	case OpenACCDirectiveKind::Data:
617	case OpenACCDirectiveKind::EnterData:
618	case OpenACCDirectiveKind::ExitData:
619	case OpenACCDirectiveKind::HostData:
620	case OpenACCDirectiveKind::Wait:
621	case OpenACCDirectiveKind::Init:
622	case OpenACCDirectiveKind::Shutdown:
623	case OpenACCDirectiveKind::Cache:
624	case OpenACCDirectiveKind::Loop:
625	case OpenACCDirectiveKind::Atomic:
626	case OpenACCDirectiveKind::Declare:
627	case OpenACCDirectiveKind::Routine:
628	case OpenACCDirectiveKind::Set:
629	case OpenACCDirectiveKind::Update:
630	return 0;
631	case OpenACCDirectiveKind::Invalid:
632	llvm_unreachable("Shouldn't be creating a scope for an invalid construct");
633	}
634	llvm_unreachable("Shouldn't be creating a scope for an invalid construct");
635	}
636
637	} // namespace
638
639	Parser::OpenACCClauseParseResult Parser::OpenACCCanContinue() {
640	return {nullptr, OpenACCParseCanContinue::Can};
641	}
642
643	Parser::OpenACCClauseParseResult Parser::OpenACCCannotContinue() {
644	return {nullptr, OpenACCParseCanContinue::Cannot};
645	}
646
647	Parser::OpenACCClauseParseResult Parser::OpenACCSuccess(OpenACCClause *Clause) {
648	return {Clause, OpenACCParseCanContinue::Can};
649	}
650
651	ExprResult Parser::ParseOpenACCConditionExpr() {
652	// FIXME: It isn't clear if the spec saying 'condition' means the same as
653	// it does in an if/while/etc (See ParseCXXCondition), however as it was
654	// written with Fortran/C in mind, we're going to assume it just means an
655	// 'expression evaluating to boolean'.
656	ExprResult ER = getActions().CorrectDelayedTyposInExpr(ER: ParseExpression());
657
658	if (!ER.isUsable())
659	return ER;
660
661	Sema::ConditionResult R =
662	getActions().ActOnCondition(S: getCurScope(), Loc: ER.get()->getExprLoc(),
663	SubExpr: ER.get(), CK: Sema::ConditionKind::Boolean);
664
665	return R.isInvalid() ? ExprError() : R.get().second;
666	}
667
668	OpenACCModifierKind Parser::tryParseModifierList(OpenACCClauseKind CK) {
669	// Use the tentative parsing to decide whether we are a comma-delmited list of
670	// identifers ending in a colon so we can do an actual parse with diagnostics.
671	{
672	RevertingTentativeParsingAction TPA{*this};
673	// capture any <ident><comma> pairs.
674	while (isTokenIdentifierOrKeyword(P&: *this, Tok: getCurToken()) &&
675	NextToken().is(K: tok::comma)) {
676	ConsumeToken();
677	ConsumeToken();
678	}
679
680	if (!isTokenIdentifierOrKeyword(P&: *this, Tok: getCurToken()) ||
681	!NextToken().is(K: tok::colon)) {
682	// No modifiers as this isn't a valid modifier-list.
683	return OpenACCModifierKind::Invalid;
684	}
685	}
686
687	auto GetModKind = [](Token T) {
688	return StringSwitch<OpenACCModifierKind>(T.getIdentifierInfo()->getName())
689	.Case(S: "always", Value: OpenACCModifierKind::Always)
690	.Case(S: "alwaysin", Value: OpenACCModifierKind::AlwaysIn)
691	.Case(S: "alwaysout", Value: OpenACCModifierKind::AlwaysOut)
692	.Case(S: "readonly", Value: OpenACCModifierKind::Readonly)
693	.Case(S: "zero", Value: OpenACCModifierKind::Zero)
694	.Case(S: "capture", Value: OpenACCModifierKind::Capture)
695	.Default(Value: OpenACCModifierKind::Invalid);
696	};
697
698	OpenACCModifierKind CurModList = OpenACCModifierKind::Invalid;
699	auto ConsumeModKind = [&]() {
700	Token IdentToken = getCurToken();
701	OpenACCModifierKind NewKind = GetModKind(IdentToken);
702
703	if (NewKind == OpenACCModifierKind::Invalid)
704	Diag(IdentToken.getLocation(), diag::err_acc_modifier)
705	<< diag::ACCModifier::Unknown << IdentToken.getIdentifierInfo() << CK;
706	else if ((NewKind & CurModList) != OpenACCModifierKind::Invalid)
707	Diag(IdentToken.getLocation(), diag::err_acc_modifier)
708	<< diag::ACCModifier::Duplicate << IdentToken.getIdentifierInfo()
709	<< CK;
710	else
711	CurModList |= NewKind;
712
713	// Consumes the identifier.
714	ConsumeToken();
715	// Consumes the comma or colon.
716	ConsumeToken();
717	};
718
719	// Inspect all but the last item. We inspected enough to know that our current
720	// token is the identifier-like thing, so just check for the comma.
721	while (NextToken().is(K: tok::comma))
722	ConsumeModKind();
723
724	// Above we confirmed that this should be correct/we should be on the last
725	// item.
726	ConsumeModKind();
727
728	return CurModList;
729	}
730
731	SmallVector<OpenACCClause *>
732	Parser::ParseOpenACCClauseList(OpenACCDirectiveKind DirKind) {
733	SmallVector<OpenACCClause *> Clauses;
734	bool FirstClause = true;
735	while (getCurToken().isNot(K: tok::annot_pragma_openacc_end)) {
736	// Comma is optional in a clause-list.
737	if (!FirstClause && getCurToken().is(K: tok::comma))
738	ConsumeToken();
739	FirstClause = false;
740
741	OpenACCClauseParseResult Result = ParseOpenACCClause(ExistingClauses: Clauses, DirKind);
742	if (OpenACCClause *Clause = Result.getPointer()) {
743	Clauses.push_back(Elt: Clause);
744	} else if (Result.getInt() == OpenACCParseCanContinue::Cannot) {
745	// Recovering from a bad clause is really difficult, so we just give up on
746	// error.
747	SkipUntilEndOfDirective(P&: *this);
748	return Clauses;
749	}
750	}
751	return Clauses;
752	}
753
754	Parser::OpenACCIntExprParseResult
755	Parser::ParseOpenACCIntExpr(OpenACCDirectiveKind DK, OpenACCClauseKind CK,
756	SourceLocation Loc) {
757	ExprResult ER = ParseAssignmentExpression();
758
759	// If the actual parsing failed, we don't know the state of the parse, so
760	// don't try to continue.
761	if (!ER.isUsable())
762	return {ER, OpenACCParseCanContinue::Cannot};
763
764	// Parsing can continue after the initial assignment expression parsing, so
765	// even if there was a typo, we can continue.
766	ER = getActions().CorrectDelayedTyposInExpr(ER);
767	if (!ER.isUsable())
768	return {ER, OpenACCParseCanContinue::Can};
769
770	return {getActions().OpenACC().ActOnIntExpr(DK, CK, Loc, IntExpr: ER.get()),
771	OpenACCParseCanContinue::Can};
772	}
773
774	bool Parser::ParseOpenACCIntExprList(OpenACCDirectiveKind DK,
775	OpenACCClauseKind CK, SourceLocation Loc,
776	llvm::SmallVectorImpl<Expr *> &IntExprs) {
777	OpenACCIntExprParseResult CurResult = ParseOpenACCIntExpr(DK, CK, Loc);
778
779	if (!CurResult.first.isUsable() &&
780	CurResult.second == OpenACCParseCanContinue::Cannot) {
781	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end,
782	Flags: Parser::StopBeforeMatch);
783	return true;
784	}
785
786	IntExprs.push_back(Elt: CurResult.first.get());
787
788	while (!getCurToken().isOneOf(K1: tok::r_paren, K2: tok::annot_pragma_openacc_end)) {
789	ExpectAndConsume(ExpectedTok: tok::comma);
790
791	CurResult = ParseOpenACCIntExpr(DK, CK, Loc);
792
793	if (!CurResult.first.isUsable() &&
794	CurResult.second == OpenACCParseCanContinue::Cannot) {
795	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end,
796	Flags: Parser::StopBeforeMatch);
797	return true;
798	}
799	IntExprs.push_back(Elt: CurResult.first.get());
800	}
801	return false;
802	}
803
804	bool Parser::ParseOpenACCDeviceTypeList(
805	llvm::SmallVector<IdentifierLoc> &Archs) {
806
807	if (expectIdentifierOrKeyword(P&: *this)) {
808	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end,
809	Flags: Parser::StopBeforeMatch);
810	return true;
811	}
812	IdentifierInfo *Ident = getCurToken().getIdentifierInfo();
813	Archs.emplace_back(Args: ConsumeToken(), Args&: Ident);
814
815	while (!getCurToken().isOneOf(K1: tok::r_paren, K2: tok::annot_pragma_openacc_end)) {
816	ExpectAndConsume(ExpectedTok: tok::comma);
817
818	if (expectIdentifierOrKeyword(P&: *this)) {
819	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end,
820	Flags: Parser::StopBeforeMatch);
821	return true;
822	}
823	Ident = getCurToken().getIdentifierInfo();
824	Archs.emplace_back(Args: ConsumeToken(), Args&: Ident);
825	}
826	return false;
827	}
828
829	ExprResult Parser::ParseOpenACCSizeExpr(OpenACCClauseKind CK) {
830	// The size-expr ends up being ambiguous when only looking at the current
831	// token, as it could be a deref of a variable/expression.
832	if (getCurToken().is(K: tok::star) &&
833	NextToken().isOneOf(K1: tok::comma, Ks: tok::r_paren,
834	Ks: tok::annot_pragma_openacc_end)) {
835	SourceLocation AsteriskLoc = ConsumeToken();
836	return getActions().OpenACC().ActOnOpenACCAsteriskSizeExpr(AsteriskLoc);
837	}
838
839	ExprResult SizeExpr =
840	getActions().CorrectDelayedTyposInExpr(ER: ParseConstantExpression());
841
842	if (!SizeExpr.isUsable())
843	return SizeExpr;
844
845	SizeExpr = getActions().OpenACC().ActOnIntExpr(
846	DK: OpenACCDirectiveKind::Invalid, CK, Loc: SizeExpr.get()->getBeginLoc(),
847	IntExpr: SizeExpr.get());
848
849	return SizeExpr;
850	}
851
852	bool Parser::ParseOpenACCSizeExprList(
853	OpenACCClauseKind CK, llvm::SmallVectorImpl<Expr *> &SizeExprs) {
854	ExprResult SizeExpr = ParseOpenACCSizeExpr(CK);
855	if (!SizeExpr.isUsable()) {
856	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end,
857	Flags: Parser::StopBeforeMatch);
858	return true;
859	}
860
861	SizeExprs.push_back(Elt: SizeExpr.get());
862
863	while (!getCurToken().isOneOf(K1: tok::r_paren, K2: tok::annot_pragma_openacc_end)) {
864	ExpectAndConsume(ExpectedTok: tok::comma);
865
866	SizeExpr = ParseOpenACCSizeExpr(CK);
867	if (!SizeExpr.isUsable()) {
868	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end,
869	Flags: Parser::StopBeforeMatch);
870	return true;
871	}
872	SizeExprs.push_back(Elt: SizeExpr.get());
873	}
874	return false;
875	}
876
877	Parser::OpenACCGangArgRes Parser::ParseOpenACCGangArg(SourceLocation GangLoc) {
878
879	if (isOpenACCSpecialToken(Kind: OpenACCSpecialTokenKind::Static, Tok: getCurToken()) &&
880	NextToken().is(K: tok::colon)) {
881	// 'static' just takes a size-expr, which is an int-expr or an asterisk.
882	ConsumeToken();
883	ConsumeToken();
884	ExprResult Res = ParseOpenACCSizeExpr(CK: OpenACCClauseKind::Gang);
885	return {OpenACCGangKind::Static, Res};
886	}
887
888	if (isOpenACCSpecialToken(Kind: OpenACCSpecialTokenKind::Dim, Tok: getCurToken()) &&
889	NextToken().is(K: tok::colon)) {
890	ConsumeToken();
891	ConsumeToken();
892	// Parse this as a const-expression, and we'll check its integer-ness/value
893	// in CheckGangExpr.
894	ExprResult Res =
895	getActions().CorrectDelayedTyposInExpr(ER: ParseConstantExpression());
896	return {OpenACCGangKind::Dim, Res};
897	}
898
899	if (isOpenACCSpecialToken(Kind: OpenACCSpecialTokenKind::Num, Tok: getCurToken()) &&
900	NextToken().is(K: tok::colon)) {
901	ConsumeToken();
902	ConsumeToken();
903	// Fallthrough to the 'int-expr' handling for when 'num' is omitted.
904	}
905
906	// This is just the 'num' case where 'num' is optional.
907	ExprResult Res = ParseOpenACCIntExpr(DK: OpenACCDirectiveKind::Invalid,
908	CK: OpenACCClauseKind::Gang, Loc: GangLoc)
909	.first;
910	return {OpenACCGangKind::Num, Res};
911	}
912
913	bool Parser::ParseOpenACCGangArgList(
914	SourceLocation GangLoc, llvm::SmallVectorImpl<OpenACCGangKind> &GKs,
915	llvm::SmallVectorImpl<Expr *> &IntExprs) {
916
917	Parser::OpenACCGangArgRes Res = ParseOpenACCGangArg(GangLoc);
918	if (!Res.second.isUsable()) {
919	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end,
920	Flags: Parser::StopBeforeMatch);
921	return true;
922	}
923
924	GKs.push_back(Elt: Res.first);
925	IntExprs.push_back(Elt: Res.second.get());
926
927	while (!getCurToken().isOneOf(K1: tok::r_paren, K2: tok::annot_pragma_openacc_end)) {
928	ExpectAndConsume(ExpectedTok: tok::comma);
929
930	Res = ParseOpenACCGangArg(GangLoc);
931	if (!Res.second.isUsable()) {
932	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end,
933	Flags: Parser::StopBeforeMatch);
934	return true;
935	}
936
937	GKs.push_back(Elt: Res.first);
938	IntExprs.push_back(Elt: Res.second.get());
939	}
940	return false;
941	}
942
943	namespace {
944	bool isUnsupportedExtensionClause(Token Tok) {
945	if (!Tok.is(K: tok::identifier))
946	return false;
947
948	return Tok.getIdentifierInfo()->getName().starts_with(Prefix: "__");
949	}
950	} // namespace
951
952	Parser::OpenACCClauseParseResult
953	Parser::ParseOpenACCClause(ArrayRef<const OpenACCClause *> ExistingClauses,
954	OpenACCDirectiveKind DirKind) {
955	// A number of clause names are actually keywords, so accept a keyword that
956	// can be converted to a name.
957	if (expectIdentifierOrKeyword(P&: *this))
958	return OpenACCCannotContinue();
959
960	OpenACCClauseKind Kind = getOpenACCClauseKind(Tok: getCurToken());
961
962	if (isUnsupportedExtensionClause(Tok: getCurToken())) {
963	Diag(getCurToken(), diag::warn_acc_unsupported_extension_clause)
964	<< getCurToken().getIdentifierInfo();
965
966	// Extension methods optionally contain balanced token sequences, so we are
967	// going to parse this.
968	ConsumeToken(); // Consume the clause name.
969	BalancedDelimiterTracker Parens(*this, tok::l_paren,
970	tok::annot_pragma_openacc_end);
971	// Consume the optional parens and tokens inside of them.
972	if (!Parens.consumeOpen())
973	Parens.skipToEnd();
974
975	return OpenACCCanContinue();
976	} else if (Kind == OpenACCClauseKind::Invalid) {
977	Diag(getCurToken(), diag::err_acc_invalid_clause)
978	<< getCurToken().getIdentifierInfo();
979	return OpenACCCannotContinue();
980	}
981
982	// Consume the clause name.
983	SourceLocation ClauseLoc = ConsumeToken();
984
985	return ParseOpenACCClauseParams(ExistingClauses, DirKind, Kind, ClauseLoc);
986	}
987
988	Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
989	ArrayRef<const OpenACCClause *> ExistingClauses,
990	OpenACCDirectiveKind DirKind, OpenACCClauseKind ClauseKind,
991	SourceLocation ClauseLoc) {
992	BalancedDelimiterTracker Parens(*this, tok::l_paren,
993	tok::annot_pragma_openacc_end);
994	SemaOpenACC::OpenACCParsedClause ParsedClause(DirKind, ClauseKind, ClauseLoc);
995
996	if (ClauseHasRequiredParens(DirKind, Kind: ClauseKind)) {
997	if (Parens.expectAndConsume()) {
998	// We are missing a paren, so assume that the person just forgot the
999	// parameter. Return 'false' so we try to continue on and parse the next
1000	// clause.
1001	SkipUntil(T1: tok::comma, T2: tok::r_paren, T3: tok::annot_pragma_openacc_end,
1002	Flags: Parser::StopBeforeMatch);
1003	return OpenACCCanContinue();
1004	}
1005	ParsedClause.setLParenLoc(Parens.getOpenLocation());
1006
1007	switch (ClauseKind) {
1008	case OpenACCClauseKind::Default: {
1009	Token DefKindTok = getCurToken();
1010
1011	if (expectIdentifierOrKeyword(P&: *this)) {
1012	Parens.skipToEnd();
1013	return OpenACCCanContinue();
1014	}
1015
1016	ConsumeToken();
1017
1018	OpenACCDefaultClauseKind DefKind =
1019	getOpenACCDefaultClauseKind(Tok: DefKindTok);
1020
1021	if (DefKind == OpenACCDefaultClauseKind::Invalid) {
1022	Diag(DefKindTok, diag::err_acc_invalid_default_clause_kind);
1023	Parens.skipToEnd();
1024	return OpenACCCanContinue();
1025	}
1026
1027	ParsedClause.setDefaultDetails(DefKind);
1028	break;
1029	}
1030	case OpenACCClauseKind::If: {
1031	ExprResult CondExpr = ParseOpenACCConditionExpr();
1032	ParsedClause.setConditionDetails(CondExpr.isUsable() ? CondExpr.get()
1033	: nullptr);
1034
1035	if (CondExpr.isInvalid()) {
1036	Parens.skipToEnd();
1037	return OpenACCCanContinue();
1038	}
1039
1040	break;
1041	}
1042	case OpenACCClauseKind::Copy:
1043	case OpenACCClauseKind::PCopy:
1044	case OpenACCClauseKind::PresentOrCopy:
1045	case OpenACCClauseKind::CopyIn:
1046	case OpenACCClauseKind::PCopyIn:
1047	case OpenACCClauseKind::PresentOrCopyIn:
1048	case OpenACCClauseKind::CopyOut:
1049	case OpenACCClauseKind::PCopyOut:
1050	case OpenACCClauseKind::PresentOrCopyOut:
1051	case OpenACCClauseKind::Create:
1052	case OpenACCClauseKind::PCreate:
1053	case OpenACCClauseKind::PresentOrCreate: {
1054	OpenACCModifierKind ModList = tryParseModifierList(CK: ClauseKind);
1055	ParsedClause.setVarListDetails(VarList: ParseOpenACCVarList(DK: DirKind, CK: ClauseKind),
1056	ModKind: ModList);
1057	break;
1058	}
1059	case OpenACCClauseKind::Reduction: {
1060	// If we're missing a clause-kind (or it is invalid), see if we can parse
1061	// the var-list anyway.
1062	OpenACCReductionOperator Op = ParseReductionOperator(P&: *this);
1063	ParsedClause.setReductionDetails(
1064	Op, VarList: ParseOpenACCVarList(DK: DirKind, CK: ClauseKind));
1065	break;
1066	}
1067	case OpenACCClauseKind::Self:
1068	// The 'self' clause is a var-list instead of a 'condition' in the case of
1069	// the 'update' clause, so we have to handle it here. Use an assert to
1070	// make sure we get the right differentiator.
1071	assert(DirKind == OpenACCDirectiveKind::Update);
1072	[[fallthrough]];
1073	case OpenACCClauseKind::Device:
1074	case OpenACCClauseKind::Host:
1075	case OpenACCClauseKind::DeviceResident:
1076	case OpenACCClauseKind::Link:
1077	case OpenACCClauseKind::Attach:
1078	case OpenACCClauseKind::Delete:
1079	case OpenACCClauseKind::Detach:
1080	case OpenACCClauseKind::DevicePtr:
1081	case OpenACCClauseKind::UseDevice:
1082	case OpenACCClauseKind::FirstPrivate:
1083	case OpenACCClauseKind::NoCreate:
1084	case OpenACCClauseKind::Present:
1085	case OpenACCClauseKind::Private:
1086	ParsedClause.setVarListDetails(VarList: ParseOpenACCVarList(DK: DirKind, CK: ClauseKind),
1087	ModKind: OpenACCModifierKind::Invalid);
1088	break;
1089	case OpenACCClauseKind::Collapse: {
1090	bool HasForce = tryParseAndConsumeSpecialTokenKind(
1091	P&: *this, Kind: OpenACCSpecialTokenKind::Force, DirOrClause: ClauseKind);
1092	ExprResult LoopCount =
1093	getActions().CorrectDelayedTyposInExpr(ER: ParseConstantExpression());
1094	if (LoopCount.isInvalid()) {
1095	Parens.skipToEnd();
1096	return OpenACCCanContinue();
1097	}
1098
1099	LoopCount = getActions().OpenACC().ActOnIntExpr(
1100	DK: OpenACCDirectiveKind::Invalid, CK: ClauseKind,
1101	Loc: LoopCount.get()->getBeginLoc(), IntExpr: LoopCount.get());
1102
1103	if (LoopCount.isInvalid()) {
1104	Parens.skipToEnd();
1105	return OpenACCCanContinue();
1106	}
1107
1108	ParsedClause.setCollapseDetails(IsForce: HasForce, LoopCount: LoopCount.get());
1109	break;
1110	}
1111	case OpenACCClauseKind::Bind: {
1112	ParsedClause.setBindDetails(ParseOpenACCBindClauseArgument());
1113
1114	// We can create an 'empty' bind clause in the event of an error
1115	if (std::holds_alternative<std::monostate>(
1116	v: ParsedClause.getBindDetails())) {
1117	Parens.skipToEnd();
1118	return OpenACCCanContinue();
1119	}
1120	break;
1121	}
1122	case OpenACCClauseKind::NumGangs: {
1123	llvm::SmallVector<Expr *> IntExprs;
1124
1125	if (ParseOpenACCIntExprList(DK: OpenACCDirectiveKind::Invalid,
1126	CK: OpenACCClauseKind::NumGangs, Loc: ClauseLoc,
1127	IntExprs)) {
1128	Parens.skipToEnd();
1129	return OpenACCCanContinue();
1130	}
1131	ParsedClause.setIntExprDetails(std::move(IntExprs));
1132	break;
1133	}
1134	case OpenACCClauseKind::NumWorkers:
1135	case OpenACCClauseKind::DeviceNum:
1136	case OpenACCClauseKind::DefaultAsync:
1137	case OpenACCClauseKind::VectorLength: {
1138	ExprResult IntExpr = ParseOpenACCIntExpr(DK: OpenACCDirectiveKind::Invalid,
1139	CK: ClauseKind, Loc: ClauseLoc)
1140	.first;
1141	if (IntExpr.isInvalid()) {
1142	Parens.skipToEnd();
1143	return OpenACCCanContinue();
1144	}
1145
1146	ParsedClause.setIntExprDetails(IntExpr.get());
1147	break;
1148	}
1149	case OpenACCClauseKind::DType:
1150	case OpenACCClauseKind::DeviceType: {
1151	llvm::SmallVector<IdentifierLoc> Archs;
1152	if (getCurToken().is(K: tok::star)) {
1153	// FIXME: We want to mark that this is an 'everything else' type of
1154	// device_type in Sema.
1155	ParsedClause.setDeviceTypeDetails(
1156	{IdentifierLoc(ConsumeToken(), nullptr)});
1157	} else if (!ParseOpenACCDeviceTypeList(Archs)) {
1158	ParsedClause.setDeviceTypeDetails(std::move(Archs));
1159	} else {
1160	Parens.skipToEnd();
1161	return OpenACCCanContinue();
1162	}
1163	break;
1164	}
1165	case OpenACCClauseKind::Tile: {
1166	llvm::SmallVector<Expr *> SizeExprs;
1167	if (ParseOpenACCSizeExprList(CK: OpenACCClauseKind::Tile, SizeExprs)) {
1168	Parens.skipToEnd();
1169	return OpenACCCanContinue();
1170	}
1171
1172	ParsedClause.setIntExprDetails(std::move(SizeExprs));
1173	break;
1174	}
1175	default:
1176	llvm_unreachable("Not a required parens type?");
1177	}
1178
1179	ParsedClause.setEndLoc(getCurToken().getLocation());
1180
1181	if (Parens.consumeClose())
1182	return OpenACCCannotContinue();
1183
1184	} else if (ClauseHasOptionalParens(DirKind, Kind: ClauseKind)) {
1185	if (!Parens.consumeOpen()) {
1186	ParsedClause.setLParenLoc(Parens.getOpenLocation());
1187	switch (ClauseKind) {
1188	case OpenACCClauseKind::Self: {
1189	assert(DirKind != OpenACCDirectiveKind::Update);
1190	ExprResult CondExpr = ParseOpenACCConditionExpr();
1191	ParsedClause.setConditionDetails(CondExpr.isUsable() ? CondExpr.get()
1192	: nullptr);
1193
1194	if (CondExpr.isInvalid()) {
1195	Parens.skipToEnd();
1196	return OpenACCCanContinue();
1197	}
1198	break;
1199	}
1200	case OpenACCClauseKind::Vector:
1201	case OpenACCClauseKind::Worker: {
1202	tryParseAndConsumeSpecialTokenKind(P&: *this,
1203	Kind: ClauseKind ==
1204	OpenACCClauseKind::Vector
1205	? OpenACCSpecialTokenKind::Length
1206	: OpenACCSpecialTokenKind::Num,
1207	DirOrClause: ClauseKind);
1208	ExprResult IntExpr = ParseOpenACCIntExpr(DK: OpenACCDirectiveKind::Invalid,
1209	CK: ClauseKind, Loc: ClauseLoc)
1210	.first;
1211	if (IntExpr.isInvalid()) {
1212	Parens.skipToEnd();
1213	return OpenACCCanContinue();
1214	}
1215	ParsedClause.setIntExprDetails(IntExpr.get());
1216	break;
1217	}
1218	case OpenACCClauseKind::Async: {
1219	ExprResult AsyncArg =
1220	ParseOpenACCAsyncArgument(DK: OpenACCDirectiveKind::Invalid,
1221	CK: OpenACCClauseKind::Async, Loc: ClauseLoc)
1222	.first;
1223	ParsedClause.setIntExprDetails(AsyncArg.isUsable() ? AsyncArg.get()
1224	: nullptr);
1225	if (AsyncArg.isInvalid()) {
1226	Parens.skipToEnd();
1227	return OpenACCCanContinue();
1228	}
1229	break;
1230	}
1231	case OpenACCClauseKind::Gang: {
1232	llvm::SmallVector<OpenACCGangKind> GKs;
1233	llvm::SmallVector<Expr *> IntExprs;
1234	if (ParseOpenACCGangArgList(GangLoc: ClauseLoc, GKs, IntExprs)) {
1235	Parens.skipToEnd();
1236	return OpenACCCanContinue();
1237	}
1238	ParsedClause.setGangDetails(GKs: std::move(GKs), IntExprs: std::move(IntExprs));
1239	break;
1240	}
1241	case OpenACCClauseKind::Wait: {
1242	OpenACCWaitParseInfo Info =
1243	ParseOpenACCWaitArgument(Loc: ClauseLoc,
1244	/*IsDirective=*/false);
1245	if (Info.Failed) {
1246	Parens.skipToEnd();
1247	return OpenACCCanContinue();
1248	}
1249
1250	ParsedClause.setWaitDetails(DevNum: Info.DevNumExpr, QueuesLoc: Info.QueuesLoc,
1251	IntExprs: std::move(Info.QueueIdExprs));
1252	break;
1253	}
1254	default:
1255	llvm_unreachable("Not an optional parens type?");
1256	}
1257	ParsedClause.setEndLoc(getCurToken().getLocation());
1258	if (Parens.consumeClose())
1259	return OpenACCCannotContinue();
1260	} else {
1261	// If we have optional parens, make sure we set the end-location to the
1262	// clause, as we are a 'single token' clause.
1263	ParsedClause.setEndLoc(ClauseLoc);
1264	}
1265	} else {
1266	ParsedClause.setEndLoc(ClauseLoc);
1267	}
1268	return OpenACCSuccess(
1269	Clause: Actions.OpenACC().ActOnClause(ExistingClauses, Clause&: ParsedClause));
1270	}
1271
1272	Parser::OpenACCIntExprParseResult
1273	Parser::ParseOpenACCAsyncArgument(OpenACCDirectiveKind DK, OpenACCClauseKind CK,
1274	SourceLocation Loc) {
1275	return ParseOpenACCIntExpr(DK, CK, Loc);
1276	}
1277
1278	Parser::OpenACCWaitParseInfo
1279	Parser::ParseOpenACCWaitArgument(SourceLocation Loc, bool IsDirective) {
1280	OpenACCWaitParseInfo Result;
1281	// [devnum : int-expr : ]
1282	if (isOpenACCSpecialToken(Kind: OpenACCSpecialTokenKind::DevNum, Tok) &&
1283	NextToken().is(K: tok::colon)) {
1284	// Consume devnum.
1285	ConsumeToken();
1286	// Consume colon.
1287	ConsumeToken();
1288
1289	OpenACCIntExprParseResult Res = ParseOpenACCIntExpr(
1290	DK: IsDirective ? OpenACCDirectiveKind::Wait
1291	: OpenACCDirectiveKind::Invalid,
1292	CK: IsDirective ? OpenACCClauseKind::Invalid : OpenACCClauseKind::Wait,
1293	Loc);
1294	if (Res.first.isInvalid() &&
1295	Res.second == OpenACCParseCanContinue::Cannot) {
1296	Result.Failed = true;
1297	return Result;
1298	}
1299
1300	if (ExpectAndConsume(ExpectedTok: tok::colon)) {
1301	Result.Failed = true;
1302	return Result;
1303	}
1304
1305	Result.DevNumExpr = Res.first.get();
1306	}
1307
1308	// [ queues : ]
1309	if (isOpenACCSpecialToken(Kind: OpenACCSpecialTokenKind::Queues, Tok) &&
1310	NextToken().is(K: tok::colon)) {
1311	// Consume queues.
1312	Result.QueuesLoc = ConsumeToken();
1313	// Consume colon.
1314	ConsumeToken();
1315	}
1316
1317
1318
1319	// OpenACC 3.3, section 2.16:
1320	// the term 'async-argument' means a nonnegative scalar integer expression, or
1321	// one of the special values 'acc_async_noval' or 'acc_async_sync', as defined
1322	// in the C header file and the Fortran opacc module.
1323	OpenACCIntExprParseResult Res = ParseOpenACCAsyncArgument(
1324	DK: IsDirective ? OpenACCDirectiveKind::Wait
1325	: OpenACCDirectiveKind::Invalid,
1326	CK: IsDirective ? OpenACCClauseKind::Invalid : OpenACCClauseKind::Wait,
1327	Loc);
1328
1329	if (Res.first.isInvalid() &&
1330	Res.second == OpenACCParseCanContinue::Cannot) {
1331	Result.Failed = true;
1332	return Result;
1333	}
1334
1335	if (Res.first.isUsable())
1336	Result.QueueIdExprs.push_back(Elt: Res.first.get());
1337
1338	while (!getCurToken().isOneOf(K1: tok::r_paren, K2: tok::annot_pragma_openacc_end)) {
1339	if (ExpectAndConsume(ExpectedTok: tok::comma)) {
1340	Result.Failed = true;
1341	return Result;
1342	}
1343
1344	OpenACCIntExprParseResult Res = ParseOpenACCAsyncArgument(
1345	DK: IsDirective ? OpenACCDirectiveKind::Wait
1346	: OpenACCDirectiveKind::Invalid,
1347	CK: IsDirective ? OpenACCClauseKind::Invalid : OpenACCClauseKind::Wait,
1348	Loc);
1349
1350	if (Res.first.isInvalid() &&
1351	Res.second == OpenACCParseCanContinue::Cannot) {
1352	Result.Failed = true;
1353	return Result;
1354	}
1355
1356	if (Res.first.isUsable())
1357	Result.QueueIdExprs.push_back(Elt: Res.first.get());
1358	}
1359
1360	return Result;
1361	}
1362
1363	ExprResult Parser::ParseOpenACCIDExpression() {
1364	ExprResult Res;
1365	if (getLangOpts().CPlusPlus) {
1366	Res = ParseCXXIdExpression(/*isAddressOfOperand=*/true);
1367	} else {
1368	// There isn't anything quite the same as ParseCXXIdExpression for C, so we
1369	// need to get the identifier, then call into Sema ourselves.
1370
1371	if (Tok.isNot(K: tok::identifier)) {
1372	Diag(Tok, diag::err_expected) << tok::identifier;
1373	return ExprError();
1374	}
1375
1376	Token FuncName = getCurToken();
1377	UnqualifiedId Name;
1378	CXXScopeSpec ScopeSpec;
1379	SourceLocation TemplateKWLoc;
1380	Name.setIdentifier(Id: FuncName.getIdentifierInfo(), IdLoc: ConsumeToken());
1381
1382	// Ensure this is a valid identifier. We don't accept causing implicit
1383	// function declarations per the spec, so always claim to not have trailing
1384	// L Paren.
1385	Res = Actions.ActOnIdExpression(S: getCurScope(), SS&: ScopeSpec, TemplateKWLoc,
1386	Id&: Name, /*HasTrailingLParen=*/false,
1387	/*isAddressOfOperand=*/IsAddressOfOperand: false);
1388	}
1389
1390	return getActions().CorrectDelayedTyposInExpr(ER: Res);
1391	}
1392
1393	std::variant<std::monostate, clang::StringLiteral *, IdentifierInfo *>
1394	Parser::ParseOpenACCBindClauseArgument() {
1395	// OpenACC 3.3 section 2.15:
1396	// The bind clause specifies the name to use when calling the procedure on a
1397	// device other than the host. If the name is specified as an identifier, it
1398	// is called as if that name were specified in the language being compiled. If
1399	// the name is specified as a string, the string is used for the procedure
1400	// name unmodified.
1401	if (getCurToken().is(K: tok::r_paren)) {
1402	Diag(getCurToken(), diag::err_acc_incorrect_bind_arg);
1403	return std::monostate{};
1404	}
1405
1406	if (getCurToken().is(K: tok::identifier)) {
1407	IdentifierInfo *II = getCurToken().getIdentifierInfo();
1408	ConsumeToken();
1409	return II;
1410	}
1411
1412	if (!tok::isStringLiteral(K: getCurToken().getKind())) {
1413	Diag(getCurToken(), diag::err_acc_incorrect_bind_arg);
1414	return std::monostate{};
1415	}
1416
1417	ExprResult Res =
1418	getActions().CorrectDelayedTyposInExpr(ER: ParseStringLiteralExpression(
1419	/*AllowUserDefinedLiteral=*/false, /*Unevaluated=*/true));
1420	if (!Res.isUsable())
1421	return std::monostate{};
1422	return cast<StringLiteral>(Val: Res.get());
1423	}
1424
1425	Parser::OpenACCVarParseResult Parser::ParseOpenACCVar(OpenACCDirectiveKind DK,
1426	OpenACCClauseKind CK) {
1427	OpenACCArraySectionRAII ArraySections(*this);
1428
1429	ExprResult Res = ParseAssignmentExpression();
1430	if (!Res.isUsable())
1431	return {Res, OpenACCParseCanContinue::Cannot};
1432
1433	Res = getActions().CorrectDelayedTyposInExpr(E: Res.get());
1434	if (!Res.isUsable())
1435	return {Res, OpenACCParseCanContinue::Can};
1436
1437	Res = getActions().OpenACC().ActOnVar(DK, CK, VarExpr: Res.get());
1438
1439	return {Res, OpenACCParseCanContinue::Can};
1440	}
1441
1442	llvm::SmallVector<Expr *> Parser::ParseOpenACCVarList(OpenACCDirectiveKind DK,
1443	OpenACCClauseKind CK) {
1444	llvm::SmallVector<Expr *> Vars;
1445
1446	auto [Res, CanContinue] = ParseOpenACCVar(DK, CK);
1447	if (Res.isUsable()) {
1448	Vars.push_back(Elt: Res.get());
1449	} else if (CanContinue == OpenACCParseCanContinue::Cannot) {
1450	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end, Flags: StopBeforeMatch);
1451	return Vars;
1452	}
1453
1454	while (!getCurToken().isOneOf(K1: tok::r_paren, K2: tok::annot_pragma_openacc_end)) {
1455	ExpectAndConsume(ExpectedTok: tok::comma);
1456
1457	auto [Res, CanContinue] = ParseOpenACCVar(DK, CK);
1458
1459	if (Res.isUsable()) {
1460	Vars.push_back(Elt: Res.get());
1461	} else if (CanContinue == OpenACCParseCanContinue::Cannot) {
1462	SkipUntil(T1: tok::r_paren, T2: tok::annot_pragma_openacc_end, Flags: StopBeforeMatch);
1463	return Vars;
1464	}
1465	}
1466	return Vars;
1467	}
1468
1469	Parser::OpenACCCacheParseInfo Parser::ParseOpenACCCacheVarList() {
1470	// If this is the end of the line, just return 'false' and count on the close
1471	// paren diagnostic to catch the issue.
1472	if (getCurToken().isAnnotation())
1473	return {};
1474
1475	OpenACCCacheParseInfo CacheInfo;
1476
1477	SourceLocation ReadOnlyLoc = getCurToken().getLocation();
1478	// The VarList is an optional `readonly:` followed by a list of a variable
1479	// specifications. Consume something that looks like a 'tag', and diagnose if
1480	// it isn't 'readonly'.
1481	if (tryParseAndConsumeSpecialTokenKind(P&: *this,
1482	Kind: OpenACCSpecialTokenKind::ReadOnly,
1483	DirOrClause: OpenACCDirectiveKind::Cache))
1484	CacheInfo.ReadOnlyLoc = ReadOnlyLoc;
1485
1486	// ParseOpenACCVarList should leave us before a r-paren, so no need to skip
1487	// anything here.
1488	CacheInfo.Vars = ParseOpenACCVarList(DK: OpenACCDirectiveKind::Cache,
1489	CK: OpenACCClauseKind::Invalid);
1490
1491	return CacheInfo;
1492	}
1493
1494	Parser::OpenACCDirectiveParseInfo
1495	Parser::ParseOpenACCDirective() {
1496	SourceLocation StartLoc = ConsumeAnnotationToken();
1497	SourceLocation DirLoc = getCurToken().getLocation();
1498	OpenACCDirectiveKind DirKind = ParseOpenACCDirectiveKind(P&: *this);
1499	Parser::OpenACCWaitParseInfo WaitInfo;
1500	Parser::OpenACCCacheParseInfo CacheInfo;
1501	OpenACCAtomicKind AtomicKind = OpenACCAtomicKind::None;
1502	ExprResult RoutineName;
1503
1504	getActions().OpenACC().ActOnConstruct(K: DirKind, DirLoc);
1505
1506	// Once we've parsed the construct/directive name, some have additional
1507	// specifiers that need to be taken care of. Atomic has an 'atomic-clause'
1508	// that needs to be parsed.
1509	if (DirKind == OpenACCDirectiveKind::Atomic)
1510	AtomicKind = ParseOpenACCAtomicKind(P&: *this);
1511
1512	// We've successfully parsed the construct/directive name, however a few of
1513	// the constructs have optional parens that contain further details.
1514	BalancedDelimiterTracker T(*this, tok::l_paren,
1515	tok::annot_pragma_openacc_end);
1516
1517	if (!T.consumeOpen()) {
1518	switch (DirKind) {
1519	default:
1520	Diag(T.getOpenLocation(), diag::err_acc_invalid_open_paren);
1521	T.skipToEnd();
1522	break;
1523	case OpenACCDirectiveKind::Routine: {
1524	// Routine has an optional paren-wrapped name of a function in the local
1525	// scope. We parse the name, emitting any diagnostics
1526	RoutineName = ParseOpenACCIDExpression();
1527	// If the routine name is invalid, just skip until the closing paren to
1528	// recover more gracefully.
1529	if (!RoutineName.isUsable()) {
1530	T.skipToEnd();
1531	} else {
1532	T.consumeClose();
1533	RoutineName =
1534	getActions().OpenACC().ActOnRoutineName(RoutineName: RoutineName.get());
1535	}
1536	break;
1537	}
1538	case OpenACCDirectiveKind::Cache:
1539	CacheInfo = ParseOpenACCCacheVarList();
1540	// The ParseOpenACCCacheVarList function manages to recover from failures,
1541	// so we can always consume the close.
1542	T.consumeClose();
1543	break;
1544	case OpenACCDirectiveKind::Wait:
1545	// OpenACC has an optional paren-wrapped 'wait-argument'.
1546	WaitInfo = ParseOpenACCWaitArgument(Loc: DirLoc, /*IsDirective=*/true);
1547	if (WaitInfo.Failed)
1548	T.skipToEnd();
1549	else
1550	T.consumeClose();
1551	break;
1552	}
1553	} else if (DirKind == OpenACCDirectiveKind::Cache) {
1554	// Cache's paren var-list is required, so error here if it isn't provided.
1555	// We know that the consumeOpen above left the first non-paren here, so
1556	// diagnose, then continue as if it was completely omitted.
1557	Diag(Tok, diag::err_expected) << tok::l_paren;
1558	}
1559
1560	// Parses the list of clauses, if present, plus set up return value.
1561	OpenACCDirectiveParseInfo ParseInfo{.DirKind: DirKind,
1562	.StartLoc: StartLoc,
1563	.DirLoc: DirLoc,
1564	.LParenLoc: T.getOpenLocation(),
1565	.RParenLoc: T.getCloseLocation(),
1566	/*EndLoc=*/SourceLocation{},
1567	.MiscLoc: (DirKind == OpenACCDirectiveKind::Wait
1568	? WaitInfo.QueuesLoc
1569	: CacheInfo.ReadOnlyLoc),
1570	.AtomicKind: AtomicKind,
1571	.Exprs: {},
1572	.Clauses: {}};
1573
1574	if (DirKind == OpenACCDirectiveKind::Wait)
1575	ParseInfo.Exprs = WaitInfo.getAllExprs();
1576	else if (DirKind == OpenACCDirectiveKind::Cache)
1577	ParseInfo.Exprs = std::move(CacheInfo.Vars);
1578	else if (DirKind == OpenACCDirectiveKind::Routine && RoutineName.isUsable())
1579	ParseInfo.Exprs = llvm::SmallVector<Expr *>(1, RoutineName.get());
1580
1581	ParseInfo.Clauses = ParseOpenACCClauseList(DirKind);
1582
1583	assert(Tok.is(tok::annot_pragma_openacc_end) &&
1584	"Didn't parse all OpenACC Clauses");
1585	ParseInfo.EndLoc = ConsumeAnnotationToken();
1586	assert(ParseInfo.EndLoc.isValid() &&
1587	"Terminating annotation token not present");
1588
1589	return ParseInfo;
1590	}
1591
1592	Parser::DeclGroupPtrTy Parser::ParseOpenACCAfterRoutineDecl(
1593	AccessSpecifier &AS, ParsedAttributes &Attrs, DeclSpec::TST TagType,
1594	Decl *TagDecl, OpenACCDirectiveParseInfo &DirInfo) {
1595	assert(DirInfo.DirKind == OpenACCDirectiveKind::Routine);
1596
1597	DeclGroupPtrTy Ptr;
1598	if (DirInfo.LParenLoc.isInvalid()) {
1599	if (Tok.isNot(K: tok::r_brace) && !isEofOrEom()) {
1600	if (AS == AS_none) {
1601	// This is either an external declaration, or inside of a C struct. If
1602	// the latter, we have to diagnose if this is the 'implicit' named
1603	// version.
1604	if (TagType == DeclSpec::TST_unspecified) {
1605	ParsedAttributes EmptyDeclSpecAttrs(AttrFactory);
1606	MaybeParseCXX11Attributes(Attrs);
1607	ParsingDeclSpec PDS(*this);
1608	Ptr = ParseExternalDeclaration(DeclAttrs&: Attrs, DeclSpecAttrs&: EmptyDeclSpecAttrs, DS: &PDS);
1609	}
1610	// The only way we can have a 'none' access specifier that is in a
1611	// not-unspecified tag-type is a C struct. Member functions and
1612	// lambdas don't work in C, so we can just count on
1613	// ActonRoutineDeclDirective to recognize that Ptr is null and diagnose.
1614	} else {
1615	Ptr = ParseCXXClassMemberDeclarationWithPragmas(AS, AccessAttrs&: Attrs, TagType,
1616	Tag: TagDecl);
1617	}
1618	}
1619	}
1620
1621	return DeclGroupPtrTy::make(
1622	P: getActions().OpenACC().ActOnEndRoutineDeclDirective(
1623	StartLoc: DirInfo.StartLoc, DirLoc: DirInfo.DirLoc, LParenLoc: DirInfo.LParenLoc,
1624	ReferencedFunc: DirInfo.Exprs.empty() ? nullptr : DirInfo.Exprs.front(),
1625	RParenLoc: DirInfo.RParenLoc, Clauses: DirInfo.Clauses, EndLoc: DirInfo.EndLoc, NextDecl: Ptr));
1626	}
1627
1628	StmtResult
1629	Parser::ParseOpenACCAfterRoutineStmt(OpenACCDirectiveParseInfo &DirInfo) {
1630	assert(DirInfo.DirKind == OpenACCDirectiveKind::Routine);
1631	// We have to know the next statement for 1 of 2 reasons:
1632	// Routine without a name needs an associated DeclStmt.
1633	// Routine WITH a name needs to see if it is a DeclStmt to diagnose.
1634	StmtResult NextStmt = StmtEmpty();
1635
1636	// Parse the next statement in the 'implicit' case, not in the 'named' case.
1637	// In the 'named' case we will use the creation of the next decl to determine
1638	// whether we should warn.
1639	if (DirInfo.LParenLoc.isInvalid()) {
1640	ParsingOpenACCDirectiveRAII DirScope(*this, /*Value=*/false);
1641	NextStmt = ParseStatement();
1642	}
1643
1644	return getActions().OpenACC().ActOnEndRoutineStmtDirective(
1645	StartLoc: DirInfo.StartLoc, DirLoc: DirInfo.DirLoc, LParenLoc: DirInfo.LParenLoc,
1646	ReferencedFunc: DirInfo.Exprs.empty() ? nullptr : DirInfo.Exprs.front(),
1647	RParenLoc: DirInfo.RParenLoc, Clauses: DirInfo.Clauses, EndLoc: DirInfo.EndLoc, NextStmt: NextStmt.get());
1648	}
1649
1650	Parser::DeclGroupPtrTy
1651	Parser::ParseOpenACCDirectiveDecl(AccessSpecifier &AS, ParsedAttributes &Attrs,
1652	DeclSpec::TST TagType, Decl *TagDecl) {
1653	assert(Tok.is(tok::annot_pragma_openacc) && "expected OpenACC Start Token");
1654
1655	ParsingOpenACCDirectiveRAII DirScope(*this);
1656
1657	OpenACCDirectiveParseInfo DirInfo = ParseOpenACCDirective();
1658
1659	if (getActions().OpenACC().ActOnStartDeclDirective(
1660	K: DirInfo.DirKind, StartLoc: DirInfo.StartLoc, Clauses: DirInfo.Clauses))
1661	return nullptr;
1662
1663	if (DirInfo.DirKind == OpenACCDirectiveKind::Routine)
1664	return ParseOpenACCAfterRoutineDecl(AS, Attrs, TagType, TagDecl, DirInfo);
1665
1666	return DeclGroupPtrTy::make(P: getActions().OpenACC().ActOnEndDeclDirective(
1667	K: DirInfo.DirKind, StartLoc: DirInfo.StartLoc, DirLoc: DirInfo.DirLoc, LParenLoc: DirInfo.LParenLoc,
1668	RParenLoc: DirInfo.RParenLoc, EndLoc: DirInfo.EndLoc, Clauses: DirInfo.Clauses));
1669	}
1670
1671	StmtResult Parser::ParseOpenACCDirectiveStmt() {
1672	assert(Tok.is(tok::annot_pragma_openacc) && "expected OpenACC Start Token");
1673
1674	ParsingOpenACCDirectiveRAII DirScope(*this);
1675
1676	OpenACCDirectiveParseInfo DirInfo = ParseOpenACCDirective();
1677	if (getActions().OpenACC().ActOnStartStmtDirective(
1678	K: DirInfo.DirKind, StartLoc: DirInfo.StartLoc, Clauses: DirInfo.Clauses))
1679	return StmtError();
1680
1681	if (DirInfo.DirKind == OpenACCDirectiveKind::Routine)
1682	return ParseOpenACCAfterRoutineStmt(DirInfo);
1683
1684	StmtResult AssocStmt;
1685	if (doesDirectiveHaveAssociatedStmt(DirKind: DirInfo.DirKind)) {
1686	SemaOpenACC::AssociatedStmtRAII AssocStmtRAII(
1687	getActions().OpenACC(), DirInfo.DirKind, DirInfo.DirLoc, {},
1688	DirInfo.Clauses);
1689	ParsingOpenACCDirectiveRAII DirScope(*this, /*Value=*/false);
1690	ParseScope ACCScope(this, getOpenACCScopeFlags(DirKind: DirInfo.DirKind));
1691
1692	AssocStmt = getActions().OpenACC().ActOnAssociatedStmt(
1693	DirectiveLoc: DirInfo.StartLoc, K: DirInfo.DirKind, AtKind: DirInfo.AtomicKind, Clauses: DirInfo.Clauses,
1694	AssocStmt: ParseStatement());
1695	}
1696
1697	return getActions().OpenACC().ActOnEndStmtDirective(
1698	K: DirInfo.DirKind, StartLoc: DirInfo.StartLoc, DirLoc: DirInfo.DirLoc, LParenLoc: DirInfo.LParenLoc,
1699	MiscLoc: DirInfo.MiscLoc, Exprs: DirInfo.Exprs, AK: DirInfo.AtomicKind, RParenLoc: DirInfo.RParenLoc,
1700	EndLoc: DirInfo.EndLoc, Clauses: DirInfo.Clauses, AssocStmt);
1701	}
1702