[clang] add Diag -Wasm-volatile for implied volatile asm stmts

[nickdesaulniers]

Summary

According to the GCC docs on inline asm
https://gcc.gnu.org/onlinedocs/gcc/Extended-Asm.html#Volatile

asm statements that have no output operands and asm goto statements,
are implicitly volatile.

We tend to see the volatile qualifier tacked on to asm statements "just
to be safe" when in reality it makes no difference in IR for asm goto
statements or output-less asm. The sideeffect flag is set on such IR
instructions.

Really the volatile asm-qualifier exists only to signal that an asm
statement should not be DCE'd (when it has outputs but they are unused),
CSE'd, or LICM'd. It is not a general compiler barrier.

Create a new warning, -Wasm-volatile and warn in either case.

[nickdesaulniers]

clang/lib/Parse/ParseStmtAsm.cpp

794

Loc is not correct if volatile is not the first asm-qualifier.

[nickdesaulniers]

clang/include/clang/Basic/DiagnosticParseKinds.td

31

should this go in clang/include/clang/Basic/DiagnosticSemaKinds.td instead?

[nickdesaulniers]

• fix volatile location, more tests

[jyknight]

I agree that's the expected semantics. I think those semantics are unfortunate, but they're not gonna change. IMO it would've been better if you had to opt-in to "no side effects" via __attribute__((const)) or so.

But I wonder why you think we should be discouraging people from writing "asm volatile" for things that are implicitly treated as having side-effects? It's quite non-obvious that adding an output argument to an existing asm statement changes whether it's treated as having side-effects, and I don't think it improves a codebase to make people lay out that trap more more frequently.

[MaskRay]

The volatile qualifier is implied but makes the intention explicit. I agree that the user should not be punished (-Wasm-volatile) by writing asm volatile("int3") instead of asm("int3").

[segher]

In D118297#3285571, @MaskRay wrote:

The volatile qualifier is implied but makes the intention explicit. I agree that the user should not be punished (-Wasm-volatile) by writing asm volatile("int3") instead of asm("int3").

Not punished no. But it is useful education (like all the other "statement without effect"
etc. warnings).

clang/include/clang/Basic/DiagnosticGroups.td

 
 def ObjCPotentiallyDirectSelector : DiagGroup<"potentially-direct-selector">;
 def ObjCStrictPotentiallyDirectSelector :
   DiagGroup<"strict-potentially-direct-selector",
             [ObjCPotentiallyDirectSelector]>;
 
 // Inline ASM warnings.
 def ASMOperandWidths : DiagGroup<"asm-operand-widths">;
+def ASMVolatile : DiagGroup<"asm-volatile">;
 def ASM : DiagGroup<"asm", [
-    ASMOperandWidths
+    ASMOperandWidths,
+    ASMVolatile
   ]>;
 
 // Linker warnings.
 def LinkerWarnings : DiagGroup<"linker-warnings">;
 
 // OpenMP warnings.
 def SourceUsesOpenMP : DiagGroup<"source-uses-openmp">;
 def OpenMPClauses : DiagGroup<"openmp-clauses">;

clang/include/clang/Basic/DiagnosticParseKinds.td

 def err_inline_ms_asm_parsing : Error<"%0">;
 def err_msasm_unsupported_arch : Error<
   "unsupported architecture '%0' for MS-style inline assembly">;
 def err_msasm_unable_to_create_target : Error<
   "MS-style inline assembly is not available: %0">;
 def err_gnu_inline_asm_disabled : Error<
   "GNU-style inline assembly is disabled">;
 def err_asm_duplicate_qual : Error<"duplicate asm qualifier '%0'">;
+def warn_asm_volatile_goto : Warning<"volatile qualifier implied by goto">, InGroup<ASMVolatile>;
+def warn_asm_volatile : Warning<"volatile qualifier implied by lack of outputs">, InGroup<ASMVolatile>;
 }
 
 let CategoryName = "Parse Issue" in {
 
 def ext_empty_translation_unit : Extension<
   "ISO C requires a translation unit to contain at least one declaration">,
   InGroup<DiagGroup<"empty-translation-unit">>;
 def warn_cxx98_compat_top_level_semi : Warning<

clang/include/clang/Parse/Parser.h

private:
   void CodeCompletePreprocessorExpression() override;
   void CodeCompleteMacroArgument(IdentifierInfo *Macro, MacroInfo *MacroInfo,
                                  unsigned ArgumentIndex) override;
   void CodeCompleteIncludedFile(llvm::StringRef Dir, bool IsAngled) override;
   void CodeCompleteNaturalLanguage() override;
 
   class GNUAsmQualifiers {
     unsigned Qualifiers = AQ_unspecified;
+    SourceLocation VolatileLoc;
 
   public:
     enum AQ {
       AQ_unspecified = 0,
       AQ_volatile    = 1,
       AQ_inline      = 2,
       AQ_goto        = 4,
     };
     static const char *getQualifierName(AQ Qualifier);
     bool setAsmQualifier(AQ Qualifier);
-    inline bool isVolatile() const { return Qualifiers & AQ_volatile; };
-    inline bool isInline() const { return Qualifiers & AQ_inline; };
+    void setGNUAsmQualifierVolatileLoc(SourceLocation Loc) {
+      VolatileLoc = Loc;
+    }
+    SourceLocation getGNUAsmQualifierVolatileLoc() const { return VolatileLoc; }
+    inline bool isVolatile() const { return Qualifiers & AQ_volatile; }
+    inline bool isInline() const { return Qualifiers & AQ_inline; }
     inline bool isGoto() const { return Qualifiers & AQ_goto; }
   };
   bool isGCCAsmStatement(const Token &TokAfterAsm) const;
   bool isGNUAsmQualifier(const Token &TokAfterAsm) const;
   GNUAsmQualifiers::AQ getGNUAsmQualifier(const Token &Tok) const;
   bool parseGNUAsmQualifierListOpt(GNUAsmQualifiers &AQ);
 };
 
 }  // end namespace clang
 
 #endif

clang/lib/Headers/x86gprintrin.h

 #endif
 
 #if !(defined(_MSC_VER) || defined(__SCE__)) || __has_feature(modules) ||      \
     defined(__CRC32__)
 #include <crc32intrin.h>
 #endif
 
 #define __SSC_MARK(Tag)                                                        \
