diff --git a/clang/include/clang/Basic/Attr.td b/clang/include/clang/Basic/Attr.td
--- a/clang/include/clang/Basic/Attr.td
+++ b/clang/include/clang/Basic/Attr.td
@@ -300,10 +300,13 @@
 }
 
 class GNU<string name> : Spelling<name, "GNU">;
-class Declspec<string name> : Spelling<name, "Declspec">;
+class Declspec<string name> : Spelling<name, "Declspec"> {
+  bit MustPrintOnLeftSide = 1;
+}
 class Microsoft<string name> : Spelling<name, "Microsoft">;
 class CXX11<string namespace, string name, int version = 1>
     : Spelling<name, "CXX11", version> {
+  bit CanPrintOnLeftSide = 0;
   string Namespace = namespace;
 }
 class C2x<string namespace, string name, int version = 1>
@@ -559,6 +562,10 @@
 def SubjectMatcherForNamed : AttrSubjectMatcherAggregateRule<Named>;
 
 class Attr {
+  // On a declaration, can this attribute be print on the left side?
+  bit CanPrintOnLeftSide = 1;
+  // On a declaration, must this attribute be print on the left side?
+  bit MustPrintOnLeftSide = 0;
   // The various ways in which an attribute can be spelled in source
   list<Spelling> Spellings;
   // The things to which an attribute can appertain
@@ -892,6 +899,7 @@
 }
 
 def AsmLabel : InheritableAttr {
+  let CanPrintOnLeftSide = 0;
   let Spellings = [CustomKeyword<"asm">, CustomKeyword<"__asm__">];
   let Args = [
     // Label specifies the mangled name for the decl.
@@ -1434,6 +1442,7 @@
 }
 
 def EnableIf : InheritableAttr {
+  let CanPrintOnLeftSide = 0;
   // Does not have a [[]] spelling because this attribute requires the ability
   // to parse function arguments but the attribute is not written in the type
   // position.
@@ -2886,6 +2895,7 @@
 }
 
 def DiagnoseIf : InheritableAttr {
+  let CanPrintOnLeftSide = 0;
   // Does not have a [[]] spelling because this attribute requires the ability
   // to parse function arguments but the attribute is not written in the type
   // position.
diff --git a/clang/include/clang/Basic/CMakeLists.txt b/clang/include/clang/Basic/CMakeLists.txt
--- a/clang/include/clang/Basic/CMakeLists.txt
+++ b/clang/include/clang/Basic/CMakeLists.txt
@@ -30,6 +30,16 @@
   SOURCE Attr.td
   TARGET ClangAttrList)
 
+clang_tablegen(AttrLeftSideCanPrintList.inc -gen-clang-attr-can-print-left-list
+  -I ${CMAKE_CURRENT_SOURCE_DIR}/../../
+  SOURCE Attr.td
+  TARGET ClangAttrCanPrintLeftList)
+
+clang_tablegen(AttrLeftSideMustPrintList.inc -gen-clang-attr-must-print-left-list
+  -I ${CMAKE_CURRENT_SOURCE_DIR}/../../
+  SOURCE Attr.td
+  TARGET ClangAttrMustPrintLeftList)
+
 clang_tablegen(AttrSubMatchRulesList.inc -gen-clang-attr-subject-match-rule-list
   -I ${CMAKE_CURRENT_SOURCE_DIR}/../../
   SOURCE Attr.td
diff --git a/clang/lib/AST/DeclPrinter.cpp b/clang/lib/AST/DeclPrinter.cpp
--- a/clang/lib/AST/DeclPrinter.cpp
+++ b/clang/lib/AST/DeclPrinter.cpp
@@ -49,6 +49,18 @@
 
     void PrintObjCTypeParams(ObjCTypeParamList *Params);
 
+    enum class AttrPrintLoc {
+      None = 0,
+      Left = 1,
+      Right = 2,
+      Any = Left | Right,
+
+      LLVM_MARK_AS_BITMASK_ENUM(/*DefaultValue=*/Any)
+    };
+
+    void prettyPrintAttributes(Decl *D, raw_ostream &out,
+                               AttrPrintLoc loc = AttrPrintLoc::Any);
+
   public:
     DeclPrinter(raw_ostream &Out, const PrintingPolicy &Policy,
                 const ASTContext &Context, unsigned Indentation = 0,
@@ -117,7 +129,11 @@
                                 const TemplateParameterList *Params);
     void printTemplateArguments(llvm::ArrayRef<TemplateArgumentLoc> Args,
                                 const TemplateParameterList *Params);
-    void prettyPrintAttributes(Decl *D);
+
+    inline void prettyPrintAttributes(Decl *D) {
+      prettyPrintAttributes(D, Out);
+    }
+
     void prettyPrintPragmas(Decl *D);
     void printDeclType(QualType T, StringRef DeclName, bool Pack = false);
   };
@@ -234,7 +250,42 @@
   return Out;
 }
 
