[OpenACC] Implement Atomic construct variants #73015
Merged
Conversation
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
`atomic` is required to be followed by a special `atomic clause`, so this patch manages the parsing of that. We are representing each of the variants of the atomic construct as separate kinds, because they have distinct rules/application/etc, and this should make it easier to check rules in the future.
llvmbot
added
clang
|
@llvm/pr-subscribers-clang Author: Erich Keane (erichkeane) Changes
Full diff: https://github.com/llvm/llvm-project/pull/73015.diff 4 Files Affected:
diff --git a/clang/include/clang/Basic/DiagnosticParseKinds.td b/clang/include/clang/Basic/DiagnosticParseKinds.td
index 54b5ba6e6414b2d..1e307f481d6c9ad 100644
--- a/clang/include/clang/Basic/DiagnosticParseKinds.td
+++ b/clang/include/clang/Basic/DiagnosticParseKinds.td
@@ -1364,6 +1364,9 @@ 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_atomic_clause
+ : Error<"%select{missing|invalid}0 OpenACC 'atomic-clause'%select{| "
+ "'%1'}0; expected 'read', 'write', 'update', or 'capture'">;
// OpenMP support.
def warn_pragma_omp_ignored : Warning<
diff --git a/clang/include/clang/Basic/OpenACCKinds.h b/clang/include/clang/Basic/OpenACCKinds.h
index 2a818638720abb0..1622f8f00274ad6 100644
--- a/clang/include/clang/Basic/OpenACCKinds.h
+++ b/clang/include/clang/Basic/OpenACCKinds.h
@@ -41,7 +41,13 @@ enum class OpenACCDirectiveKind {
SerialLoop,
KernelsLoop,
- // FIXME: atomic Construct variants.
+ // Atomic Construct. The OpenACC standard considers these as a single
+ // construct, however the atomic-clause (read, write, update, capture) are
+ // important for legalization of the application of this to statements/blocks.
+ AtomicRead,
+ AtomicWrite,
+ AtomicUpdate,
+ AtomicCapture,
// Declare Directive.
Declare,
diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp
index a0f8fa97f6fa701..8a4f7a5da636e30 100644
--- a/clang/lib/Parse/ParseOpenACC.cpp
+++ b/clang/lib/Parse/ParseOpenACC.cpp
@@ -29,7 +29,8 @@ enum class OpenACCDirectiveKindEx {
// 'enter data' and 'exit data'
Enter,
Exit,
- // FIXME: Atomic Variants
+ // 'atomic read', 'atomic write', 'atomic update', and 'atomic capture'.
+ Atomic,
};
// Translate single-token string representations to the OpenACC Directive Kind.
@@ -59,9 +60,21 @@ OpenACCDirectiveKindEx getOpenACCDirectiveKind(StringRef Name) {
return llvm::StringSwitch<OpenACCDirectiveKindEx>(Name)
.Case("enter", OpenACCDirectiveKindEx::Enter)
.Case("exit", OpenACCDirectiveKindEx::Exit)
+ .Case("atomic", OpenACCDirectiveKindEx::Atomic)
.Default(OpenACCDirectiveKindEx::Invalid);
}
+// Since 'atomic' is effectively a compound directive, this will decode the
+// second part of the directive.
+OpenACCDirectiveKind getOpenACCAtomicDirectiveKind(StringRef Name) {
+ return llvm::StringSwitch<OpenACCDirectiveKind>(Name)
+ .Case("read", OpenACCDirectiveKind::AtomicRead)
+ .Case("write", OpenACCDirectiveKind::AtomicWrite)
+ .Case("update", OpenACCDirectiveKind::AtomicUpdate)
+ .Case("capture", OpenACCDirectiveKind::AtomicCapture)
+ .Default(OpenACCDirectiveKind::Invalid);
+}
+
bool isOpenACCDirectiveKind(OpenACCDirectiveKind Kind, StringRef Tok) {
switch (Kind) {
case OpenACCDirectiveKind::Parallel:
@@ -82,6 +95,10 @@ bool isOpenACCDirectiveKind(OpenACCDirectiveKind Kind, StringRef Tok) {
case OpenACCDirectiveKind::KernelsLoop:
case OpenACCDirectiveKind::EnterData:
case OpenACCDirectiveKind::ExitData:
+ case OpenACCDirectiveKind::AtomicRead:
+ case OpenACCDirectiveKind::AtomicWrite:
+ case OpenACCDirectiveKind::AtomicUpdate:
+ case OpenACCDirectiveKind::AtomicCapture:
return false;
case OpenACCDirectiveKind::Declare:
@@ -126,6 +143,28 @@ ParseOpenACCEnterExitDataDirective(Parser &P, Token FirstTok,
: OpenACCDirectiveKind::ExitData;
}
+OpenACCDirectiveKind ParseOpenACCAtomicDirective(Parser &P) {
+ Token AtomicClauseToken = P.getCurToken();
+
+ if (AtomicClauseToken.isAnnotation()) {
+ P.Diag(AtomicClauseToken, diag::err_acc_invalid_atomic_clause) << 0;
+ return OpenACCDirectiveKind::Invalid;
+ }
+
+ std::string AtomicClauseSpelling =
+ P.getPreprocessor().getSpelling(AtomicClauseToken);
+
+ OpenACCDirectiveKind DirKind =
+ getOpenACCAtomicDirectiveKind(AtomicClauseSpelling);
+
+ if (DirKind == OpenACCDirectiveKind::Invalid)
+ P.Diag(AtomicClauseToken, diag::err_acc_invalid_atomic_clause)
+ << 1 << AtomicClauseSpelling;
+
+ P.ConsumeToken();
+ return DirKind;
+}
+
// Parse and consume the tokens for OpenACC Directive/Construct kinds.
OpenACCDirectiveKind ParseOpenACCDirectiveKind(Parser &P) {
Token FirstTok = P.getCurToken();
@@ -158,6 +197,8 @@ OpenACCDirectiveKind ParseOpenACCDirectiveKind(Parser &P) {
case OpenACCDirectiveKindEx::Exit:
return ParseOpenACCEnterExitDataDirective(P, FirstTok, FirstTokSpelling,
ExDirKind);
+ case OpenACCDirectiveKindEx::Atomic:
+ return ParseOpenACCAtomicDirective(P);
}
}
diff --git a/clang/test/ParserOpenACC/parse-constructs.c b/clang/test/ParserOpenACC/parse-constructs.c
index a5270daf6034cf8..28410cff7f15d44 100644
--- a/clang/test/ParserOpenACC/parse-constructs.c
+++ b/clang/test/ParserOpenACC/parse-constructs.c
@@ -94,6 +94,37 @@ void func() {
#pragma acc kernels loop
for(;;){}
+ int i = 0, j = 0, k = 0;
+ // expected-error@+2{{missing OpenACC 'atomic-clause'; expected 'read', 'write', 'update', or 'capture'}}
+ // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc atomic
+ i = j;
+ // expected-error@+2{{invalid OpenACC 'atomic-clause' 'garbage'; expected 'read', 'write', 'update', or 'capture'}}
+ // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc atomic garbage
+ i = j;
+ // expected-warning@+3{{OpenACC clause parsing not yet implemented}}
+ // expected-error@+2{{invalid OpenACC 'atomic-clause' 'garbage'; expected 'read', 'write', 'update', or 'capture'}}
+ // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc atomic garbage clause list
+ i = j;
+ // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc atomic read
+ i = j;
+ // expected-warning@+2{{OpenACC clause parsing not yet implemented}}
+ // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc atomic write clause list
+ i = i + j;
+ // expected-warning@+2{{OpenACC clause parsing not yet implemented}}
+ // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc atomic update clause list
+ i++;
+ // expected-warning@+2{{OpenACC clause parsing not yet implemented}}
+ // expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
+#pragma acc atomic capture clause list
+ i = j++;
+
+
// expected-warning@+2{{OpenACC clause parsing not yet implemented}}
// expected-warning@+1{{OpenACC directives not yet implemented, pragma ignored}}
#pragma acc declare clause list
|
After review, it was pointed out that this is a list that has changed recently in OpenMP and thus might change here. This patch extracts the parsing of atomic-clause to happen after directive determination.
| // this as a clause list, which, despite being invalid, is likely what the | ||
| // user was trying to do. | ||
| if (AtomicKind == OpenACCAtomicKind::Invalid) | ||
| return OpenACCAtomicKind::Update; |
There was a problem hiding this comment.
Shall you consume token here before return?
There was a problem hiding this comment.
I don't believe so, since this is not diagnosing on it, I'm treating it as just #pragma acc atomic and treating the 'rest' as a 'clause-list'. That way we can diagnose the rest as a 'clauses not allowed here'.
alexey-bataev
approved these changes
Nov 21, 2023
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
atomicis required to be followed by a specialatomic clause, so this patch manages the parsing of that. We are representing each of the variants of the atomic construct as separate kinds, because they have distinct rules/application/etc, and this should make it easier to check rules in the future.