-  __asm__ __volatile__("mov {%%ebx, %%eax|eax, ebx}; "                      \
-                       "mov {%0, %%ebx|ebx, %0}; "                          \
-                       ".byte 0x64, 0x67, 0x90; "                              \
-                       "mov {%%eax, %%ebx|ebx, eax};" ::"i"(Tag)            \
-                       : "%eax");
+  __asm__("mov {%%ebx, %%eax|eax, ebx}; "                                      \
+          "mov {%0, %%ebx|ebx, %0}; "                                          \
+          ".byte 0x64, 0x67, 0x90; "                                           \
+          "mov {%%eax, %%ebx|ebx, eax};" ::"i"(Tag)                            \
 
 #endif /* __X86GPRINTRIN_H */

clang/lib/Parse/ParseStmtAsm.cpp

while (true) {
     if (A == GNUAsmQualifiers::AQ_unspecified) {
       if (Tok.isNot(tok::l_paren)) {
         Diag(Tok.getLocation(), diag::err_asm_qualifier_ignored);
         SkipUntil(tok::r_paren, StopAtSemi);
         return true;
       }
       return false;
     }
+    if (A == GNUAsmQualifiers::AQ_volatile)
+      AQ.setGNUAsmQualifierVolatileLoc(Tok.getLocation());
     if (AQ.setAsmQualifier(A))
       Diag(Tok.getLocation(), diag::err_asm_duplicate_qual)
           << GNUAsmQualifiers::getQualifierName(A);
     ConsumeToken();
   }
   return false;
 }
 
StmtResult Parser::ParseAsmStatement(bool &msAsm) {
   }
 
   SmallVector<IdentifierInfo *, 4> Names;
   ExprVector Constraints;
   ExprVector Exprs;
   ExprVector Clobbers;
 
   if (Tok.is(tok::r_paren)) {
-    // We have a simple asm expression like 'asm("foo")'.
+    // We have a simple asm expression like 'asm [asm-qualifier-list] ("foo")',
+    // no outputs, inputs, or labels.
+    if (GAQ.isVolatile())
+      Diag(GAQ.getGNUAsmQualifierVolatileLoc(), diag::warn_asm_volatile);
+
     T.consumeClose();
     return Actions.ActOnGCCAsmStmt(
         AsmLoc, /*isSimple*/ true, GAQ.isVolatile(),
         /*NumOutputs*/ 0, /*NumInputs*/ 0, nullptr, Constraints, Exprs,
         AsmString.get(), Clobbers, /*NumLabels*/ 0, T.getCloseLocation());
   }
 
   // Parse Outputs, if present.
   bool AteExtraColon = false;
   if (Tok.is(tok::colon) || Tok.is(tok::coloncolon)) {
     // In C++ mode, parse "::" like ": :".
     AteExtraColon = Tok.is(tok::coloncolon);
     ConsumeToken();
 
     if (!AteExtraColon && ParseAsmOperandsOpt(Names, Constraints, Exprs))
       return StmtError();
   }
 
   unsigned NumOutputs = Names.size();
 
+  if (GAQ.isVolatile()) {
+    if (GAQ.isGoto())
+      Diag(GAQ.getGNUAsmQualifierVolatileLoc(), diag::warn_asm_volatile_goto);
+    else if (NumOutputs == 0)
+      Diag(GAQ.getGNUAsmQualifierVolatileLoc(), diag::warn_asm_volatile);
+  }
+
   // Parse Inputs, if present.
   if (AteExtraColon || Tok.is(tok::colon) || Tok.is(tok::coloncolon)) {
     // In C++ mode, parse "::" like ": :".
     if (AteExtraColon)
       AteExtraColon = false;
     else {
       AteExtraColon = Tok.is(tok::coloncolon);
       ConsumeToken();

clang/test/CodeGen/X86/amx_inline_asm.c

 // RUN: %clang_cc1 %s -ffreestanding -triple=x86_64-unknown-unknown  -target-feature +amx-int8 -target-feature +amx-bf16 -emit-llvm -o - -Wall -Werror -pedantic | FileCheck %s
 
 void f_tilemul(short a)
 {
   //CHECK:  call void asm sideeffect "tileloadd 0(%rsi,%r13,4), %tmm0   \0A\09tileloadd 0(%rdx,%r14,4), %tmm6   \0A\09tdpbf16ps %tmm6, %tmm0, %tmm7    \0A\09tilestored %tmm7, 0(%r12,%r15,4) \0A\09", "~{memory},~{tmm0},~{tmm6},~{tmm7},~{dirflag},~{fpsr},~{flags}"()
-  __asm__ volatile ("tileloadd 0(%%rsi,%%r13,4), %%tmm0   \n\t"
+  __asm__ ("tileloadd 0(%%rsi,%%r13,4), %%tmm0   \n\t"
                     "tileloadd 0(%%rdx,%%r14,4), %%tmm6   \n\t"
                     "tdpbf16ps %%tmm6, %%tmm0, %%tmm7    \n\t"
                     "tilestored %%tmm7, 0(%%r12,%%r15,4) \n\t"
           ::: "memory", "tmm0", "tmm6", "tmm7");
 }

clang/test/CodeGen/X86/x86-64-inline-asm.c

 // REQUIRES: x86-registered-target
 // RUN: %clang_cc1 -triple x86_64 %s -S -o /dev/null -DWARN -verify
 // RUN: %clang_cc1 -triple x86_64 %s -S -o /dev/null -Werror -verify
 // RUN: %clang_cc1 -triple x86_64-linux-gnu %s -S -o - | FileCheck %s
 void f() {
   asm("movaps %xmm3, (%esi, 2)");
 // expected-note@1 {{instantiated into assembly here}}
 #ifdef WARN
 // expected-warning@-3 {{scale factor without index register is ignored}}
 #else
 // expected-error@-5 {{scale factor without index register is ignored}}
 #endif
 }
 
 static unsigned var[1] = {};
-void g(void) { asm volatile("movd %%xmm0, %0"
+void g(void) { asm ("movd %%xmm0, %0"
                             :
                             : "m"(var)); }
 
 void pr40890(void) {
   struct s {
     int a, b;
   } s;
-  __asm__ __volatile__("\n#define S_A abcd%0\n" : : "n"(&((struct s*)0)->a));
-  __asm__ __volatile__("\n#define S_B abcd%0\n" : : "n"(&((struct s*)0)->b));
-  __asm__ __volatile__("\n#define BEEF abcd%0\n" : : "n"((int*)0xdeadbeeeeeef));
-  __asm__ __volatile__("movabsq %0, %%rax" : : "n"(4624529908474429119));
+  __asm__ ("\n#define S_A abcd%0\n" : : "n"(&((struct s*)0)->a));
+  __asm__ ("\n#define S_B abcd%0\n" : : "n"(&((struct s*)0)->b));
+  __asm__ ("\n#define BEEF abcd%0\n" : : "n"((int*)0xdeadbeeeeeef));
+  __asm__ ("movabsq %0, %%rax" : : "n"(4624529908474429119));
 
 // CHECK-LABEL: pr40890
 // CHECK: #define S_A abcd$0
 // CHECK: #define S_B abcd$4
 // CHECK: #define BEEF abcd$244837814038255
 // CHECK: movabsq $4624529908474429119, %rax
 }

clang/test/CodeGen/X86/x86_32-inline-asm.c

 __m128 val128;
 __m256 val256;
 __m512 val512;
 
 int func1() {
   // Error out if size is > 32-bits.
   uint32_t msr = 0x8b;
   uint64_t val = 0;
-  __asm__ volatile("wrmsr"
+  __asm__ ("wrmsr"
                    :
                    : "c" (msr),
                      "a" ((val & 0xFFFFFFFFUL)), // expected-error {{invalid input size for constraint 'a'}}
                      "d" (((val >> 32) & 0xFFFFFFFFUL)));
 
   // Don't error out if the size of the destination is <= 32 bits.
   unsigned char data;
   unsigned int port;
-  __asm__ volatile("outb %0, %w1" : : "a" (data), "Nd" (port)); // No error expected.
+  __asm__ ("outb %0, %w1" : : "a" (data), "Nd" (port)); // No error expected.
 
-  __asm__ volatile("outb %0, %w1" : : "R" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'R'}}
-  __asm__ volatile("outb %0, %w1" : : "q" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'q'}}
-  __asm__ volatile("outb %0, %w1" : : "Q" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'Q'}}
-  __asm__ volatile("outb %0, %w1" : : "b" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'b'}}
-  __asm__ volatile("outb %0, %w1" : : "c" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'c'}}
-  __asm__ volatile("outb %0, %w1" : : "d" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'd'}}
-  __asm__ volatile("outb %0, %w1" : : "S" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'S'}}
-  __asm__ volatile("outb %0, %w1" : : "D" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'D'}}
-  __asm__ volatile("foo1 %0" : : "A" (val128)); // expected-error {{invalid input size for constraint 'A'}}
-  __asm__ volatile("foo1 %0" : : "f" (val256)); // expected-error {{invalid input size for constraint 'f'}}
-  __asm__ volatile("foo1 %0" : : "t" (val256)); // expected-error {{invalid input size for constraint 't'}}
-  __asm__ volatile("foo1 %0" : : "u" (val256)); // expected-error {{invalid input size for constraint 'u'}}
-  __asm__ volatile("foo1 %0" : : "x" (val512)); // expected-error {{invalid input size for constraint 'x'}}
+  __asm__ ("outb %0, %w1" : : "R" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'R'}}
+  __asm__ ("outb %0, %w1" : : "q" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'q'}}
+  __asm__ ("outb %0, %w1" : : "Q" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'Q'}}
+  __asm__ ("outb %0, %w1" : : "b" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'b'}}
+  __asm__ ("outb %0, %w1" : : "c" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'c'}}
+  __asm__ ("outb %0, %w1" : : "d" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'd'}}
+  __asm__ ("outb %0, %w1" : : "S" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'S'}}
+  __asm__ ("outb %0, %w1" : : "D" (val), "Nd" (port)); // expected-error {{invalid input size for constraint 'D'}}
+  __asm__ ("foo1 %0" : : "A" (val128)); // expected-error {{invalid input size for constraint 'A'}}
+  __asm__ ("foo1 %0" : : "f" (val256)); // expected-error {{invalid input size for constraint 'f'}}
+  __asm__ ("foo1 %0" : : "t" (val256)); // expected-error {{invalid input size for constraint 't'}}
+  __asm__ ("foo1 %0" : : "u" (val256)); // expected-error {{invalid input size for constraint 'u'}}
+  __asm__ ("foo1 %0" : : "x" (val512)); // expected-error {{invalid input size for constraint 'x'}}
 
   __asm__ volatile("foo1 %0" : "=R" (val)); // expected-error {{invalid output size for constraint '=R'}}
   __asm__ volatile("foo1 %0" : "=q" (val)); // expected-error {{invalid output size for constraint '=q'}}
   __asm__ volatile("foo1 %0" : "=Q" (val)); // expected-error {{invalid output size for constraint '=Q'}}
   __asm__ volatile("foo1 %0" : "=a" (val)); // expected-error {{invalid output size for constraint '=a'}}
   __asm__ volatile("foo1 %0" : "=b" (val)); // expected-error {{invalid output size for constraint '=b'}}
   __asm__ volatile("foo1 %0" : "=c" (val)); // expected-error {{invalid output size for constraint '=c'}}
   __asm__ volatile("foo1 %0" : "=d" (val)); // expected-error {{invalid output size for constraint '=d'}}
   __asm__ volatile("foo1 %0" : "=S" (val)); // expected-error {{invalid output size for constraint '=S'}}
   __asm__ volatile("foo1 %0" : "=D" (val)); // expected-error {{invalid output size for constraint '=D'}}
   __asm__ volatile("foo1 %0" : "=A" (val128)); // expected-error {{invalid output size for constraint '=A'}}
   __asm__ volatile("foo1 %0" : "=t" (val256)); // expected-error {{invalid output size for constraint '=t'}}
   __asm__ volatile("foo1 %0" : "=u" (val256)); // expected-error {{invalid output size for constraint '=u'}}
   __asm__ volatile("foo1 %0" : "=x" (val512)); // expected-error {{invalid output size for constraint '=x'}}
 
 #ifdef __AVX__