-void DeclPrinter::prettyPrintAttributes(Decl *D) {
+static bool canPrintOnLeftSide(attr::Kind kind) {
+  switch (kind) {
+#include "clang/Basic/AttrLeftSideCanPrintList.inc"
+    return true;
+  default:
+    return false;
+  }
+}
+
+static bool canPrintOnLeftSide(const Attr *A) {
+  if (A->isStandardAttributeSyntax()) {
+    return false;
+  }
+
+  return canPrintOnLeftSide(A->getKind());
+}
+
+static bool mustPrintOnLeftSide(attr::Kind kind) {
+  switch (kind) {
+#include "clang/Basic/AttrLeftSideMustPrintList.inc"
+    return true;
+  default:
+    return false;
+  }
+}
+
+static bool mustPrintOnLeftSide(const Attr *A) {
+  if (A->isDeclspecAttribute()) {
+    return true;
+  }
+
+  return mustPrintOnLeftSide(A->getKind());
+}
+
+void DeclPrinter::prettyPrintAttributes(Decl *D, llvm::raw_ostream &Out,
+                                        AttrPrintLoc Loc) {
   if (Policy.PolishForDeclaration)
     return;
 
@@ -243,14 +294,33 @@
     for (auto *A : Attrs) {
       if (A->isInherited() || A->isImplicit())
         continue;
-      switch (A->getKind()) {
-#define ATTR(X)
-#define PRAGMA_SPELLING_ATTR(X) case attr::X:
-#include "clang/Basic/AttrList.inc"
-        break;
-      default:
+
+      AttrPrintLoc attrloc = AttrPrintLoc::Right;
+      if (mustPrintOnLeftSide(A)) {
+        // If we must always print on left side (e.g. declspec), then mark as
+        // so.
+        attrloc = AttrPrintLoc::Left;
+      } else if (canPrintOnLeftSide(A)) {
+        // For functions with body defined we print the attributes on the left
+        // side so that GCC accept our dumps as well.
+        if (const FunctionDecl *FD = dyn_cast<FunctionDecl>(D);
+            FD && FD->isThisDeclarationADefinition()) {
+          // In case Decl is a function with a body, then attrs should be print
+          // on the left side.
+          attrloc = AttrPrintLoc::Left;
+
+          // In case it is a variable declaration with a ctor, then allow
+          // printing on the left side for readbility.
+        } else if (const VarDecl *VD = dyn_cast<VarDecl>(D);
+                   VD && VD->getInit() &&
+                   VD->getInitStyle() == VarDecl::CallInit) {
+          attrloc = AttrPrintLoc::Left;
+        }
+      }
+
+      // Only print the side matches the user requested.
+      if ((Loc & attrloc) != AttrPrintLoc::None) {
         A->printPretty(Out, Policy);
-        break;
       }
     }
   }
@@ -611,6 +681,22 @@
       printTemplateParameters(D->getTemplateParameterList(I));
   }
 
+  std::string LeftsideAttrs;
+  llvm::raw_string_ostream LSAS(LeftsideAttrs);
+
+  prettyPrintAttributes(D, LSAS, AttrPrintLoc::Left);
+
+  // prettyPrintAttributes print a space on left side of the attribute.
+  if (LeftsideAttrs[0] == ' ') {
+    // Skip the space prettyPrintAttributes generated.
+    LeftsideAttrs.erase(0, LeftsideAttrs.find_first_not_of(' '));
+
+    // Add a single space between the attribute and the Decl name.
+    LSAS << ' ';
+  }
+
+  Out << LeftsideAttrs;
+
   CXXConstructorDecl *CDecl = dyn_cast<CXXConstructorDecl>(D);
   CXXConversionDecl *ConversionDecl = dyn_cast<CXXConversionDecl>(D);
   CXXDeductionGuideDecl *GuideDecl = dyn_cast<CXXDeductionGuideDecl>(D);
@@ -772,7 +858,7 @@
     Ty.print(Out, Policy, Proto);
   }
 
-  prettyPrintAttributes(D);
+  prettyPrintAttributes(D, Out, AttrPrintLoc::Right);
 
   if (D->isPure())
     Out << " = 0";
@@ -865,6 +951,23 @@
 void DeclPrinter::VisitVarDecl(VarDecl *D) {
   prettyPrintPragmas(D);
 
+  std::string LeftSide;
+  llvm::raw_string_ostream LeftSideStream(LeftSide);
+
+  // Print attributes that should be placed on the left, such as __declspec.
+  prettyPrintAttributes(D, LeftSideStream, AttrPrintLoc::Left);
+
+  // prettyPrintAttributes print a space on left side of the attribute.
+  if (LeftSide[0] == ' ') {
+    // Skip the space prettyPrintAttributes generated.
+    LeftSide.erase(0, LeftSide.find_first_not_of(' '));
+
+    // Add a single space between the attribute and the Decl name.
+    LeftSideStream << ' ';
+  }
+
+  Out << LeftSide;
+
   QualType T = D->getTypeSourceInfo()
     ? D->getTypeSourceInfo()->getType()
     : D->getASTContext().getUnqualifiedObjCPointerType(D->getType());
@@ -897,10 +1000,19 @@
     }
   }
 
-  printDeclType(T, (isa<ParmVarDecl>(D) && Policy.CleanUglifiedParameters &&
-                    D->getIdentifier())
-                       ? D->getIdentifier()->deuglifiedName()
-                       : D->getName());
+  std::string Name;
+
+  Name += (isa<ParmVarDecl>(D) && Policy.CleanUglifiedParameters &&
+           D->getIdentifier())
+              ? D->getIdentifier()->deuglifiedName().str()
+              : D->getName().str();
+
+  printDeclType(T, StringRef(Name));
+
+  // Print the attributes that should be placed right before the end of the
+  // decl.
+  prettyPrintAttributes(D, Out, AttrPrintLoc::Right);
+
   Expr *Init = D->getInit();
   if (!Policy.SuppressInitializers && Init) {
     bool ImplicitInit = false;
@@ -929,7 +1041,6 @@
         Out << ")";
     }
   }