-  __asm__ volatile("foo1 %0" : : "x" (val256)); // No error.
+  __asm__ ("foo1 %0" : : "x" (val256)); // No error.
   __asm__ volatile("foo1 %0" : "=x" (val256));  // No error.
 #else
-  __asm__ volatile("foo1 %0" : : "x" (val256)); // expected-error {{invalid input size for constraint 'x'}}
+  __asm__ ("foo1 %0" : : "x" (val256)); // expected-error {{invalid input size for constraint 'x'}}
   __asm__ volatile("foo1 %0" : "=x" (val256)); // expected-error {{invalid output size for constraint '=x'}}
 #endif
 }
 
 int __attribute__((__target__("sse"))) _func2() {
-  __asm__ volatile("foo1 %0" : : "x" (val128)); // No error.
+  __asm__ ("foo1 %0" : : "x" (val128)); // No error.
   __asm__ volatile("foo1 %0" : "=x" (val128));  // No error.
 #ifdef __AVX__
-  __asm__ volatile("foo1 %0" : : "x" (val256)); // No error.
+  __asm__ ("foo1 %0" : : "x" (val256)); // No error.
   __asm__ volatile("foo1 %0" : "=x" (val256));  // No error.
 #else
-  __asm__ volatile("foo1 %0" : : "x" (val256)); // expected-error {{invalid input size for constraint 'x'}}
+  __asm__ ("foo1 %0" : : "x" (val256)); // expected-error {{invalid input size for constraint 'x'}}
   __asm__ volatile("foo1 %0" : "=x" (val256)); // expected-error {{invalid output size for constraint '=x'}}
 #endif
-  __asm__ volatile("foo1 %0" : : "x" (val512)); // expected-error {{invalid input size for constraint 'x'}}
+  __asm__ ("foo1 %0" : : "x" (val512)); // expected-error {{invalid input size for constraint 'x'}}
   __asm__ volatile("foo1 %0" : "=x" (val512)); // expected-error {{invalid output size for constraint '=x'}}
 }
 
 int __attribute__((__target__("avx"))) _func3() {
-  __asm__ volatile("foo1 %0" : : "x" (val128)); // No error.
+  __asm__ ("foo1 %0" : : "x" (val128)); // No error.
   __asm__ volatile("foo1 %0" : "=x" (val128));  // No error.
-  __asm__ volatile("foo1 %0" : : "x" (val256)); // No error.
+  __asm__ ("foo1 %0" : : "x" (val256)); // No error.
   __asm__ volatile("foo1 %0" : "=x" (val256));  // No error.
-  __asm__ volatile("foo1 %0" : : "x" (val512)); // expected-error {{invalid input size for constraint 'x'}}
+  __asm__ ("foo1 %0" : : "x" (val512)); // expected-error {{invalid input size for constraint 'x'}}
   __asm__ volatile("foo1 %0" : "=x" (val512)); // expected-error {{invalid output size for constraint '=x'}}
 }
 
 int __attribute__((__target__("avx512f"))) _func4() {
-  __asm__ volatile("foo1 %0" : : "x" (val128)); // No error.
+  __asm__ ("foo1 %0" : : "x" (val128)); // No error.
   __asm__ volatile("foo1 %0" : "=x" (val128));  // No error.
-  __asm__ volatile("foo1 %0" : : "x" (val256)); // No error.
+  __asm__ ("foo1 %0" : : "x" (val256)); // No error.
   __asm__ volatile("foo1 %0" : "=x" (val256));  // No error.
-  __asm__ volatile("foo1 %0" : : "x" (val512)); // No error.
+  __asm__ ("foo1 %0" : : "x" (val512)); // No error.
   __asm__ volatile("foo1 %0" : "=x" (val512)); // No error.
 }