-  prettyPrintAttributes(D);
 }
 
 void DeclPrinter::VisitParmVarDecl(ParmVarDecl *D) {
diff --git a/clang/test/AST/ast-print-attr-knr.c b/clang/test/AST/ast-print-attr-knr.c
new file mode 100644
--- /dev/null
+++ b/clang/test/AST/ast-print-attr-knr.c
@@ -0,0 +1,17 @@
+// This file contain tests for attribute arguments on K&R functions.
+
+// RUN: %clang_cc1 -ast-print -x c -std=c89 -fms-extensions %s -o - | FileCheck %s
+
+// CHECK: int knr(i)
+// CHECK-NEXT: int i __attribute__((unused));
+// CHECK-NEXT: {
+// CHECK-NEXT: return 0;
+// CHECK-NEXT: }
+int knr(i) int i __attribute__((unused)); { return 0; }
+
+// CHECK: __attribute__((unused)) int knr2(i)
+// CHECK-NEXT: int i;
+// CHECK-NEXT: {
+// CHECK-NEXT: return 0;
+// CHECK-NEXT: }
+__attribute__((unused)) int knr2(i) int i; { return 0; }
diff --git a/clang/test/AST/ast-print-attr.c b/clang/test/AST/ast-print-attr.c
--- a/clang/test/AST/ast-print-attr.c
+++ b/clang/test/AST/ast-print-attr.c
@@ -32,3 +32,9 @@
 
 // CHECK: void fun_holds(int *a) __attribute__((ownership_holds(fun_holds, 1)));
 void fun_holds(int *a) __attribute__((ownership_holds(fun_holds, 1)));
+
+// CHECK: int fun_var_unused() {
+// CHECK-NEXT: int x __attribute__((unused)) = 0;
+// CHECK-NEXT: return x;
+// CHECK-NEXT: }
+int fun_var_unused() { int x __attribute__((unused)) = 0; return x; }
diff --git a/clang/test/AST/ast-print-pragmas.cpp b/clang/test/AST/ast-print-pragmas.cpp
--- a/clang/test/AST/ast-print-pragmas.cpp
+++ b/clang/test/AST/ast-print-pragmas.cpp
@@ -93,7 +93,7 @@
 #ifdef MS_EXT
 #pragma init_seg(compiler)
 // MS-EXT: #pragma init_seg (.CRT$XCC){{$}}
-// MS-EXT-NEXT: int x = 3 __declspec(thread);
-int __declspec(thread) x = 3;
+// MS-EXT-NEXT: __declspec(thread) int x = 3;
+__declspec(thread) int x = 3;
 #endif //MS_EXT
 
diff --git a/clang/test/Analysis/blocks.mm b/clang/test/Analysis/blocks.mm
--- a/clang/test/Analysis/blocks.mm
+++ b/clang/test/Analysis/blocks.mm
@@ -78,7 +78,7 @@
 // CHECK-NEXT:   1: 5
 // WARNINGS-NEXT:   2: [B1.1] (CXXConstructExpr, StructWithCopyConstructor)
 // ANALYZER-NEXT:   2: [B1.1] (CXXConstructExpr, [B1.3], StructWithCopyConstructor)
-// CHECK-NEXT:   3: StructWithCopyConstructor s(5) __attribute__((blocks("byref")));
+// CHECK-NEXT:   3: __attribute__((blocks("byref"))) StructWithCopyConstructor s(5);
 // CHECK-NEXT:   4: ^{ }
 // CHECK-NEXT:   5: (void)([B1.4]) (CStyleCastExpr, ToVoid, void)
 // CHECK-NEXT:   Preds (1): B2
diff --git a/clang/test/OpenMP/assumes_codegen.cpp b/clang/test/OpenMP/assumes_codegen.cpp
--- a/clang/test/OpenMP/assumes_codegen.cpp
+++ b/clang/test/OpenMP/assumes_codegen.cpp
@@ -67,41 +67,41 @@
 }
 #pragma omp end assumes
 
-// AST:      void foo() __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) {
+// AST:      __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) void foo() {
 // AST-NEXT: }
 // AST-NEXT: class BAR {
 // AST-NEXT: public:
-// AST-NEXT:     BAR() __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))     {
+// AST-NEXT:     __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) BAR()      {
 // AST-NEXT:     }
-// AST-NEXT:     void bar1() __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))     {
+// AST-NEXT:     __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) void bar1()     {
 // AST-NEXT:     }
-// AST-NEXT:     static void bar2() __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))     {
+// AST-NEXT:     __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) static void bar2()      {
 // AST-NEXT:     }
 // AST-NEXT: };
-// AST-NEXT: void bar() __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) {
+// AST-NEXT:  __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) void bar() {
 // AST-NEXT:     BAR b;
 // AST-NEXT: }
 // AST-NEXT: void baz() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")));
 // AST-NEXT: template <typename T> class BAZ {
 // AST-NEXT: public:
-// AST-NEXT:     BAZ<T>() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))     {
+// AST-NEXT:     __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) BAZ<T>()      {
 // AST-NEXT:     }
-// AST-NEXT:     void baz1() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))     {
+// AST-NEXT:     __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))  void baz1()     {
 // AST-NEXT:     }
-// AST-NEXT:     static void baz2() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))     {
+// AST-NEXT:     __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) static void baz2()      {
 // AST-NEXT:     }
 // AST-NEXT: };
 // AST-NEXT: template<> class BAZ<float> {
 // AST-NEXT: public:
-// AST-NEXT:     BAZ() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))    {
+// AST-NEXT:     __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))  BAZ()    {
 // AST-NEXT:     }
 // AST-NEXT:     void baz1() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")));
 // AST-NEXT:     static void baz2() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")));
 // AST-NEXT: };
-// AST-NEXT: void baz() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) {
+// AST-NEXT: __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) void baz() {
 // AST-NEXT:     BAZ<float> b;
 // AST-NEXT: }
-// AST-NEXT: int lambda_outer() __attribute__((assume("ompx_lambda_assumption"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) {
+// AST-NEXT: __attribute__((assume("ompx_lambda_assumption"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) int lambda_outer() {
 // AST-NEXT:     auto lambda_inner = []() {
 // AST-NEXT:         return 42;
 // AST-NEXT:     };
diff --git a/clang/test/OpenMP/assumes_print.cpp b/clang/test/OpenMP/assumes_print.cpp
--- a/clang/test/OpenMP/assumes_print.cpp
+++ b/clang/test/OpenMP/assumes_print.cpp
@@ -37,8 +37,8 @@
 }
 #pragma omp end assumes
 
-// CHECK: void foo() __attribute__((assume("omp_no_openmp_routines"))) __attribute__((assume("omp_no_openmp")))
-// CHECK: void bar() __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines"))) __attribute__((assume("omp_no_openmp")))
-// CHECK: void baz() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines"))) __attribute__((assume("omp_no_openmp")))
+// CHECK: __attribute__((assume("omp_no_openmp_routines"))) __attribute__((assume("omp_no_openmp"))) void foo()
+// CHECK: __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines"))) __attribute__((assume("omp_no_openmp"))) void bar()
+// CHECK: __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines"))) __attribute__((assume("omp_no_openmp"))) void baz()
 
 #endif