clang/test/CodeGen/avr-unsupported-inline-asm-constraints.c

 // RUN: %clang_cc1 -triple avr-unknown-unknown -verify %s
 
 const unsigned char val = 0;
 
 int foo() {
-  __asm__ volatile("foo %0, 1" : : "fo" (val)); // expected-error {{invalid input constraint 'fo' in asm}}
-  __asm__ volatile("foo %0, 1" : : "Nd" (val)); // expected-error {{invalid input constraint 'Nd' in asm}}
+  __asm__ ("foo %0, 1" : : "fo" (val)); // expected-error {{invalid input constraint 'fo' in asm}}
+  __asm__ ("foo %0, 1" : : "Nd" (val)); // expected-error {{invalid input constraint 'Nd' in asm}}
 }

clang/test/PCH/asm.h

 // Header for the PCH test asm.c
 
 void f() {
   int i,cond;
 
   asm ("foo\n" : : "a" (i + 2));
   asm ("foo\n" : [symbolic_name] "=a" (i) : "[symbolic_name]" (i));
-  asm volatile goto("testl %0, %0; jne %l1;" :: "r"(cond)::label_true, loop);
+  asm goto("testl %0, %0; jne %l1;" :: "r"(cond)::label_true, loop);
 label_true:
 loop:
   return;
 }
 
 void clobbers() {
   asm ("nop" : : : "ax", "#ax", "%ax");
   asm ("nop" : : : "eax", "rax", "ah", "al");
   asm ("nop" : : : "0", "%0", "#0");
 }