diff --git a/clang/test/OpenMP/assumes_template_print.cpp b/clang/test/OpenMP/assumes_template_print.cpp
--- a/clang/test/OpenMP/assumes_template_print.cpp
+++ b/clang/test/OpenMP/assumes_template_print.cpp
@@ -17,7 +17,7 @@
 struct S {
   int a;
 // CHECK: template <typename T> struct S {
-// CHECK:     void foo() __attribute__((assume("ompx_global_assumption")))     {
+// CHECK:     __attribute__((assume("ompx_global_assumption"))) void foo()     {
   void foo() {
     #pragma omp parallel
     {}
@@ -25,15 +25,15 @@
 };
 
 // CHECK: template<> struct S<int> {
-// CHECK:     void foo() __attribute__((assume("ompx_global_assumption")))     {
+// CHECK:     __attribute__((assume("ompx_global_assumption"))) void foo()     {
 
 #pragma omp begin assumes no_openmp
-// CHECK: void S_with_assumes_no_call() __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) {
+// CHECK: __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) void S_with_assumes_no_call() {
 void S_with_assumes_no_call() {
   S<int> s;
   s.a = 0;
 }
-// CHECK: void S_with_assumes_call() __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) {
+// CHECK: __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) void S_with_assumes_call() {
 void S_with_assumes_call() {
   S<int> s;
   s.a = 0;
@@ -42,7 +42,7 @@
 }
 #pragma omp end assumes
 
-// CHECK: void S_without_assumes() __attribute__((assume("ompx_global_assumption"))) {
+// CHECK: __attribute__((assume("ompx_global_assumption"))) void S_without_assumes() {
 void S_without_assumes() {
   S<int> s;
   s.foo();
@@ -54,7 +54,7 @@
 template <typename T>
 struct P {
 // CHECK: template <typename T> struct P {
-// CHECK:     void foo() __attribute__((assume("ompx_global_assumption")))     {
+// CHECK:    __attribute__((assume("ompx_global_assumption"))) void foo()      {
   int a;
   void foo() {
     #pragma omp parallel
@@ -65,21 +65,21 @@
 // TODO: Avoid the duplication here:
 
 // CHECK: template<> struct P<int> {
-// CHECK:     void foo() __attribute__((assume("ompx_global_assumption"))) __attribute__((assume("ompx_global_assumption")))     {
+// CHECK:      __attribute__((assume("ompx_global_assumption"))) __attribute__((assume("ompx_global_assumption"))) void foo()   {
 
-// CHECK: void P_without_assumes() __attribute__((assume("ompx_global_assumption"))) {
+// CHECK: __attribute__((assume("ompx_global_assumption"))) void P_without_assumes() {
 void P_without_assumes() {
   P<int> p;
   p.foo();
 }
 
 #pragma omp begin assumes no_openmp
-// CHECK: void P_with_assumes_no_call() __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) {
+// CHECK: __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) void P_with_assumes_no_call() {
 void P_with_assumes_no_call() {
   P<int> p;
   p.a = 0;
 }
-// CHECK: void P_with_assumes_call() __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) {
+// CHECK: __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) void P_with_assumes_call() {
 void P_with_assumes_call() {
   P<int> p;
   p.a = 0;
diff --git a/clang/test/OpenMP/declare_simd_ast_print.cpp b/clang/test/OpenMP/declare_simd_ast_print.cpp
--- a/clang/test/OpenMP/declare_simd_ast_print.cpp
+++ b/clang/test/OpenMP/declare_simd_ast_print.cpp
@@ -60,11 +60,11 @@
 
 class VV {
   // CHECK: #pragma omp declare simd uniform(this, a) linear(val(b): a)
-  // CHECK-NEXT: int add(int a, int b) __attribute__((cold))    {
+  // CHECK-NEXT:  __attribute__((cold)) int add(int a, int b)     {
   // CHECK-NEXT: return a + b;
   // CHECK-NEXT: }
   #pragma omp declare simd uniform(this, a) linear(val(b): a)
-  int add(int a, int b) __attribute__((cold)) { return a + b; }
+  __attribute__((cold)) int add(int a, int b) { return a + b; }
 
   // CHECK: #pragma omp declare simd aligned(b: 4) aligned(a) linear(ref(b): 4) linear(val(this)) linear(val(a))
   // CHECK-NEXT: float taddpf(float *a, float *&b)     {
diff --git a/clang/test/Sema/attr-print.c b/clang/test/Sema/attr-print.c
--- a/clang/test/Sema/attr-print.c
+++ b/clang/test/Sema/attr-print.c
@@ -3,8 +3,7 @@
 // CHECK: int x __attribute__((aligned(4)));
 int x __attribute__((aligned(4)));
 
-// FIXME: Print this at a valid location for a __declspec attr.
-// CHECK: int y __declspec(align(4));
+// CHECK: __declspec(align(4)) int y;
 __declspec(align(4)) int y;
 
 // CHECK: short arr[3] __attribute__((aligned));
diff --git a/clang/test/SemaCXX/attr-print.cpp b/clang/test/SemaCXX/attr-print.cpp
--- a/clang/test/SemaCXX/attr-print.cpp
+++ b/clang/test/SemaCXX/attr-print.cpp
@@ -3,8 +3,7 @@
 // CHECK: int x __attribute__((aligned(4)));
 int x __attribute__((aligned(4)));
 
-// FIXME: Print this at a valid location for a __declspec attr.
-// CHECK: int y __declspec(align(4));
+// CHECK: __declspec(align(4)) int y;
 __declspec(align(4)) int y;
 
 // CHECK: void foo() __attribute__((const));
diff --git a/clang/test/SemaCXX/cxx11-attr-print.cpp b/clang/test/SemaCXX/cxx11-attr-print.cpp
--- a/clang/test/SemaCXX/cxx11-attr-print.cpp
+++ b/clang/test/SemaCXX/cxx11-attr-print.cpp
@@ -3,8 +3,7 @@
 // CHECK: int x __attribute__((aligned(4)));
 int x __attribute__((aligned(4)));
 
-// FIXME: Print this at a valid location for a __declspec attr.
-// CHECK: int y __declspec(align(4));
+// CHECK: __declspec(align(4)) int y;
 __declspec(align(4)) int y;
 
 // CHECK: int z {{\[}}[gnu::aligned(4)]];
@@ -66,7 +65,7 @@
 // CHECK: int m __attribute__((aligned(4
 // CHECK: int n alignas(4
 // CHECK: int p alignas(int
-// CHECK: static int f() __attribute__((pure))
+// CHECK: __attribute__((pure)) static int f()
 // CHECK: static int g() {{\[}}[gnu::pure]]
 template <typename T> struct S {
   __attribute__((aligned(4))) int m;
@@ -82,7 +81,7 @@
 
 // CHECK: int m __attribute__((aligned(4
 // CHECK: int n alignas(4
-// CHECK: static int f() __attribute__((pure))
+// CHECK: __attribute__((pure)) static int f()
 // CHECK: static int g() {{\[}}[gnu::pure]]
 template struct S<int>;
 
diff --git a/clang/utils/TableGen/ClangAttrEmitter.cpp b/clang/utils/TableGen/ClangAttrEmitter.cpp
--- a/clang/utils/TableGen/ClangAttrEmitter.cpp
+++ b/clang/utils/TableGen/ClangAttrEmitter.cpp
@@ -3163,6 +3163,27 @@
   OS << "#undef PRAGMA_SPELLING_ATTR\n";
 }
 
+// Emits the enumeration list for attributes.
+void EmitClangAttrPrintList(const std::string &FieldName, RecordKeeper &Records,
+                            raw_ostream &OS) {
+  emitSourceFileHeader(
+      "List of attributes that can be print on the left side of a decl", OS);
+
+  AttrClassHierarchy Hierarchy(Records);
+
+  std::vector<Record *> Attrs = Records.getAllDerivedDefinitions("Attr");
+  std::vector<Record *> PragmaAttrs;
+  for (auto *Attr : Attrs) {
+    if (!Attr->getValueAsBit("ASTNode"))
+      continue;
+
+    if (!Attr->getValueAsBit(FieldName))
+      continue;
+
+    OS << "case attr::" << Attr->getName() << ":\n";
+  }
+}
+
 // Emits the enumeration list for attributes.
 void EmitClangAttrSubjectMatchRuleList(RecordKeeper &Records, raw_ostream &OS) {
   emitSourceFileHeader(
diff --git a/clang/utils/TableGen/TableGen.cpp b/clang/utils/TableGen/TableGen.cpp
--- a/clang/utils/TableGen/TableGen.cpp
+++ b/clang/utils/TableGen/TableGen.cpp
@@ -31,6 +31,8 @@
   GenClangAttrSubjectMatchRulesParserStringSwitches,
   GenClangAttrImpl,
   GenClangAttrList,
+  GenClangAttrCanPrintLeftList,
+  GenClangAttrMustPrintLeftList,
   GenClangAttrDocTable,
   GenClangAttrSubjectMatchRuleList,
   GenClangAttrPCHRead,
@@ -127,6 +129,14 @@
                    "Generate clang attribute implementations"),
         clEnumValN(GenClangAttrList, "gen-clang-attr-list",
                    "Generate a clang attribute list"),
+        clEnumValN(GenClangAttrCanPrintLeftList,
+                   "gen-clang-attr-can-print-left-list",
+                   "Generate list of attributes that can be printed on left "
+                   "side of a decl"),
+        clEnumValN(GenClangAttrMustPrintLeftList,
+                   "gen-clang-attr-must-print-left-list",
+                   "Generate list of attributes that must be printed on left "
+                   "side of a decl"),
         clEnumValN(GenClangAttrDocTable, "gen-clang-attr-doc-table",
                    "Generate a table of attribute documentation"),
         clEnumValN(GenClangAttrSubjectMatchRuleList,
@@ -269,11 +279,14 @@
                    "Generate riscv_vector_builtin_cg.inc for clang"),
         clEnumValN(GenRISCVVectorBuiltinSema, "gen-riscv-vector-builtin-sema",
                    "Generate riscv_vector_builtin_sema.inc for clang"),
-        clEnumValN(GenRISCVSiFiveVectorBuiltins, "gen-riscv-sifive-vector-builtins",
+        clEnumValN(GenRISCVSiFiveVectorBuiltins,
+                   "gen-riscv-sifive-vector-builtins",
                    "Generate riscv_sifive_vector_builtins.inc for clang"),
-        clEnumValN(GenRISCVSiFiveVectorBuiltinCG, "gen-riscv-sifive-vector-builtin-codegen",
+        clEnumValN(GenRISCVSiFiveVectorBuiltinCG,
+                   "gen-riscv-sifive-vector-builtin-codegen",
                    "Generate riscv_sifive_vector_builtin_cg.inc for clang"),
-        clEnumValN(GenRISCVSiFiveVectorBuiltinSema, "gen-riscv-sifive-vector-builtin-sema",
+        clEnumValN(GenRISCVSiFiveVectorBuiltinSema,
+                   "gen-riscv-sifive-vector-builtin-sema",
                    "Generate riscv_sifive_vector_builtin_sema.inc for clang"),
         clEnumValN(GenAttrDocs, "gen-attr-docs",
                    "Generate attribute documentation"),
@@ -315,6 +328,12 @@
   case GenClangAttrList:
     EmitClangAttrList(Records, OS);
     break;
+  case GenClangAttrCanPrintLeftList:
+    EmitClangAttrPrintList("CanPrintOnLeftSide", Records, OS);
+    break;
+  case GenClangAttrMustPrintLeftList:
+    EmitClangAttrPrintList("MustPrintOnLeftSide", Records, OS);
+    break;
   case GenClangAttrDocTable:
     EmitClangAttrDocTable(Records, OS);
     break;
diff --git a/clang/utils/TableGen/TableGenBackends.h b/clang/utils/TableGen/TableGenBackends.h
--- a/clang/utils/TableGen/TableGenBackends.h
+++ b/clang/utils/TableGen/TableGenBackends.h
@@ -39,6 +39,8 @@
 void EmitClangAttrClass(llvm::RecordKeeper &Records, llvm::raw_ostream &OS);
 void EmitClangAttrImpl(llvm::RecordKeeper &Records, llvm::raw_ostream &OS);
 void EmitClangAttrList(llvm::RecordKeeper &Records, llvm::raw_ostream &OS);
+void EmitClangAttrPrintList(const std::string &FieldName,
+                            llvm::RecordKeeper &Records, llvm::raw_ostream &OS);
 void EmitClangAttrSubjectMatchRuleList(llvm::RecordKeeper &Records,
                                        llvm::raw_ostream &OS);
 void EmitClangAttrPCHRead(llvm::RecordKeeper &Records, llvm::raw_ostream &OS);