clang/test/Parser/asm-goto.c

 // RUN: %clang_cc1 -triple i386-pc-linux-gnu -fsyntax-only -verify %s
 // RUN: %clang_cc1 -triple x86_64-pc-linux-gnu -fsyntax-only -verify %s
 
 #if !__has_extension(gnu_asm)
 #error Extension 'gnu_asm' should be available by default
 #endif
 #if !__has_extension(gnu_asm_goto_with_outputs)
 #error Extension 'gnu_asm_goto_with_outputs' should be available by default
 #endif
 
 int a, b, c, d, e, f, g, h, i, j, k, l;
 
 void test(void) {
-  __asm__ volatile goto (""
+  __asm__ goto (""
             :: [a] "r" (a), [b] "r" (b), [c] "r" (c), [d] "r" (d),
                [e] "r" (e), [f] "r" (f), [g] "r" (g), [h] "r" (h),
                [i] "r" (i), [j] "r" (j), [k] "r" (k), [l] "r" (l)
             ::lab1,lab2);
 lab1: return;
 lab2: return;
 }
 
 void test2(void) {
-  __asm__ volatile goto (""
+  __asm__ goto (""
             :: [a] "r,m" (a), [b] "r,m" (b), [c] "r,m" (c), [d] "r,m" (d),
                [e] "r,m" (e), [f] "r,m" (f), [g] "r,m" (g), [h] "r,m" (h),
                [i] "r,m" (i), [j] "r,m" (j), [k] "r,m" (k), [l] "r,m" (l)
             :: lab);
   lab: return;
 }
 
 int test3(int x) {
-  __asm__ volatile goto ("decl %0; jnz %l[a]"
+  __asm__ goto ("decl %0; jnz %l[a]"
                          : "=r" (x) : "m" (x) : "memory" : a);
 a:
   return -x;
 }
 
 int test4(int x) {
   int y;
   if (x > 42)
-    __asm__ volatile goto ("decl %0; jnz %l[a]"
+    __asm__ goto ("decl %0; jnz %l[a]"
                            : "=r" (x), "=r" (y) : "m" (x) : "memory" : a);
   else
-    __asm__ volatile goto ("decl %0; jnz %l[b]"
+    __asm__ goto ("decl %0; jnz %l[b]"
                            : "=r" (x), "=r" (y) : "m" (x) : "memory" : b);
   x = y + 42;
 a:
   return -x;
 b:
   return +x;
 }
 

clang/test/Parser/asm-goto.cpp

 // RUN: %clang_cc1 -triple i386-pc-linux-gnu -fsyntax-only -verify -std=c++11 %s
 // RUN: %clang_cc1 -triple x86_64-pc-linux-gnu -fsyntax-only -verify -std=c++11 %s
 
 int a, b, c, d, e, f, g, h, i, j, k, l;
 
 void test1(void) {
-  __asm__ volatile goto (""
+  __asm__ goto (""
             :: [a] "r" (a), [b] "r" (b), [c] "r" (c), [d] "r" (d),
                [e] "r" (e), [f] "r" (f), [g] "r" (g), [h] "r" (h),
                [i] "r" (i), [j] "r" (j), [k] "r" (k), [l] "r" (l)
             ::lab1,lab2);
 lab1: return;
 lab2: return;
 }
 
 void test2(void) {
-  __asm__ volatile goto (""
+  __asm__ goto (""
             :: [a] "r,m" (a), [b] "r,m" (b), [c] "r,m" (c), [d] "r,m" (d),
                [e] "r,m" (e), [f] "r,m" (f), [g] "r,m" (g), [h] "r,m" (h),
                [i] "r,m" (i), [j] "r,m" (j), [k] "r,m" (k), [l] "r,m" (l)
             :: lab);
   lab: return;
 }
 
 int test3(int x) {
-  __asm__ volatile goto ("decl %0; jnz %l[a]"
+  __asm__ goto ("decl %0; jnz %l[a]"
                          : "=r" (x) : "m" (x) : "memory" : a);
 a:
   return -x;
 }
 
 int test4(int x) {
   int y;
   if (x > 42)
-    __asm__ volatile goto ("decl %0; jnz %l[a]"
+    __asm__ goto ("decl %0; jnz %l[a]"
                            : "=r" (x), "=r" (y) : "m" (x) : "memory" : a);
   else
-    __asm__ volatile goto ("decl %0; jnz %l[b]"
+    __asm__ goto ("decl %0; jnz %l[b]"
                            : "=r" (x), "=r" (y) : "m" (x) : "memory" : b);
   x = y + 42;
 a:
   return -x;
 b:
   return +x;
 }
 

clang/test/Parser/asm-qualifiers.c

 // RUN: %clang_cc1 -triple x86_64-linux-gnu -fsyntax-only -verify %s
 
 void qualifiers(void) {
   asm("");
-  asm volatile("");
+  asm volatile(""); // expected-warning {{volatile qualifier implied by lack of outputs}}
   asm inline("");
   asm goto("" ::::foo);
 foo:;
 }
 
 void unknown_qualifiers(void) {
   asm noodle(""); // expected-error {{expected 'volatile', 'inline', 'goto', or '('}}
   asm goto noodle("" ::::foo); // expected-error {{expected 'volatile', 'inline', 'goto', or '('}}
   asm volatile noodle inline(""); // expected-error {{expected 'volatile', 'inline', 'goto', or '('}}
 foo:;
 }
 
 void underscores(void) {
   __asm__("");
-  __asm__ __volatile__("");
+  __asm__ __volatile__(""); // expected-warning {{volatile qualifier implied by lack of outputs}}
   __asm__ __inline__("");
   // Note: goto is not supported with underscore prefix+suffix.
   __asm__ goto("" ::::foo);
 foo:;
 }
 
 void permutations(void) {
-  asm goto inline volatile("" ::::foo);
+  asm goto inline volatile("" ::::foo); // expected-warning {{volatile qualifier implied by goto}}
   asm goto inline("");
-  asm goto volatile inline("" ::::foo);
-  asm goto volatile("");
-  asm inline goto volatile("" ::::foo);
+  asm goto volatile inline("" ::::foo); // expected-warning {{volatile qualifier implied by goto}}
+  asm goto volatile("");                // expected-warning {{volatile qualifier implied by lack of outputs}}
+  asm inline goto volatile("" ::::foo); // expected-warning {{volatile qualifier implied by goto}}
   asm inline goto("" ::::foo);
-  asm inline volatile goto("" ::::foo);
-  asm inline volatile("");
-  asm volatile goto("" ::::foo);
-  asm volatile inline goto("" ::::foo);
-  asm volatile inline("");
+  asm inline volatile goto("" ::::foo); // expected-warning {{volatile qualifier implied by goto}}
+  asm inline volatile("");              // expected-warning {{volatile qualifier implied by lack of outputs}}
+  asm volatile goto("" ::::foo);        // expected-warning {{volatile qualifier implied by goto}}
+  asm volatile inline goto("" ::::foo); // expected-warning {{volatile qualifier implied by goto}}
+  asm volatile inline("");              // expected-warning {{volatile qualifier implied by lack of outputs}}
 foo:;
 }
 
 void duplicates(void) {
   asm volatile volatile("");             // expected-error {{duplicate asm qualifier 'volatile'}}
+                                         // expected-warning@-1 {{volatile qualifier implied by lack of outputs}}
   __asm__ __volatile__ __volatile__(""); // expected-error {{duplicate asm qualifier 'volatile'}}
+                                         // expected-warning@-1 {{volatile qualifier implied by lack of outputs}}
   asm inline inline("");                 // expected-error {{duplicate asm qualifier 'inline'}}
   __asm__ __inline__ __inline__("");     // expected-error {{duplicate asm qualifier 'inline'}}
   asm goto goto("" ::::foo);             // expected-error {{duplicate asm qualifier 'goto'}}
   __asm__ goto goto("" ::::foo);         // expected-error {{duplicate asm qualifier 'goto'}}
 foo:;
 }
 
 // globals
 asm ("");
 // <rdar://problem/7574870>
 asm volatile (""); // expected-error {{meaningless 'volatile' on asm outside function}}
 asm inline (""); // expected-error {{meaningless 'inline' on asm outside function}}
 asm goto (""::::noodle); // expected-error {{meaningless 'goto' on asm outside function}}
 // expected-error@-1 {{expected ')'}}
 // expected-note@-2 {{to match this '('}}
+
+// https://gcc.gnu.org/onlinedocs/gcc/Extended-Asm.html#Volatile
+// asm statements that have no output operands and asm goto statements, are
+// implicitly volatile.
+void implied_volatile (void) {
+  int y;
+  const int z = 0;
+
+  asm volatile ("");                   // expected-warning {{volatile qualifier implied by lack of outputs}}
+  asm volatile ("" : "=r"(y));
+  asm volatile ("" : "=r"(y) : "r"(z));
+  asm volatile ("" : "=r"(y) : "r"(z));
+  asm volatile ("" : "=r"(y) : "r"(z) : "cc");
+  asm volatile ("" : : "r"(z) : "cc"); // expected-warning {{volatile qualifier implied by lack of outputs}}
+  asm volatile ("" : : : "cc");        // expected-warning {{volatile qualifier implied by lack of outputs}}
+  asm volatile ("" : "=r"(y) : : "cc");
+  //
+  asm ("");
+  asm ("" : "=r"(y));
+  asm ("" : "=r"(y) : "r"(z));
+  asm ("" : "=r"(y) : "r"(z));
+  asm ("" : "=r"(y) : "r"(z) : "cc");
+  asm ("" : : "r"(z) : "cc");
+  asm ("" : : : "cc");
+  asm ("" : "=r"(y) : : "cc");
+}

clang/test/Parser/cxx-stmt.cpp

void f4() {
   int i = 0;
   switch (Kind) {
     case Type: i = 7; break;  // no error.
   }
 }
 
 // PR5500
 void f5() {
-  asm volatile ("":: :"memory");
-  asm volatile ("": ::"memory");
+  asm ("":: :"memory");
+  asm ("": ::"memory");
 }
 
 int f6() {
   int k, // expected-note {{change this ',' to a ';' to call 'f6'}}
   f6(), // expected-error {{expected ';'}} expected-warning {{interpreted as a function declaration}} expected-note {{replace paren}}
   int n = 0, // expected-error {{expected ';'}}
   return f5(), // ok
   int(n);
 }

clang/test/Sema/arm-asm.c

 // RUN: %clang_cc1 %s -triple armv7-apple-darwin -verify -fsyntax-only
 
 void f (void) {
   int Val;
   asm volatile ("lw (r1), %0[val]": "=&b"(Val)); // expected-error {{invalid output constraint '=&b' in asm}}
   return;
 }
 
 void test_64bit_r(void) {
   long long foo = 0, bar = 0;
   asm volatile("INST %0, %1" : "=r"(foo) : "r"(bar));
 }
 
 void test_clobber_conflict(void) {
   register int x asm("r1");
-  asm volatile("nop" :: "r"(x) : "%r1"); // expected-error {{conflicts with asm clobber list}}
-  asm volatile("nop" :: "l"(x) : "%r1"); // expected-error {{conflicts with asm clobber list}}
+  asm ("nop" :: "r"(x) : "%r1");          // expected-error {{conflicts with asm clobber list}}
+  asm ("nop" :: "l"(x) : "%r1");          // expected-error {{conflicts with asm clobber list}}
   asm volatile("nop" : "=r"(x) :: "%r1"); // expected-error {{conflicts with asm clobber list}}
 }

clang/test/Sema/arm64-inline-asm.c

 // RUN: %clang_cc1 -triple arm64-apple-ios7.1 -fsyntax-only -verify %s
 
 void foo() {
-  asm volatile("USE(%0)" :: "z"(0LL));
-  asm volatile("USE(%x0)" :: "z"(0LL));
-  asm volatile("USE(%w0)" :: "z"(0));
+  asm ("USE(%0)" :: "z"(0LL));
+  asm ("USE(%x0)" :: "z"(0LL));
+  asm ("USE(%w0)" :: "z"(0));
 
-  asm volatile("USE(%0)" :: "z"(0)); // expected-warning {{value size does not match register size specified by the constraint and modifier}} expected-note {{use constraint modifier "w"}}
+  asm ("USE(%0)" :: "z"(0)); // expected-warning {{value size does not match register size specified by the constraint and modifier}} expected-note {{use constraint modifier "w"}}
 }
 
 void test_clobber_conflict(void) {
   register long x asm("x1");
-  asm volatile("nop" :: "r"(x) : "%x1"); // expected-error {{conflicts with asm clobber list}}
+  asm ("nop" :: "r"(x) : "%x1");          // expected-error {{conflicts with asm clobber list}}
   asm volatile("nop" : "=r"(x) :: "%x1"); // expected-error {{conflicts with asm clobber list}}
 }

clang/test/Sema/asm-goto.cpp

 BAR:
   return;
 }
 
 int test3(int n)
 {
   // expected-error@+2 {{cannot jump from this asm goto statement to one of its possible targets}}
   // expected-error@+1 {{cannot jump from this asm goto statement to one of its possible targets}}
-  asm volatile goto("testl %0, %0; jne %l1;" :: "r"(n)::label_true, loop);
+  asm goto("testl %0, %0; jne %l1;" :: "r"(n)::label_true, loop);
   // expected-note@+2 {{jump bypasses initialization of variable length array}}
   // expected-note@+1 {{possible target of asm goto statement}}
   return ({int a[n];label_true: 2;});
   // expected-note@+1 {{jump bypasses initialization of variable length array}}
   int b[n];
 // expected-note@+1 {{possible target of asm goto statement}}
 loop:
   return 0;
 }

clang/test/Sema/asm-qualifiers.c

+// RUN: %clang_cc1 -Wno-asm-volatile -triple x86_64-linux-gnu -fsyntax-only -verify %s
+
+// Test that -Wno-asm-volatile disables -Wasm-volatile as expected.
+void qualifiers(void) {
+  asm volatile("");
+  __asm__ __volatile__("");
+  asm goto inline volatile("" ::::foo);
+  asm goto volatile inline("" ::::foo);
+  asm goto volatile("");
+  asm inline goto volatile("" ::::foo);
+  asm inline volatile goto("" ::::foo);
+  asm inline volatile("");
+  asm volatile goto("" ::::foo);
+  asm volatile inline goto("" ::::foo);
+  asm volatile inline("");
+foo:;
+  asm volatile volatile("");             // expected-error {{duplicate asm qualifier 'volatile'}}
+  __asm__ __volatile__ __volatile__(""); // expected-error {{duplicate asm qualifier 'volatile'}}
+  const int z = 0;
+  asm volatile ("" : : "r"(z) : "cc");
+  asm volatile ("" : : : "cc");
+}

clang/test/Sema/attr-naked.c

 // RUN: %clang_cc1 %s -verify -fsyntax-only -triple i686-pc-linux
 
 int a __attribute__((naked)); // expected-warning {{'naked' attribute only applies to functions}}
 
 __attribute__((naked)) int t0(void) {
-  __asm__ volatile("mov r0, #0");
+  __asm__ ("mov r0, #0");
 }
 
 void t1() __attribute__((naked));
 
 void t2() __attribute__((naked(2))); // expected-error {{'naked' attribute takes no arguments}}
 
 __attribute__((naked)) int t3() { // expected-note{{attribute is here}}
   return 42; // expected-error{{non-ASM statement in naked function is not supported}}

clang/test/Sema/inline-asm-validate-riscv.c

 // RUN: %clang_cc1 -triple riscv32 -fsyntax-only -verify %s
 // RUN: %clang_cc1 -triple riscv64 -fsyntax-only -verify %s
 
 void I(int i) {
   static const int BelowMin = -2049;
   static const int AboveMax = 2048;
-  asm volatile ("" :: "I"(BelowMin)); // expected-error{{value '-2049' out of range for constraint 'I'}}
-  asm volatile ("" :: "I"(AboveMax)); // expected-error{{value '2048' out of range for constraint 'I'}}
+  asm ("" :: "I"(BelowMin)); // expected-error{{value '-2049' out of range for constraint 'I'}}
+  asm ("" :: "I"(AboveMax)); // expected-error{{value '2048' out of range for constraint 'I'}}
 }
 
 void J(int j) {
   static const int BelowMin = -1;
   static const int AboveMax = 1;
-  asm volatile ("" :: "J"(BelowMin)); // expected-error{{value '-1' out of range for constraint 'J'}}
-  asm volatile ("" :: "J"(AboveMax)); // expected-error{{value '1' out of range for constraint 'J'}}
+  asm ("" :: "J"(BelowMin)); // expected-error{{value '-1' out of range for constraint 'J'}}
+  asm ("" :: "J"(AboveMax)); // expected-error{{value '1' out of range for constraint 'J'}}
 }
 
 void K(int k) {
   static const int BelowMin = -1;
   static const int AboveMax = 32;
-  asm volatile ("" :: "K"(BelowMin)); // expected-error{{value '-1' out of range for constraint 'K'}}
-  asm volatile ("" :: "K"(AboveMax)); // expected-error{{value '32' out of range for constraint 'K'}}
+  asm ("" :: "K"(BelowMin)); // expected-error{{value '-1' out of range for constraint 'K'}}
+  asm ("" :: "K"(AboveMax)); // expected-error{{value '32' out of range for constraint 'K'}}
 }
 
 void test_clobber_conflict(void) {
   register long x10 asm("x10");
-  asm volatile("" :: "r"(x10) : "x10"); // expected-error {{conflicts with asm clobber list}}
-  asm volatile("" :: "r"(x10) : "a0"); // expected-error {{conflicts with asm clobber list}}
+  asm ("" :: "r"(x10) : "x10"); // expected-error {{conflicts with asm clobber list}}
+  asm ("" :: "r"(x10) : "a0");  // expected-error {{conflicts with asm clobber list}}
   asm volatile("" : "=r"(x10) :: "x10"); // expected-error {{conflicts with asm clobber list}}
   asm volatile("" : "=r"(x10) :: "a0"); // expected-error {{conflicts with asm clobber list}}
 }

clang/test/Sema/inline-asm-validate-x86.c

 }
 
 void pr40890(void) {
   struct s {
     int a, b;
   };
   static struct s s;
   // This null pointer can be used as an integer constant expression.
-  __asm__ __volatile__("\n#define S_A abcd%0\n" : : "n"(&((struct s*)0)->a));
+  __asm__ ("\n#define S_A abcd%0\n" : : "n"(&((struct s*)0)->a));
   // This offset-from-null pointer can be used as an integer constant expression.
-  __asm__ __volatile__("\n#define S_B abcd%0\n" : : "n"(&((struct s*)0)->b));
+  __asm__ ("\n#define S_B abcd%0\n" : : "n"(&((struct s*)0)->b));
 #ifdef AMD64
   // This arbitrary pointer is fine.
-  __asm__ __volatile__("\n#define BEEF abcd%0\n" : : "n"((int*)0xdeadbeeeeeef));
+  __asm__ ("\n#define BEEF abcd%0\n" : : "n"((int*)0xdeadbeeeeeef));
 #endif
 }

clang/test/Sema/riscv-asm.c

 // RUN: %clang_cc1 %s -triple riscv32 -verify -fsyntax-only
 // RUN: %clang_cc1 %s -triple riscv64 -verify -fsyntax-only
 
 // expected-no-diagnostics
 
 void i (void) {
-  asm volatile ("" ::: "x0",  "x1",  "x2",  "x3",  "x4",  "x5",  "x6",  "x7");
-  asm volatile ("" ::: "x8",  "x9",  "x10", "x11", "x12", "x13", "x14", "x15");
-  asm volatile ("" ::: "x16", "x17", "x18", "x19", "x20", "x21", "x22", "x23");
-  asm volatile ("" ::: "x24", "x25", "x26", "x27", "x28", "x29", "x30", "x31");
+  asm ("" ::: "x0",  "x1",  "x2",  "x3",  "x4",  "x5",  "x6",  "x7");
+  asm ("" ::: "x8",  "x9",  "x10", "x11", "x12", "x13", "x14", "x15");
+  asm ("" ::: "x16", "x17", "x18", "x19", "x20", "x21", "x22", "x23");
+  asm ("" ::: "x24", "x25", "x26", "x27", "x28", "x29", "x30", "x31");
 
-  asm volatile ("" ::: "zero", "ra", "sp",  "gp",  "tp", "t0", "t1", "t2");
-  asm volatile ("" ::: "s0",   "s1", "a0",  "a1",  "a2", "a3", "a4", "a5");
-  asm volatile ("" ::: "a6",   "a7", "s2",  "s3",  "s4", "s5", "s6", "s7");
-  asm volatile ("" ::: "s8",   "s9", "s10", "s11", "t3", "t4", "t5", "t6");
+  asm ("" ::: "zero", "ra", "sp",  "gp",  "tp", "t0", "t1", "t2");
+  asm ("" ::: "s0",   "s1", "a0",  "a1",  "a2", "a3", "a4", "a5");
+  asm ("" ::: "a6",   "a7", "s2",  "s3",  "s4", "s5", "s6", "s7");
+  asm ("" ::: "s8",   "s9", "s10", "s11", "t3", "t4", "t5", "t6");
 }
 
 void f (void) {
-  asm volatile ("" ::: "f0",  "f1",  "f2",  "f3",  "f4",  "f5",  "f6",  "f7");
-  asm volatile ("" ::: "f8",  "f9",  "f10", "f11", "f12", "f13", "f14", "f15");
-  asm volatile ("" ::: "f16", "f17", "f18", "f19", "f20", "f21", "f22", "f23");
-  asm volatile ("" ::: "f24", "f25", "f26", "f27", "f28", "f29", "f30", "f31");
+  asm ("" ::: "f0",  "f1",  "f2",  "f3",  "f4",  "f5",  "f6",  "f7");
+  asm ("" ::: "f8",  "f9",  "f10", "f11", "f12", "f13", "f14", "f15");
+  asm ("" ::: "f16", "f17", "f18", "f19", "f20", "f21", "f22", "f23");
+  asm ("" ::: "f24", "f25", "f26", "f27", "f28", "f29", "f30", "f31");
 
-  asm volatile ("" ::: "ft0", "ft1", "ft2",  "ft3",  "ft4", "ft5", "ft6",  "ft7");
-  asm volatile ("" ::: "fs0", "fs1", "fa0",  "fa1",  "fa2", "fa3", "fa4",  "fa5");
-  asm volatile ("" ::: "fa6", "fa7", "fs2",  "fs3",  "fs4", "fs5", "fs6",  "fs7");
-  asm volatile ("" ::: "fs8", "fs9", "fs10", "fs11", "ft8", "ft9", "ft10", "ft11");
+  asm ("" ::: "ft0", "ft1", "ft2",  "ft3",  "ft4", "ft5", "ft6",  "ft7");
+  asm ("" ::: "fs0", "fs1", "fa0",  "fa1",  "fa2", "fa3", "fa4",  "fa5");
+  asm ("" ::: "fa6", "fa7", "fs2",  "fs3",  "fs4", "fs5", "fs6",  "fs7");
+  asm ("" ::: "fs8", "fs9", "fs10", "fs11", "ft8", "ft9", "ft10", "ft11");
 }

compiler-rt/lib/sanitizer_common/sanitizer_atomic_clang.h

 // - 64-bit atomic operations are not implemented on x86_32
 // (http://llvm.org/bugs/show_bug.cgi?id=15034)
 // - they are not implemented on ARM
 // error: undefined reference to '__atomic_load_4'
 
 // See http://www.cl.cam.ac.uk/~pes20/cpp/cpp0xmappings.html
 // for mappings of the memory model to different processors.
 
-inline void atomic_signal_fence(memory_order) {
-  __asm__ __volatile__("" ::: "memory");
-}
+inline void atomic_signal_fence(memory_order) { __asm__("" ::: "memory"); }
 
 inline void atomic_thread_fence(memory_order) {
   __sync_synchronize();
 }
 
 template<typename T>
 inline typename T::Type atomic_fetch_add(volatile T *a,
     typename T::Type v, memory_order mo) {

compiler-rt/lib/sanitizer_common/sanitizer_atomic_clang_x86.h

 //===----------------------------------------------------------------------===//
 
 #ifndef SANITIZER_ATOMIC_CLANG_X86_H
 #define SANITIZER_ATOMIC_CLANG_X86_H
 
 namespace __sanitizer {
 
 inline void proc_yield(int cnt) {
-  __asm__ __volatile__("" ::: "memory");
-  for (int i = 0; i < cnt; i++)
-    __asm__ __volatile__("pause");
-  __asm__ __volatile__("" ::: "memory");
+  __asm__("" ::: "memory");
+  for (int i = 0; i < cnt; i++) __asm__("pause");
+  __asm__("" ::: "memory");
 }
 
 template<typename T>
 inline typename T::Type atomic_load(
     const volatile T *a, memory_order mo) {
   DCHECK(mo & (memory_order_relaxed | memory_order_consume
       | memory_order_acquire | memory_order_seq_cst));
   DCHECK(!((uptr)a % sizeof(*a)));
   typename T::Type v;
 
   if (sizeof(*a) < 8 || sizeof(void*) == 8) {
     // Assume that aligned loads are atomic.
     if (mo == memory_order_relaxed) {
       v = a->val_dont_use;
     } else if (mo == memory_order_consume) {
       // Assume that processor respects data dependencies
       // (and that compiler won't break them).
-      __asm__ __volatile__("" ::: "memory");
+      __asm__("" ::: "memory");
       v = a->val_dont_use;
-      __asm__ __volatile__("" ::: "memory");
+      __asm__("" ::: "memory");
     } else if (mo == memory_order_acquire) {
-      __asm__ __volatile__("" ::: "memory");
+      __asm__("" ::: "memory");
       v = a->val_dont_use;
       // On x86 loads are implicitly acquire.
-      __asm__ __volatile__("" ::: "memory");
+      __asm__("" ::: "memory");
     } else {  // seq_cst
       // On x86 plain MOV is enough for seq_cst store.
-      __asm__ __volatile__("" ::: "memory");
+      __asm__("" ::: "memory");
       v = a->val_dont_use;
-      __asm__ __volatile__("" ::: "memory");
+      __asm__("" ::: "memory");
     }
   } else {
     // 64-bit load on 32-bit platform.
     __asm__ __volatile__(
         "movq %1, %%mm0;"  // Use mmx reg for 64-bit atomic moves
         "movq %%mm0, %0;"  // (ptr could be read-only)
         "emms;"            // Empty mmx state/Reset FP regs
         : "=m" (v)
inline void atomic_store(volatile T *a, typename T::Type v, memory_order mo) {
   DCHECK(!((uptr)a % sizeof(*a)));
 
   if (sizeof(*a) < 8 || sizeof(void*) == 8) {
     // Assume that aligned loads are atomic.
     if (mo == memory_order_relaxed) {
       a->val_dont_use = v;
     } else if (mo == memory_order_release) {
       // On x86 stores are implicitly release.
-      __asm__ __volatile__("" ::: "memory");
+      __asm__("" ::: "memory");
       a->val_dont_use = v;
-      __asm__ __volatile__("" ::: "memory");
+      __asm__("" ::: "memory");
     } else {  // seq_cst
       // On x86 stores are implicitly release.
-      __asm__ __volatile__("" ::: "memory");
+      __asm__("" ::: "memory");
       a->val_dont_use = v;
       __sync_synchronize();
     }
   } else {
     // 64-bit store on 32-bit platform.
     __asm__ __volatile__(
         "movq %1, %%mm0;"  // Use mmx reg for 64-bit atomic moves
         "movq %%mm0, %0;"

compiler-rt/lib/sanitizer_common/sanitizer_common.h

 // Make the compiler think that something is going on there.
 // Use this inside a loop that looks like memset/memcpy/etc to prevent the
 // compiler from recognising it and turning it into an actual call to
 // memset/memcpy/etc.
 static inline void SanitizerBreakOptimization(void *arg) {
 #if defined(_MSC_VER) && !defined(__clang__)
   _ReadWriteBarrier();
 #else
-  __asm__ __volatile__("" : : "r" (arg) : "memory");
+  __asm__("" : : "r"(arg) : "memory");
 #endif
 }
 
 struct SignalContext {
   void *siginfo;
   void *context;
   uptr addr;
   uptr pc;

compiler-rt/lib/sanitizer_common/tests/sanitizer_test_utils.h

 #endif  // ATTRIBUTE_NO_SANITIZE_ADDRESS
 
 #if __LP64__ || defined(_WIN64)
 #  define SANITIZER_WORDSIZE 64
 #else
 #  define SANITIZER_WORDSIZE 32
 #endif
 
-// Make the compiler thinks that something is going on there.
+// Make the compiler think that something is going on there.
 inline void break_optimization(void *arg) {
 #if !defined(_WIN32) || defined(__clang__)
-  __asm__ __volatile__("" : : "r" (arg) : "memory");
+  __asm__("" : : "r"(arg) : "memory");
 #endif
 }
 
 // This function returns its parameter but in such a way that compiler
 // can not prove it.
 template<class T>
 NOINLINE
 static T Ident(T t) {

compiler-rt/lib/scudo/standalone/common.h

for (u32 I = N - 1; I > 0; I--)
     Swap(A[I], A[getRandomModN(&State, I + 1)]);
   *RandState = State;
 }
 
 // Hardware specific inlinable functions.
 
 inline void yieldProcessor(u8 Count) {
 #if defined(__i386__) || defined(__x86_64__)
-  __asm__ __volatile__("" ::: "memory");
+  __asm__("" ::: "memory");
   for (u8 I = 0; I < Count; I++)
-    __asm__ __volatile__("pause");
+    __asm__("pause");
 #elif defined(__aarch64__) || defined(__arm__)
-  __asm__ __volatile__("" ::: "memory");
+  __asm__("" ::: "memory");
   for (u8 I = 0; I < Count; I++)
-    __asm__ __volatile__("yield");
+    __asm__("yield");
 #endif
-  __asm__ __volatile__("" ::: "memory");
+  __asm__("" ::: "memory");
 }
 
 // Platform specific functions.
 
 extern uptr PageSizeCached;
 uptr getPageSizeSlow();
 inline uptr getPageSizeCached() {
   // Bionic uses a hardcoded value.

compiler-rt/lib/scudo/standalone/tests/wrappers_cpp_test.cpp

TEST(ScudoWrappersCppTest, AllocAfterFork) {
   std::vector<std::thread *> Threads;
   for (size_t N = 0; N < 5; N++) {
     std::thread *T = new std::thread([&Stop] {
       while (!Stop) {
         for (size_t SizeLog = 3; SizeLog <= 20; SizeLog++) {
           char *P = new char[1UL << SizeLog];
           EXPECT_NE(P, nullptr);
           // Make sure this value is not optimized away.
-          asm volatile("" : : "r,m"(P) : "memory");
+          asm("" : : "r,m"(P) : "memory");
           delete[] P;
         }
       }
     });
     Threads.push_back(T);
   }
 
   // Create a thread to fork and allocate.
   for (size_t N = 0; N < 50; N++) {
     pid_t Pid;
     if ((Pid = fork()) == 0) {
       for (size_t SizeLog = 3; SizeLog <= 20; SizeLog++) {
         char *P = new char[1UL << SizeLog];
         EXPECT_NE(P, nullptr);
         // Make sure this value is not optimized away.
-        asm volatile("" : : "r,m"(P) : "memory");
+        asm("" : : "r,m"(P) : "memory");
         // Make sure we can touch all of the allocation.
         memset(P, 0x32, 1U << SizeLog);
         // EXPECT_LE(1U << SizeLog, malloc_usable_size(ptr));
         delete[] P;
       }
       _exit(10);
     }
     EXPECT_NE(-1, Pid);