From patchwork Tue Oct 18 18:09:54 2011
Content-Type: text/plain; charset="utf-8"
MIME-Version: 1.0
Content-Transfer-Encoding: 7bit
X-Patchwork-Submitter: Torvald Riegel <triegel@redhat.com>
X-Patchwork-Id: 120476
Return-Path: 
 <gcc-patches-return-304879-incoming=patchwork.ozlabs.org@gcc.gnu.org>
X-Original-To: incoming@patchwork.ozlabs.org
Delivered-To: patchwork-incoming@bilbo.ozlabs.org
Received: from sourceware.org (server1.sourceware.org [209.132.180.131])
	by ozlabs.org (Postfix) with SMTP id 243A3B71BB
	for <incoming@patchwork.ozlabs.org>;
	Wed, 19 Oct 2011 05:10:45 +1100 (EST)
Received: (qmail 1001 invoked by alias); 18 Oct 2011 18:10:42 -0000
Received: (qmail 961 invoked by uid 22791); 18 Oct 2011 18:10:30 -0000
X-SWARE-Spam-Status: No, hits=-5.9 required=5.0	tests=AWL, BAYES_40,
	RCVD_IN_DNSWL_HI, RP_MATCHES_RCVD, SPF_HELO_PASS, TW_RX, TW_TM,
	TW_TR, TW_TX
X-Spam-Check-By: sourceware.org
Received: from mx1.redhat.com (HELO mx1.redhat.com) (209.132.183.28) by
	sourceware.org (qpsmtpd/0.43rc1) with ESMTP;
	Tue, 18 Oct 2011 18:09:59 +0000
Received: from int-mx02.intmail.prod.int.phx2.redhat.com
	(int-mx02.intmail.prod.int.phx2.redhat.com [10.5.11.12])	by
	mx1.redhat.com (8.14.4/8.14.4) with ESMTP id
	p9II9xfO024136	(version=TLSv1/SSLv3 cipher=DHE-RSA-AES256-SHA
	bits=256 verify=OK)	for <gcc-patches@gcc.gnu.org>;
	Tue, 18 Oct 2011 14:09:59 -0400
Received: from [10.36.6.64] (vpn1-6-64.ams2.redhat.com [10.36.6.64])	by
	int-mx02.intmail.prod.int.phx2.redhat.com (8.13.8/8.13.8)
	with ESMTP id p9II9tqe009451; Tue, 18 Oct 2011 14:09:56 -0400
Subject: [trans-mem] Rename/split __transaction into __transaction_atomic
	and __transaction_relaxed.
From: Torvald Riegel <triegel@redhat.com>
To: GCC Patches <gcc-patches@gcc.gnu.org>
Cc: Richard Henderson <rth@redhat.com>, Aldy Hernandez <aldyh@redhat.com>
Date: Tue, 18 Oct 2011 20:09:54 +0200
Message-ID: <1318961394.5756.429.camel@triegel.csb>
Mime-Version: 1.0
X-IsSubscribed: yes
Mailing-List: contact gcc-patches-help@gcc.gnu.org; run by ezmlm
Precedence: bulk
List-Id: <gcc-patches.gcc.gnu.org>
List-Unsubscribe: 
 <mailto:gcc-patches-unsubscribe-incoming=patchwork.ozlabs.org@gcc.gnu.org>
List-Archive: <http://gcc.gnu.org/ml/gcc-patches/>
List-Post: <mailto:gcc-patches@gcc.gnu.org>
List-Help: <mailto:gcc-patches-help@gcc.gnu.org>
Sender: gcc-patches-owner@gcc.gnu.org
Delivered-To: mailing list gcc-patches@gcc.gnu.org

The attached patch renames/splits __transaction (with optional
[[atomic]] and [[relaxed]] attributes) into two keywords that explicitly
select an atomic or relaxed transaction, with no default. This is a
recent change in the C++ TM spec.

patch10: Contains all the work on the GCC side. I do keep the single
internal transaction construct together with the internal flags for
relaxed and atomic outer transactions, but the parser is changed to only
accept the new syntax.

patch11: Adjusts the syntax used in the libitm test cases accordingly.

Several ICEs in the TM tests on C++, but I think they are old. C TM
tests work except some missing optimizations (old failures too).
c-c++-common/tm/wrap-[12].c fail on C++, but not on C:
wrap-1.c:5:57: error: 'transaction_wrap' argument not an identifier
I believe this is an old error too?
(I don't yet have a testcase summary of before the changes, so I can't
compare properly. But will send an follow-up if there are regressions.)

OK for branch?
commit bc14ebc1d062d9e6cca6ef571aa47071ffe741dc
Author: Torvald Riegel <triegel@redhat.com>
Date:   Tue Oct 18 02:01:05 2011 +0200

    Rename/split __transaction into __transaction_atomic and __transaction_relaxed.
    
    	* c-common.h (RID_TRANSACTION): Split into RID_TRANSACTION_ATOMIC
    	and RID_TRANSACTION_RELAXED.
    	* c-common.c (RID_TRANSACTION): Same.
    	(  parse_tm_stmt_attr): Do not accept "atomic" and "relaxed"
    	attributes anymore.
    	* c-family/common.h: Same.
    	* c-family/common.c: Same.
    	* c-parser.c (c_parser_transaction_expression, c_parser_transaction,
    	c_parser_statement_after_labels, c_parser_unary_expression):
    	Handle both RID_TRANSACTION_ATOMIC and RID_TRANSACTION_RELAXED.
    	(c_parser_attribute_any_word): Only RID_TRANSACTION_ATOMIC accepts
    	attributes.
    	(c_parser_transaction_cancel): Update error messages.
    	* tree-pretty-print.c: Adjust to new syntax.
    	* gimple-pretty-print.c: Same.
    	* cp/parser.c (cp_parser_transaction, cp_parser_function_transaction,
    	cp_parser_transaction_expression, cp_parser_unary_expression,
    	cp_parser_statement, cp_parser_function_definition_after_declarator,
    	cp_parser_token_starts_function_definition_p):
    	Handle both RID_TRANSACTION_ATOMIC and RID_TRANSACTION_RELAXED.
    	(RT_TRANSACTION): Split into RT_TRANSACTION_ATOMIC and
    	RT_TRANSACTION_RELAXED.
    	(cp_parser_required_error): Same.
    	(cp_parser_transaction_cancel): Update error messages.
    	* cp/parser.h (cp_parser): Update comment.
    	* gimple.h: Same.
    	* gimple.def: Same.
    	* cp/semantics.c (begin_transaction_stmt): Add flags parameter.
    	* cp/cp-tree.h: Same.
    	* cp/pt.c (tsubst_expr): Same.
    
    	* testsuite/c-c++-common/tm/*: Adjust for new syntax.
    	* testsuite/g++.dg/tm/*: Same.
    	* testsuite/gcc.dg/tm/*: Same.
commit 3e61c60240993744ba921e32b32fb91dda253d9d
Author: Torvald Riegel <triegel@redhat.com>
Date:   Tue Oct 18 01:58:44 2011 +0200

    Change __transaction to __transaction_atomic or __transaction_relaxed.
    
    	* testsuite/*: Change __transaction to __transaction_atomic or
    	__transaction_relaxed.

diff --git a/libitm/testsuite/libitm.c++/dropref.C b/libitm/testsuite/libitm.c++/dropref.C
index ee4f1bb..c9bdfff 100644
--- a/libitm/testsuite/libitm.c++/dropref.C
+++ b/libitm/testsuite/libitm.c++/dropref.C
@@ -5,7 +5,7 @@ char *pp;
 
 int main()
 {
-  __transaction {
+  __transaction_atomic {
     _ITM_dropReferences (pp, 555);
   }
   return 0;
diff --git a/libitm/testsuite/libitm.c++/eh-1.C b/libitm/testsuite/libitm.c++/eh-1.C
index d572dd3..49e2213 100644
--- a/libitm/testsuite/libitm.c++/eh-1.C
+++ b/libitm/testsuite/libitm.c++/eh-1.C
@@ -14,7 +14,7 @@ static void f1()
 
 static void f2()
 {
-  __transaction {
+  __transaction_atomic {
     f1();
   }
 }
diff --git a/libitm/testsuite/libitm.c++/static_ctor.C b/libitm/testsuite/libitm.c++/static_ctor.C
index 1dcd5d5..e25b56b 100644
--- a/libitm/testsuite/libitm.c++/static_ctor.C
+++ b/libitm/testsuite/libitm.c++/static_ctor.C
@@ -14,7 +14,7 @@ int f(int x)
 static void *thread (void *)
 {
   int bar;
-  __transaction { bar = f(10); }
+  __transaction_atomic { bar = f(10); }
   if (bar != 100)
     abort();
   return 0;
@@ -30,7 +30,7 @@ int main()
   pthread_join(pt, NULL);
 
   // Now y should already be initialized.
-  __transaction { bar = f(20); }
+  __transaction_atomic { bar = f(20); }
   if (bar != 200)
     abort();
 
diff --git a/libitm/testsuite/libitm.c++/throwdown.C b/libitm/testsuite/libitm.c++/throwdown.C
index 188fdbe..a53cf92 100644
--- a/libitm/testsuite/libitm.c++/throwdown.C
+++ b/libitm/testsuite/libitm.c++/throwdown.C
@@ -7,7 +7,7 @@ static void throwit() {
 }
 
 void tranfunc() {
-	__transaction {
+	__transaction_atomic {
 		throwit();
 	}
 }
diff --git a/libitm/testsuite/libitm.c/dropref-2.c b/libitm/testsuite/libitm.c/dropref-2.c
index 2386b18..5bb2726 100644
--- a/libitm/testsuite/libitm.c/dropref-2.c
+++ b/libitm/testsuite/libitm.c/dropref-2.c
@@ -13,7 +13,7 @@ int main()
   for(i=0; i < 100; ++i)
     pp[i]=0x22;
 
-  __transaction {
+  __transaction_atomic {
     for(i=0; i < 100; ++i)
       pp[i]=0x33;
 
diff --git a/libitm/testsuite/libitm.c/dropref.c b/libitm/testsuite/libitm.c/dropref.c
index ee4f1bb..c9bdfff 100644
--- a/libitm/testsuite/libitm.c/dropref.c
+++ b/libitm/testsuite/libitm.c/dropref.c
@@ -5,7 +5,7 @@ char *pp;
 
 int main()
 {
-  __transaction {
+  __transaction_atomic {
     _ITM_dropReferences (pp, 555);
   }
   return 0;
diff --git a/libitm/testsuite/libitm.c/reentrant.c b/libitm/testsuite/libitm.c/reentrant.c
index aeb9a0e..d67e8ac 100644
--- a/libitm/testsuite/libitm.c/reentrant.c
+++ b/libitm/testsuite/libitm.c/reentrant.c
@@ -10,7 +10,7 @@ int x = 0;
 
 int __attribute__((transaction_pure)) pure(int i)
 {
-  __transaction {
+  __transaction_atomic {
     x++;
   }
   if (_ITM_inTransaction() == outsideTransaction)
@@ -22,7 +22,7 @@ int __attribute__((transaction_unsafe)) unsafe(int i)
 {
   if (_ITM_inTransaction() != inIrrevocableTransaction)
     abort();
-  __transaction {
+  __transaction_atomic {
     x++;
   }
   if (_ITM_inTransaction() != inIrrevocableTransaction)
@@ -32,10 +32,10 @@ int __attribute__((transaction_unsafe)) unsafe(int i)
 
 static void *thread (void *dummy __attribute__((unused)))
 {
-  __transaction {
+  __transaction_atomic {
     pure(1);
   }
-  __transaction[[relaxed]] {
+  __transaction_relaxed {
     unsafe(1);
   }
   return 0;
@@ -46,10 +46,10 @@ int main()
   pthread_t pt;
   int r = 0;
 
-  __transaction {
+  __transaction_atomic {
     r += pure(1) + x;
   }
-  __transaction[[relaxed]] {
+  __transaction_relaxed {
     r += unsafe(1) + x;
   }
   if (r != 7)
diff --git a/libitm/testsuite/libitm.c/simple-1.c b/libitm/testsuite/libitm.c/simple-1.c
index 0456232..de41740 100644
--- a/libitm/testsuite/libitm.c/simple-1.c
+++ b/libitm/testsuite/libitm.c/simple-1.c
@@ -9,7 +9,7 @@ static int x;
 
 static void start (void)
 {
-  __transaction { x++; }
+  __transaction_atomic { x++; }
 }
 
 int main()
diff --git a/libitm/testsuite/libitm.c/simple-2.c b/libitm/testsuite/libitm.c/simple-2.c
index 3402fc2..e92c917 100644
--- a/libitm/testsuite/libitm.c/simple-2.c
+++ b/libitm/testsuite/libitm.c/simple-2.c
@@ -7,7 +7,7 @@ static int x;
 
 static void *start (void *dummy __attribute__((unused)))
 {
-  __transaction { x++; }
+  __transaction_atomic { x++; }
   return NULL;
 }
 
diff --git a/libitm/testsuite/libitm.c/txrelease.c b/libitm/testsuite/libitm.c/txrelease.c
index 8d8b697..7f3b49b 100644
--- a/libitm/testsuite/libitm.c/txrelease.c
+++ b/libitm/testsuite/libitm.c/txrelease.c
@@ -18,7 +18,7 @@ thread_exit_handler(void *dummy __attribute__((unused)))
     {
       // ??? It would be good if we could check here that the transaction has
       // indeed been released.
-      __transaction { round++; }
+      __transaction_atomic { round++; }
       if (pthread_setspecific(key, &round))
         abort();
     }
@@ -32,7 +32,7 @@ static void *thread (void *dummy __attribute__((unused)))
     abort();
   if (pthread_setspecific(key, &round))
     abort();
-  __transaction { round++; }
+  __transaction_atomic { round++; }
   return NULL;
 }

--- a/gcc/ChangeLog.tm
+++ b/gcc/ChangeLog.tm
@@ -1,3 +1,40 @@
+2011-10-17  Torvald Riegel  <triegel@redhat.com>
+
+	* c-common.h (RID_TRANSACTION): Split into RID_TRANSACTION_ATOMIC
+	and RID_TRANSACTION_RELAXED.
+	* c-common.c (RID_TRANSACTION): Same.
+	(  parse_tm_stmt_attr): Do not accept "atomic" and "relaxed"
+	attributes anymore.
+	* c-family/common.h: Same.
+	* c-family/common.c: Same.
+	* c-parser.c (c_parser_transaction_expression, c_parser_transaction,
+	c_parser_statement_after_labels, c_parser_unary_expression):
+	Handle both RID_TRANSACTION_ATOMIC and RID_TRANSACTION_RELAXED.
+	(c_parser_attribute_any_word): Only RID_TRANSACTION_ATOMIC accepts
+	attributes.
+	(c_parser_transaction_cancel): Update error messages.
+	* tree-pretty-print.c: Adjust to new syntax.
+	* gimple-pretty-print.c: Same.
+	* cp/parser.c (cp_parser_transaction, cp_parser_function_transaction,
+	cp_parser_transaction_expression, cp_parser_unary_expression,
+	cp_parser_statement, cp_parser_function_definition_after_declarator,
+	cp_parser_token_starts_function_definition_p):
+	Handle both RID_TRANSACTION_ATOMIC and RID_TRANSACTION_RELAXED.
+	(RT_TRANSACTION): Split into RT_TRANSACTION_ATOMIC and
+	RT_TRANSACTION_RELAXED.
+	(cp_parser_required_error): Same.
+	(cp_parser_transaction_cancel): Update error messages.
+	* cp/parser.h (cp_parser): Update comment.
+	* gimple.h: Same.
+	* gimple.def: Same.
+	* cp/semantics.c (begin_transaction_stmt): Add flags parameter.
+	* cp/cp-tree.h: Same.
+	* cp/pt.c (tsubst_expr): Same.
+	
+	* testsuite/c-c++-common/tm/*: Adjust for new syntax.
+	* testsuite/g++.dg/tm/*: Same.
+	* testsuite/gcc.dg/tm/*: Same.
+
 2011-09-20  Richard Henderson  <rth@redhat.com>
 
 	* trans-mem.c (transaction_invariant_address_p): Handle MEM_REF.
diff --git a/gcc/c-common.c b/gcc/c-common.c
index 99061ea..198c6ae 100644
--- a/gcc/c-common.c
+++ b/gcc/c-common.c
@@ -618,7 +618,8 @@ const struct c_common_resword c_common_reswords[] =
   { "__signed",		RID_SIGNED,	0 },
   { "__signed__",	RID_SIGNED,	0 },
   { "__thread",		RID_THREAD,	0 },
-  { "__transaction",	RID_TRANSACTION, 0 },
+  { "__transaction_atomic", RID_TRANSACTION_ATOMIC, 0 },
+  { "__transaction_relaxed", RID_TRANSACTION_RELAXED, 0 },
   { "__transaction_cancel", RID_TRANSACTION_CANCEL, 0 },
   { "__typeof",		RID_TYPEOF,	0 },
   { "__typeof__",	RID_TYPEOF,	0 },
@@ -7222,10 +7223,6 @@ parse_tm_stmt_attr (tree attrs, int allowed)
 
       if (is_attribute_p ("outer", a))
 	m = TM_STMT_ATTR_OUTER;
-      else if (is_attribute_p ("atomic", a))
-	m = TM_STMT_ATTR_ATOMIC;
-      else if (is_attribute_p ("relaxed", a))
-	m = TM_STMT_ATTR_RELAXED;
 
       if ((m & allowed) == 0)
 	{
diff --git a/gcc/c-common.h b/gcc/c-common.h
index b045a9b..47b5657 100644
--- a/gcc/c-common.h
+++ b/gcc/c-common.h
@@ -86,7 +86,7 @@ enum rid
   RID_CXX_COMPAT_WARN,
 
   /* GNU transactional memory extension */
-  RID_TRANSACTION, RID_TRANSACTION_CANCEL,
+  RID_TRANSACTION_ATOMIC, RID_TRANSACTION_RELAXED, RID_TRANSACTION_CANCEL,
 
   /* Too many ways of getting the name of a function as a string */
   RID_FUNCTION_NAME, RID_PRETTY_FUNCTION_NAME, RID_C99_FUNCTION_NAME,
diff --git a/gcc/c-family/c-common.c b/gcc/c-family/c-common.c
index 847a24f..af4bbb1 100644
--- a/gcc/c-family/c-common.c
+++ b/gcc/c-family/c-common.c
@@ -472,7 +472,8 @@ const struct c_common_resword c_common_reswords[] =
   { "__signed",		RID_SIGNED,	0 },
   { "__signed__",	RID_SIGNED,	0 },
   { "__thread",		RID_THREAD,	0 },
-  { "__transaction",   RID_TRANSACTION, 0 },
+  { "__transaction_atomic", RID_TRANSACTION_ATOMIC, 0 },
+  { "__transaction_relaxed", RID_TRANSACTION_RELAXED, 0 },
   { "__transaction_cancel", RID_TRANSACTION_CANCEL, 0 },
   { "__typeof",		RID_TYPEOF,	0 },
   { "__typeof__",	RID_TYPEOF,	0 },
@@ -7325,10 +7326,6 @@ parse_tm_stmt_attr (tree attrs, int allowed)
 
       if (is_attribute_p ("outer", a))
 	m = TM_STMT_ATTR_OUTER;
-      else if (is_attribute_p ("atomic", a))
-	m = TM_STMT_ATTR_ATOMIC;
-      else if (is_attribute_p ("relaxed", a))
-	m = TM_STMT_ATTR_RELAXED;
 
       if ((m & allowed) == 0)
 	{
diff --git a/gcc/c-family/c-common.h b/gcc/c-family/c-common.h
index 090e132..8d7a9d8 100644
--- a/gcc/c-family/c-common.h
+++ b/gcc/c-family/c-common.h
@@ -112,7 +112,7 @@ enum rid
   RID_CXX_COMPAT_WARN,
 
   /* GNU transactional memory extension */
-  RID_TRANSACTION, RID_TRANSACTION_CANCEL,
+  RID_TRANSACTION_ATOMIC, RID_TRANSACTION_RELAXED, RID_TRANSACTION_CANCEL,
 
   /* Too many ways of getting the name of a function as a string */
   RID_FUNCTION_NAME, RID_PRETTY_FUNCTION_NAME, RID_C99_FUNCTION_NAME,
diff --git a/gcc/c-parser.c b/gcc/c-parser.c
index 1b57b1b..5c8724e 100644
--- a/gcc/c-parser.c
+++ b/gcc/c-parser.c
@@ -1172,8 +1172,8 @@ static struct c_expr c_parser_postfix_expression_after_paren_type (c_parser *,
 static struct c_expr c_parser_postfix_expression_after_primary (c_parser *,
 								location_t loc,
 								struct c_expr);
-static tree c_parser_transaction (c_parser *);
-static struct c_expr c_parser_transaction_expression (c_parser *);
+static tree c_parser_transaction (c_parser *, enum rid);
+static struct c_expr c_parser_transaction_expression (c_parser *, enum rid);
 static tree c_parser_transaction_cancel (c_parser *);
 static struct c_expr c_parser_expression (c_parser *);
 static struct c_expr c_parser_expression_conv (c_parser *);
@@ -3402,7 +3402,7 @@ c_parser_attribute_any_word (c_parser *parser)
 	case RID_FRACT:
 	case RID_ACCUM:
 	case RID_SAT:
-	case RID_TRANSACTION:
+	case RID_TRANSACTION_ATOMIC:
 	case RID_TRANSACTION_CANCEL:
 	  ok = true;
 	  break;
@@ -4454,8 +4454,10 @@ c_parser_statement_after_labels (c_parser *parser)
 	case RID_ASM:
 	  stmt = c_parser_asm_statement (parser);
 	  break;
-	case RID_TRANSACTION:
-	  stmt = c_parser_transaction (parser);
+	case RID_TRANSACTION_ATOMIC:
+	case RID_TRANSACTION_RELAXED:
+	  stmt = c_parser_transaction (parser,
+	      c_parser_peek_token (parser)->keyword);
 	  break;
 	case RID_TRANSACTION_CANCEL:
 	  stmt = c_parser_transaction_cancel (parser);
@@ -5897,8 +5899,10 @@ c_parser_unary_expression (c_parser *parser)
 	  op = c_parser_cast_expression (parser, NULL);
 	  op = default_function_array_conversion (exp_loc, op);
 	  return parser_build_unary_op (op_loc, IMAGPART_EXPR, op);
-	case RID_TRANSACTION:
-	  return c_parser_transaction_expression (parser);
+	case RID_TRANSACTION_ATOMIC:
+	case RID_TRANSACTION_RELAXED:
+	  return c_parser_transaction_expression (parser,
+	      c_parser_peek_token (parser)->keyword);
 	default:
 	  return c_parser_postfix_expression (parser);
 	}
@@ -10447,33 +10451,36 @@ c_parser_transaction_attributes (c_parser *parser)
   return attr;
 }
 
-/* Parse a __transaction statement (GCC Extension).
+/* Parse a __transaction_atomic or __transaction_relaxed statement
+   (GCC Extension).
 
    transaction-statement:
-     __transaction attributes[opt] compound-statement
+     __transaction_atomic attributes[opt] compound-statement
+     __transaction_relaxed compound-statement
 
-   Note that the only valid attributes are: "atomic", "relaxed", "outer".
+   Note that the only valid attribute is: "outer".
 */
 
 static tree
-c_parser_transaction (c_parser *parser)
+c_parser_transaction (c_parser *parser, enum rid keyword)
 {
   unsigned int old_in = parser->in_transaction;
   unsigned int this_in = 1, new_in;
   location_t loc = c_parser_peek_token (parser)->location;
   tree stmt, attrs;
 
-  gcc_assert (c_parser_next_token_is_keyword (parser, RID_TRANSACTION));
+  gcc_assert ((keyword == RID_TRANSACTION_ATOMIC
+      || keyword == RID_TRANSACTION_RELAXED)
+      && c_parser_next_token_is_keyword (parser, keyword));
   c_parser_consume_token (parser);
 
-  attrs = c_parser_transaction_attributes (parser);
-  if (attrs)
+  if (keyword == RID_TRANSACTION_RELAXED)
+    this_in |= TM_STMT_ATTR_RELAXED;
+  else
     {
-      this_in |= parse_tm_stmt_attr (attrs, (TM_STMT_ATTR_OUTER
-					     | TM_STMT_ATTR_ATOMIC
-					     | TM_STMT_ATTR_RELAXED));
-      /* The [[ atomic ]] attribute is the same as no attribute.  */
-      this_in &= ~TM_STMT_ATTR_ATOMIC;
+      attrs = c_parser_transaction_attributes (parser);
+      if (attrs)
+        this_in |= parse_tm_stmt_attr (attrs, TM_STMT_ATTR_OUTER);
     }
 
   /* Keep track if we're in the lexical scope of an outer transaction.  */
@@ -10486,22 +10493,24 @@ c_parser_transaction (c_parser *parser)
   if (flag_tm)
     stmt = c_finish_transaction (loc, stmt, this_in);
   else
-    error_at (loc, "%<__transaction%> without "
-	      "transactional memory support enabled");
+    error_at (loc, (keyword == RID_TRANSACTION_ATOMIC ?
+        "%<__transaction_atomic%> without transactional memory support enabled"
+        : "%<__transaction_relaxed %> "
+        "without transactional memory support enabled"));
 
   return stmt;
 }
 
-/* Parse a __transaction expression (GCC Extension).
+/* Parse a __transaction_atomic or __transaction_relaxed expression
+   (GCC Extension).
 
    transaction-expression:
-     __transaction attributes[opt] ( expression )
-
-   Note that the only valid attributes are: "atomic" and "relaxed".
+     __transaction_atomic ( expression )
+     __transaction_relaxed ( expression )
 */
 
 static struct c_expr
-c_parser_transaction_expression (c_parser *parser)
+c_parser_transaction_expression (c_parser *parser, enum rid keyword)
 {
   struct c_expr ret;
   unsigned int old_in = parser->in_transaction;
@@ -10509,16 +10518,18 @@ c_parser_transaction_expression (c_parser *parser)
   location_t loc = c_parser_peek_token (parser)->location;
   tree attrs;
 
-  gcc_assert (c_parser_next_token_is_keyword (parser, RID_TRANSACTION));
+  gcc_assert ((keyword == RID_TRANSACTION_ATOMIC
+      || keyword == RID_TRANSACTION_RELAXED)
+      && c_parser_next_token_is_keyword (parser, keyword));
   c_parser_consume_token (parser);
 
-  attrs = c_parser_transaction_attributes (parser);
-  if (attrs)
+  if (keyword == RID_TRANSACTION_RELAXED)
+    this_in |= TM_STMT_ATTR_RELAXED;
+  else
     {
-      this_in |= parse_tm_stmt_attr (attrs, (TM_STMT_ATTR_ATOMIC
-					     | TM_STMT_ATTR_RELAXED));
-      /* The [[ atomic ]] attribute is the same as no attribute.  */
-      this_in &= ~TM_STMT_ATTR_ATOMIC;
+      attrs = c_parser_transaction_attributes (parser);
+      if (attrs)
+        this_in |= parse_tm_stmt_attr (attrs, 0);
     }
 
   parser->in_transaction = this_in;
@@ -10542,8 +10553,10 @@ c_parser_transaction_expression (c_parser *parser)
   parser->in_transaction = old_in;
 
   if (!flag_tm)
-    error_at (loc, "%<__transaction%> without "
-	      "transactional memory support enabled");
+    error_at (loc, (keyword == RID_TRANSACTION_ATOMIC ?
+        "%<__transaction_atomic%> without transactional memory support enabled"
+        : "%<__transaction_relaxed %> "
+        "without transactional memory support enabled"));
 
   return ret;
 }
@@ -10579,7 +10592,7 @@ c_parser_transaction_cancel(c_parser *parser)
   else if (parser->in_transaction & TM_STMT_ATTR_RELAXED)
     {
       error_at (loc, "%<__transaction_cancel%> within a "
-		"relaxed %<__transaction%>");
+		"%<__transaction_relaxed%>");
       goto ret_error;
     }
   else if (is_outer)
@@ -10588,14 +10601,15 @@ c_parser_transaction_cancel(c_parser *parser)
 	  && !is_tm_may_cancel_outer (current_function_decl))
 	{
 	  error_at (loc, "outer %<__transaction_cancel%> not "
-		    "within outer %<__transaction%>");
+		    "within outer %<__transaction_atomic%>");
 	  error_at (loc, "  or a %<transaction_may_cancel_outer%> function");
 	  goto ret_error;
 	}
     }
   else if (parser->in_transaction == 0)
     {
-      error_at (loc, "%<__transaction_cancel%> not within %<__transaction%>");
+      error_at (loc, "%<__transaction_cancel%> not within "
+                "%<__transaction_atomic%>");
       goto ret_error;
     }
 
diff --git a/gcc/cp/cp-tree.h b/gcc/cp/cp-tree.h
index a71018b..ccc6239 100644
--- a/gcc/cp/cp-tree.h
+++ b/gcc/cp/cp-tree.h
@@ -5458,7 +5458,7 @@ extern void finish_omp_atomic			(enum tree_code, enum tree_code,
 extern void finish_omp_barrier			(void);
 extern void finish_omp_flush			(void);
 extern void finish_omp_taskwait			(void);
-extern tree begin_transaction_stmt		(location_t, tree *);
+extern tree begin_transaction_stmt		(location_t, tree *, int);
 extern void finish_transaction_stmt		(tree, tree, int);
 extern void finish_omp_taskyield		(void);
 extern bool cxx_omp_create_clause_info		(tree, tree, bool, bool, bool);
diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c
index 0235a08..0c2a915 100644
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -173,7 +173,8 @@ typedef enum required_token {
   RT_CLASS_KEY, /* class-key */
   RT_CLASS_TYPENAME_TEMPLATE /* class, typename, or template */
 
-  , RT_TRANSACTION, /* __transaction */
+  , RT_TRANSACTION_ATOMIC, /* __transaction_atomic */
+  RT_TRANSACTION_RELAXED, /* __transaction_relaxed */
   RT_TRANSACTION_CANCEL /* __transaction_cancel */
 } required_token;
 
@@ -1867,11 +1868,11 @@ static void cp_parser_label_declaration
 /* Transactional Memory Extensions */
 
 static tree cp_parser_transaction
-  (cp_parser *);
+  (cp_parser *, enum rid);
 static tree cp_parser_transaction_expression
-  (cp_parser *);
+  (cp_parser *, enum rid);
 static bool cp_parser_function_transaction
-  (cp_parser *);
+  (cp_parser *, enum rid);
 static tree cp_parser_transaction_cancel
   (cp_parser *);
 
@@ -5843,8 +5844,9 @@ cp_parser_unary_expression (cp_parser *parser, bool address_p, bool cast_p,
 	  }
 	  break;
 
-	case RID_TRANSACTION:
-	  return cp_parser_transaction_expression (parser);
+	case RID_TRANSACTION_ATOMIC:
+	case RID_TRANSACTION_RELAXED:
+	  return cp_parser_transaction_expression (parser, keyword);
 
 	case RID_NOEXCEPT:
 	  {
@@ -8026,8 +8028,9 @@ cp_parser_statement (cp_parser* parser, tree in_statement_expr,
 	  cp_parser_declaration_statement (parser);
 	  return;
 	  
-	case RID_TRANSACTION:
-	  statement = cp_parser_transaction (parser);
+	case RID_TRANSACTION_ATOMIC:
+	case RID_TRANSACTION_RELAXED:
+	  statement = cp_parser_transaction (parser, keyword);
 	  break;
 	case RID_TRANSACTION_CANCEL:
 	  statement = cp_parser_transaction_cancel (parser);
@@ -20112,11 +20115,17 @@ cp_parser_function_definition_after_declarator (cp_parser* parser,
 
   start_lambda_scope (current_function_decl);
 
-  /* If the next token is `try' or `__transaction', then we are looking at
-     either function-try-block or function-transaction-block.  Note that
-     both of these include the function-body.  */
-  if (cp_lexer_next_token_is_keyword (parser->lexer, RID_TRANSACTION))
-    ctor_initializer_p = cp_parser_function_transaction (parser);
+  /* If the next token is `try', `__transaction_atomic', or
+     `__transaction_relaxed`, then we are looking at either function-try-block
+     or function-transaction-block.  Note that all of these include the
+     function-body.  */
+  if (cp_lexer_next_token_is_keyword (parser->lexer, RID_TRANSACTION_ATOMIC))
+    ctor_initializer_p = cp_parser_function_transaction (parser,
+        RID_TRANSACTION_ATOMIC);
+  else if (cp_lexer_next_token_is_keyword (parser->lexer,
+      RID_TRANSACTION_RELAXED))
+    ctor_initializer_p = cp_parser_function_transaction (parser,
+        RID_TRANSACTION_RELAXED);
   else if (cp_lexer_next_token_is_keyword (parser->lexer, RID_TRY))
     ctor_initializer_p = cp_parser_function_try_block (parser);
   else
@@ -21167,8 +21176,11 @@ cp_parser_required_error (cp_parser *parser,
       case RT_AT_THROW:
 	cp_parser_error (parser, "expected %<@throw%>");
 	return;
-      case RT_TRANSACTION:
-	cp_parser_error (parser, "expected %<__transaction%>");
+      case RT_TRANSACTION_ATOMIC:
+	cp_parser_error (parser, "expected %<__transaction_atomic%>");
+	return;
+      case RT_TRANSACTION_RELAXED:
+	cp_parser_error (parser, "expected %<__transaction_relaxed%>");
 	return;
       default:
 	break;
@@ -21400,8 +21412,10 @@ cp_parser_token_starts_function_definition_p (cp_token* token)
 	  || token->type == CPP_COLON
 	  /* A function-try-block begins with `try'.  */
 	  || token->keyword == RID_TRY
-	  /* A function-transaction-block begins with `__transaction'.  */
-	  || token->keyword == RID_TRANSACTION
+	  /* A function-transaction-block begins with `__transaction_atomic'
+	     or `__transaction_relaxed'.  */
+	  || token->keyword == RID_TRANSACTION_ATOMIC
+	  || token->keyword == RID_TRANSACTION_RELAXED
 	  /* The named return value extension begins with `return'.  */
 	  || token->keyword == RID_RETURN);
 }
@@ -25762,40 +25776,44 @@ cp_parser_txn_attribute_opt (cp_parser *parser)
   return attr;
 }
 
-/* Parse a __transaction statement.
+/* Parse a __transaction_atomic or __transaction_relaxed statement.
 
    transaction-statement:
-     __transaction txn-attribute[opt] txn-exception-spec[opt]
-	compound-statement
+     __transaction_atomic txn-attribute[opt] txn-exception-spec[opt]
+       compound-statement
+     __transaction_relaxed txn-exception-spec[opt] compound-statement
 
    ??? The exception specification is not yet implemented.
 */
 
 static tree
-cp_parser_transaction (cp_parser *parser)
+cp_parser_transaction (cp_parser *parser, enum rid keyword)
 {
   unsigned char old_in = parser->in_transaction;
   unsigned char this_in = 1, new_in;
   cp_token *token;
   tree stmt, attrs;
 
-  token = cp_parser_require_keyword (parser, RID_TRANSACTION, RT_TRANSACTION);
+  gcc_assert (keyword == RID_TRANSACTION_ATOMIC
+      || keyword == RID_TRANSACTION_RELAXED);
+  token = cp_parser_require_keyword (parser, keyword,
+      (keyword == RID_TRANSACTION_ATOMIC ? RT_TRANSACTION_ATOMIC
+          : RT_TRANSACTION_RELAXED));
   gcc_assert (token != NULL);
 
-  attrs = cp_parser_txn_attribute_opt (parser);
-  if (attrs)
+  if (keyword == RID_TRANSACTION_RELAXED)
+    this_in |= TM_STMT_ATTR_RELAXED;
+  else
     {
-      this_in |= parse_tm_stmt_attr (attrs, (TM_STMT_ATTR_OUTER
-					     | TM_STMT_ATTR_ATOMIC
-					     | TM_STMT_ATTR_RELAXED));
-      /* The [[ atomic ]] attribute is the same as no attribute.  */
-      this_in &= ~TM_STMT_ATTR_ATOMIC;
+      attrs = cp_parser_txn_attribute_opt (parser);
+      if (attrs)
+        this_in |= parse_tm_stmt_attr (attrs, TM_STMT_ATTR_OUTER);
     }
 
   /* Keep track if we're in the lexical scope of an outer transaction.  */
   new_in = this_in | (old_in & TM_STMT_ATTR_OUTER);
 
-  stmt = begin_transaction_stmt (token->location, NULL);
+  stmt = begin_transaction_stmt (token->location, NULL, this_in);
 
   parser->in_transaction = new_in;
   cp_parser_compound_statement (parser, NULL, false, false);
@@ -25806,33 +25824,37 @@ cp_parser_transaction (cp_parser *parser)
   return stmt;
 }
 
-/* Parse a __transaction expression.
+/* Parse a __transaction_atomic or __transaction_relaxed expression.
 
    transaction-expression:
-     __transaction txn-attribute[opt] txn-exception-spec[opt]
-	compound-statement
+     __transaction_atomic txn-exception-spec[opt] compound-statement
+     __transaction_relaxed txn-exception-spec[opt] compound-statement
 
    ??? The exception specification is not yet implemented.
 */
 
 static tree
-cp_parser_transaction_expression (cp_parser *parser)
+cp_parser_transaction_expression (cp_parser *parser, enum rid keyword)
 {
   unsigned char old_in = parser->in_transaction;
   unsigned char this_in = 1;
   cp_token *token;
   tree ret, attrs;
 
-  token = cp_parser_require_keyword (parser, RID_TRANSACTION, RT_TRANSACTION);
+  gcc_assert (keyword == RID_TRANSACTION_ATOMIC
+      || keyword == RID_TRANSACTION_RELAXED);
+  token = cp_parser_require_keyword (parser, keyword,
+      (keyword == RID_TRANSACTION_ATOMIC ? RT_TRANSACTION_ATOMIC
+          : RT_TRANSACTION_RELAXED));
   gcc_assert (token != NULL);
 
-  attrs = cp_parser_txn_attribute_opt (parser);
-  if (attrs)
+  if (keyword == RID_TRANSACTION_RELAXED)
+    this_in |= TM_STMT_ATTR_RELAXED;
+  else
     {
-      this_in |= parse_tm_stmt_attr (attrs, (TM_STMT_ATTR_ATOMIC
-					     | TM_STMT_ATTR_RELAXED));
-      /* The [[ atomic ]] attribute is the same as no attribute.  */
-      this_in &= ~TM_STMT_ATTR_ATOMIC;
+      attrs = cp_parser_txn_attribute_opt (parser);
+      if (attrs)
+        this_in |= parse_tm_stmt_attr (attrs, 0);
     }
 
   parser->in_transaction = this_in;
@@ -25857,12 +25879,15 @@ cp_parser_transaction_expression (cp_parser *parser)
 /* Parse a function-transaction-block.
 
    function-transaction-block:
-     __transaction txn-attribute[opt] ctor-initializer[opt] function-body
-     __transaction txn-attribute[opt] function-try-block
+     __transaction_atomic txn-attribute[opt] ctor-initializer[opt]
+         function-body
+     __transaction_atomic txn-attribute[opt] function-try-block
+     __transaction_relaxed ctor-initializer[opt] function-body
+     __transaction_relaxed function-try-block
 */
 
 static bool
-cp_parser_function_transaction (cp_parser *parser)
+cp_parser_function_transaction (cp_parser *parser, enum rid keyword)
 {
   unsigned char old_in = parser->in_transaction;
   unsigned char new_in = 1;
@@ -25870,20 +25895,23 @@ cp_parser_function_transaction (cp_parser *parser)
   bool ctor_initializer_p;
   cp_token *token;
 
-  token = cp_parser_require_keyword (parser, RID_TRANSACTION, RT_TRANSACTION);
+  gcc_assert (keyword == RID_TRANSACTION_ATOMIC
+      || keyword == RID_TRANSACTION_RELAXED);
+  token = cp_parser_require_keyword (parser, keyword,
+      (keyword == RID_TRANSACTION_ATOMIC ? RT_TRANSACTION_ATOMIC
+          : RT_TRANSACTION_RELAXED));
   gcc_assert (token != NULL);
   
-  attrs = cp_parser_txn_attribute_opt (parser);
-  if (attrs)
+  if (keyword == RID_TRANSACTION_RELAXED)
+    new_in |= TM_STMT_ATTR_RELAXED;
+  else
     {
-      new_in |= parse_tm_stmt_attr (attrs, (TM_STMT_ATTR_OUTER
-					    | TM_STMT_ATTR_ATOMIC
-					    | TM_STMT_ATTR_RELAXED));
-      /* The [[ atomic ]] attribute is the same as no attribute.  */
-      new_in &= ~TM_STMT_ATTR_ATOMIC;
+      attrs = cp_parser_txn_attribute_opt (parser);
+      if (attrs)
+        new_in |= parse_tm_stmt_attr (attrs, TM_STMT_ATTR_OUTER);
     }
 
-  stmt = begin_transaction_stmt (token->location, &compound_stmt);
+  stmt = begin_transaction_stmt (token->location, &compound_stmt, new_in);
 
   parser->in_transaction = new_in;
 
@@ -25936,7 +25964,7 @@ cp_parser_transaction_cancel (cp_parser *parser)
   else if (parser->in_transaction & TM_STMT_ATTR_RELAXED)
     {
       error_at (token->location, "%<__transaction_cancel%> within a "
-		"relaxed %<__transaction%>");
+		"%<__transaction_relaxed%>");
       return error_mark_node;
     }
   else if (is_outer)
@@ -25945,7 +25973,7 @@ cp_parser_transaction_cancel (cp_parser *parser)
 	  && !is_tm_may_cancel_outer (current_function_decl))
 	{
 	  error_at (token->location, "outer %<__transaction_cancel%> not "
-		    "within outer %<__transaction%>");
+		    "within outer %<__transaction_atomic%>");
 	  error_at (token->location,
 		    "  or a %<transaction_may_cancel_outer%> function");
 	  return error_mark_node;
@@ -25954,7 +25982,7 @@ cp_parser_transaction_cancel (cp_parser *parser)
   else if (parser->in_transaction == 0)
     {
       error_at (token->location, "%<__transaction_cancel%> not within "
-		"%<__transaction%>");
+		"%<__transaction_atomic%>");
       return error_mark_node;
     }
 
diff --git a/gcc/cp/parser.h b/gcc/cp/parser.h
index be19099..8d7665d 100644
--- a/gcc/cp/parser.h
+++ b/gcc/cp/parser.h
@@ -325,7 +325,8 @@ typedef struct GTY(()) cp_parser {
      a local class.  */
   bool in_function_body;
 
-  /* Nonzero if we're processing a __transaction statement.  */
+  /* Nonzero if we're processing a __transaction_atomic or
+     __transaction_relaxed statement.  */
   unsigned char in_transaction;
 
   /* TRUE if we can auto-correct a colon to a scope operator.  */
diff --git a/gcc/cp/pt.c b/gcc/cp/pt.c
index 5b7e44c..4213aa3 100644
--- a/gcc/cp/pt.c
+++ b/gcc/cp/pt.c
@@ -12907,7 +12907,7 @@ tsubst_expr (tree t, tree args, tsubst_flags_t complain, tree in_decl,
         flags |= (TRANSACTION_EXPR_OUTER (t) ? TM_STMT_ATTR_OUTER : 0);
         flags |= (TRANSACTION_EXPR_RELAXED (t) ? TM_STMT_ATTR_RELAXED : 0);
 
-        stmt = begin_transaction_stmt (input_location, NULL);
+        stmt = begin_transaction_stmt (input_location, NULL, flags);
         tmp = RECUR (TRANSACTION_EXPR_BODY (t));
         finish_transaction_stmt (stmt, NULL, flags);
       }
diff --git a/gcc/cp/semantics.c b/gcc/cp/semantics.c
index 750909c..934cc37 100644
--- a/gcc/cp/semantics.c
+++ b/gcc/cp/semantics.c
@@ -4834,12 +4834,12 @@ finish_omp_taskyield (void)
   finish_expr_stmt (stmt);
 }
 
-/* Begin a __transaction statement.  If PCOMPOUND is non-null, this is
-   for a function-transaction-block, and we should create an extra
-   compound stmt.  */
+/* Begin a __transaction_atomic or __transaction_relaxed statement.
+   If PCOMPOUND is non-null, this is for a function-transaction-block, and we
+   should create an extra compound stmt.  */
 
 tree
-begin_transaction_stmt (location_t loc, tree *pcompound)
+begin_transaction_stmt (location_t loc, tree *pcompound, int flags)
 {
   tree r;
 
@@ -4852,15 +4852,19 @@ begin_transaction_stmt (location_t loc, tree *pcompound)
   if (flag_tm)
     add_stmt (r);
   else
-    error_at (loc, "%<__transaction%> without transactional memory "
-	      "support enabled");
+    error_at (loc, ((flags & TM_STMT_ATTR_RELAXED) != 0 ?
+        "%<__transaction_relaxed%> without transactional memory "
+	      "support enabled"
+        : "%<__transaction_atomic%> without transactional memory "
+	      "support enabled"));
 
   TRANSACTION_EXPR_BODY (r) = push_stmt_list ();
   return r;
 }
 
-/* End a __transaction statement.  If COMPOUND_STMT is non-null, this is
-   for a function-transaction-block, and we should end the compound.  */
+/* End a __transaction_atomic or __transaction_relaxed statement.
+   If COMPOUND_STMT is non-null, this is for a function-transaction-block,
+   and we should end the compound.  */
 
 void
 finish_transaction_stmt (tree stmt, tree compound_stmt, int flags)
diff --git a/gcc/gimple-pretty-print.c b/gcc/gimple-pretty-print.c
index 988dcdb..e25178a 100644
--- a/gcc/gimple-pretty-print.c
+++ b/gcc/gimple-pretty-print.c
@@ -1353,11 +1353,12 @@ dump_gimple_transaction (pretty_printer *buffer, gimple gs, int spc, int flags)
     }
   else
     {
-      pp_string (buffer, "__transaction");
       if (subcode & GTMA_IS_OUTER)
-	pp_string (buffer, " [[outer]]");
+	pp_string (buffer, "__transaction_atomic [[outer]]");
       else if (subcode & GTMA_IS_RELAXED)
-	pp_string (buffer, " [[relaxed]]");
+	pp_string (buffer, "__transaction_relaxed");
+      else
+	pp_string (buffer, "__transaction_atomic");
       subcode &= ~GTMA_DECLARATION_MASK;
 
       if (subcode || gimple_transaction_label (gs))
diff --git a/gcc/gimple.def b/gcc/gimple.def
index 2dc6fc6..8139a61 100644
--- a/gcc/gimple.def
+++ b/gcc/gimple.def
@@ -124,7 +124,8 @@ DEFGSCODE(GIMPLE_ASM, "gimple_asm", GSS_ASM)
     CHAIN is the optional static chain link for nested functions.  */
 DEFGSCODE(GIMPLE_CALL, "gimple_call", GSS_CALL)
 
-/* GIMPLE_TRANSACTION <BODY, LABEL> represents __transaction blocks.
+/* GIMPLE_TRANSACTION <BODY, LABEL> represents __transaction_atomic and
+   __transaction_relaxed blocks.
    BODY is the sequence of statements inside the transaction.
    LABEL is a label for the statement immediately following the
    transaction.  This is before RETURN so that it has MEM_OPS,
diff --git a/gcc/gimple.h b/gcc/gimple.h
index fad049e..14c198f 100644
--- a/gcc/gimple.h
+++ b/gcc/gimple.h
@@ -771,7 +771,8 @@ struct GTY(()) gimple_statement_omp_atomic_store {
 
 /* Bits to be stored in the GIMPLE_TRANSACTION subcode.  */
 
-/* The __transaction was declared [[outer]] or [[relaxed]].  */
+/* The __transaction_atomic was declared [[outer]] or it is
+   __transaction_relaxed.  */
 #define GTMA_IS_OUTER			(1u << 0)
 #define GTMA_IS_RELAXED			(1u << 1)
 #define GTMA_DECLARATION_MASK		(GTMA_IS_OUTER | GTMA_IS_RELAXED)
diff --git a/gcc/testsuite/c-c++-common/tm/20100127.c b/gcc/testsuite/c-c++-common/tm/20100127.c
index c390222..c25336d 100644
--- a/gcc/testsuite/c-c++-common/tm/20100127.c
+++ b/gcc/testsuite/c-c++-common/tm/20100127.c
@@ -15,7 +15,7 @@ static int set_remove(int * val)
 {
   int result;
   int * v;
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     v = next->val;
     result = (v == val);
     if (result)
diff --git a/gcc/testsuite/c-c++-common/tm/abort-2.c b/gcc/testsuite/c-c++-common/tm/abort-2.c
index 65c2c7d..727c634 100644
--- a/gcc/testsuite/c-c++-common/tm/abort-2.c
+++ b/gcc/testsuite/c-c++-common/tm/abort-2.c
@@ -4,7 +4,7 @@
 int g;
 void f(void)
 {
-  __transaction {
+  __transaction_atomic {
     if (g == 0)
       __transaction_cancel;
   }
diff --git a/gcc/testsuite/c-c++-common/tm/abort-3.c b/gcc/testsuite/c-c++-common/tm/abort-3.c
index 075f17c..f2cf5b5 100644
--- a/gcc/testsuite/c-c++-common/tm/abort-3.c
+++ b/gcc/testsuite/c-c++-common/tm/abort-3.c
@@ -2,7 +2,7 @@
 
 void f(void)
 {
-  __transaction {	   /* { dg-error "__transaction. without trans" } */
+  __transaction_atomic {   /* { dg-error "__transaction_atomic. without trans" } */
     __transaction_cancel;  /* { dg-error "_cancel. without trans" } */
   }
 }
diff --git a/gcc/testsuite/c-c++-common/tm/atomic-1.c b/gcc/testsuite/c-c++-common/tm/atomic-1.c
index d9d37c7..e301f1f 100644
--- a/gcc/testsuite/c-c++-common/tm/atomic-1.c
+++ b/gcc/testsuite/c-c++-common/tm/atomic-1.c
@@ -3,7 +3,7 @@
 int g;
 void f(void)
 {
-  __transaction {	/* { dg-error "without transactional memory" } */
+  __transaction_atomic {	/* { dg-error "without transactional memory" } */
     g++;
   }
 }
diff --git a/gcc/testsuite/c-c++-common/tm/atomic-2.c b/gcc/testsuite/c-c++-common/tm/atomic-2.c
index 27db749..f232766 100644
--- a/gcc/testsuite/c-c++-common/tm/atomic-2.c
+++ b/gcc/testsuite/c-c++-common/tm/atomic-2.c
@@ -4,7 +4,7 @@
 int g;
 void f(void)
 {
-  __transaction {
+  __transaction_atomic {
     g++;
   }
 }
diff --git a/gcc/testsuite/c-c++-common/tm/inline-asm.c b/gcc/testsuite/c-c++-common/tm/inline-asm.c
index 6b028e1..eefd347 100644
--- a/gcc/testsuite/c-c++-common/tm/inline-asm.c
+++ b/gcc/testsuite/c-c++-common/tm/inline-asm.c
@@ -10,7 +10,7 @@ inline_death ()
 void
 tranfunction ()
 {
-  __transaction
+  __transaction_atomic
     {
       inline_death ();
     }
diff --git a/gcc/testsuite/c-c++-common/tm/ipa-1.c b/gcc/testsuite/c-c++-common/tm/ipa-1.c
index 7610590..961f7fe 100644
--- a/gcc/testsuite/c-c++-common/tm/ipa-1.c
+++ b/gcc/testsuite/c-c++-common/tm/ipa-1.c
@@ -8,7 +8,7 @@ extern void func();
 int set_remove(void)
 {
   int result = 8;
-  __transaction  {
+  __transaction_atomic  {
     result = george;
     if (val)
       goto out;
diff --git a/gcc/testsuite/c-c++-common/tm/malloc.c b/gcc/testsuite/c-c++-common/tm/malloc.c
index 1c6f696..4ae82e9 100644
--- a/gcc/testsuite/c-c++-common/tm/malloc.c
+++ b/gcc/testsuite/c-c++-common/tm/malloc.c
@@ -8,7 +8,7 @@ char *z;
 void foobar(void)
 {
     char *p, *q;
-    __transaction {
+    __transaction_atomic {
         p = (char *)malloc(123);
         q = (char *)calloc(555,1);
         free(q);
diff --git a/gcc/testsuite/c-c++-common/tm/omp.c b/gcc/testsuite/c-c++-common/tm/omp.c
index 1bff42a..b9fcc76 100644
--- a/gcc/testsuite/c-c++-common/tm/omp.c
+++ b/gcc/testsuite/c-c++-common/tm/omp.c
@@ -15,7 +15,7 @@ void ParClassify()
 #pragma omp parallel private(Parent)
   {
     times_t inside;
-    __transaction [[atomic]] {
+    __transaction_atomic {
        inside.enter = rdtsc();
     }
   }
diff --git a/gcc/testsuite/c-c++-common/tm/outer-1.c b/gcc/testsuite/c-c++-common/tm/outer-1.c
index e8bf216..7dbf2e8 100644
--- a/gcc/testsuite/c-c++-common/tm/outer-1.c
+++ b/gcc/testsuite/c-c++-common/tm/outer-1.c
@@ -7,13 +7,13 @@ void
 f(void)
 {
   mco();			/* { dg-error "" } */
-  __transaction {
+  __transaction_atomic {
     mco();			/* { dg-error "" } */
   }
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     mco();			/* { dg-error "" } */
   }
-  __transaction [[outer]] {
+  __transaction_atomic [[outer]] {
     mco();
   }
 }
@@ -22,10 +22,10 @@ void __attribute__((transaction_may_cancel_outer))
 g(void)
 {
   mco();
-  __transaction {
+  __transaction_atomic {
     mco();
   }
-  __transaction [[outer]] {	/* { dg-error "" } */
+  __transaction_atomic [[outer]] {	/* { dg-error "" } */
     mco();
   }
 }
diff --git a/gcc/testsuite/c-c++-common/tm/safe-2.c b/gcc/testsuite/c-c++-common/tm/safe-2.c
index e254c53..b6d0486 100644
--- a/gcc/testsuite/c-c++-common/tm/safe-2.c
+++ b/gcc/testsuite/c-c++-common/tm/safe-2.c
@@ -7,19 +7,19 @@ void
 f(void)
 {
   mco();			/* { dg-error "" } */
-  __transaction {
+  __transaction_atomic {
     mco();			/* { dg-error "" } */
   }
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     mco();			/* { dg-error "" } */
   }
-  __transaction [[outer]] {
+  __transaction_atomic [[outer]] {
     mco();
   }
-  __transaction [[outer]] {
-    __transaction {
-      __transaction {
-        __transaction {
+  __transaction_atomic [[outer]] {
+    __transaction_atomic {
+      __transaction_atomic {
+        __transaction_atomic {
           mco();
         }
       }
@@ -31,10 +31,10 @@ void __attribute__((transaction_may_cancel_outer))
 g(void)
 {
   mco();
-  __transaction {
-    __transaction {
-      __transaction {
-        __transaction {
+  __transaction_atomic {
+    __transaction_atomic {
+      __transaction_atomic {
+        __transaction_atomic {
           mco();
         }
       }
diff --git a/gcc/testsuite/c-c++-common/tm/safe-3.c b/gcc/testsuite/c-c++-common/tm/safe-3.c
index eebc03f..8a883db 100644
--- a/gcc/testsuite/c-c++-common/tm/safe-3.c
+++ b/gcc/testsuite/c-c++-common/tm/safe-3.c
@@ -28,17 +28,17 @@ test_mco (void)
 void
 test_atomic (void)
 {
-  __transaction {
+  __transaction_atomic {
     f_extern ();	/* { dg-error "unsafe function call" } */
     f_first ();
     f_later ();
   }
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     f_extern ();
     f_first ();
     f_later ();
   }
-  __transaction [[outer]] {
+  __transaction_atomic [[outer]] {
     f_extern ();	/* { dg-error "unsafe function call" } */
     f_first ();
     f_later ();
diff --git a/gcc/testsuite/c-c++-common/tm/trxn-expr.c b/gcc/testsuite/c-c++-common/tm/trxn-expr.c
index 2409878..53d4677 100644
--- a/gcc/testsuite/c-c++-common/tm/trxn-expr.c
+++ b/gcc/testsuite/c-c++-common/tm/trxn-expr.c
@@ -5,7 +5,7 @@ int y, x, york;
 
 void foobar(void)
 {
-  x = y + __transaction (york);
+  x = y + __transaction_atomic (york);
 }
 
 /* { dg-final { scan-tree-dump-times "_ITM_RU.*york" 1 "tmmark" } } */
diff --git a/gcc/testsuite/c-c++-common/tm/wrap-1.c b/gcc/testsuite/c-c++-common/tm/wrap-1.c
index 8c63a31..04b5b6f 100644
--- a/gcc/testsuite/c-c++-common/tm/wrap-1.c
+++ b/gcc/testsuite/c-c++-common/tm/wrap-1.c
@@ -4,7 +4,7 @@
 void orig(void);
 void xyzzy(void) __attribute__((transaction_wrap (orig)));
 
-void foo() { __transaction [[relaxed]] { orig (); } }
+void foo() { __transaction_relaxed { orig (); } }
 
 /* { dg-final { scan-tree-dump-times "xyzzy" 1 "optimized" } } */
 /* { dg-final { cleanup-tree-dump "optimized" } } */
diff --git a/gcc/testsuite/g++.dg/tm/20100429.C b/gcc/testsuite/g++.dg/tm/20100429.C
index 9054d2f..087ce32 100644
--- a/gcc/testsuite/g++.dg/tm/20100429.C
+++ b/gcc/testsuite/g++.dg/tm/20100429.C
@@ -7,7 +7,7 @@ int
 bar(int a)
 {
   int r;
-  __transaction
+  __transaction_atomic
     {
       r = foo(a); // { dg-error "unsafe function call 'int foo\\(int\\)'" }
     }
diff --git a/gcc/testsuite/g++.dg/tm/fatomic-1.C b/gcc/testsuite/g++.dg/tm/fatomic-1.C
index e589405..dac850a 100644
--- a/gcc/testsuite/g++.dg/tm/fatomic-1.C
+++ b/gcc/testsuite/g++.dg/tm/fatomic-1.C
@@ -7,4 +7,4 @@ struct S
   S();
 };
 
-S::S() __transaction : i(1), j(2), k(3) { }
+S::S() __transaction_atomic : i(1), j(2), k(3) { }
diff --git a/gcc/testsuite/g++.dg/tm/nested-1.C b/gcc/testsuite/g++.dg/tm/nested-1.C
index d30315e..1f4e19e 100644
--- a/gcc/testsuite/g++.dg/tm/nested-1.C
+++ b/gcc/testsuite/g++.dg/tm/nested-1.C
@@ -15,7 +15,7 @@ __attribute__((transaction_safe))
 int HashTree::add_element2()
 {
  int tt;
-  __transaction [[atomic]] {
+  __transaction_atomic {
     tt = Count;
  }
  return tt;
diff --git a/gcc/testsuite/g++.dg/tm/nested-2.C b/gcc/testsuite/g++.dg/tm/nested-2.C
index 2995d6a..c6d5d98 100644
--- a/gcc/testsuite/g++.dg/tm/nested-2.C
+++ b/gcc/testsuite/g++.dg/tm/nested-2.C
@@ -29,11 +29,11 @@ int HashTree::add_element()
  int tt;
  if (Hash_table[val] == __null)
  {
-  __transaction [[atomic]] {
+  __transaction_atomic {
     Hash_table[val] = new HashTree;
   }
  }
- __transaction [[atomic]] {
+ __transaction_atomic {
  tt = Count++;
  enter_exit_times_inside.exit_atomicsec_time = 5;
  }
diff --git a/gcc/testsuite/g++.dg/tm/nested-3.C b/gcc/testsuite/g++.dg/tm/nested-3.C
index 4bca488..19718c3 100644
--- a/gcc/testsuite/g++.dg/tm/nested-3.C
+++ b/gcc/testsuite/g++.dg/tm/nested-3.C
@@ -31,11 +31,11 @@ int HashTree::add_element()
  int tt;
  if (Hash_table[val] == __null)
  {
-  __transaction [[atomic]] {
+  __transaction_atomic {
     Hash_table[val] = new HashTree;
   }
  }
- __transaction [[atomic]] {
+ __transaction_atomic {
  tt = Count++;
  enter_exit_times_inside.exit_atomicsec_time = 5;
  }
diff --git a/gcc/testsuite/g++.dg/tm/opt-1.C b/gcc/testsuite/g++.dg/tm/opt-1.C
index b9d2094..8dd41a5 100644
--- a/gcc/testsuite/g++.dg/tm/opt-1.C
+++ b/gcc/testsuite/g++.dg/tm/opt-1.C
@@ -6,4 +6,4 @@ struct S
   virtual void f() __attribute__((transaction_safe));
 };
 
-void f(S *s) { __transaction { s->f(); } }
+void f(S *s) { __transaction_atomic { s->f(); } }
diff --git a/gcc/testsuite/g++.dg/tm/pr45940-2.C b/gcc/testsuite/g++.dg/tm/pr45940-2.C
index bc4b3a4..e740716 100644
--- a/gcc/testsuite/g++.dg/tm/pr45940-2.C
+++ b/gcc/testsuite/g++.dg/tm/pr45940-2.C
@@ -24,7 +24,7 @@ public:
 sp_counted_base *base;
 
 void here(){
-  __transaction[[atomic]] {
+  __transaction_atomic {
     base->release();
   }
 }
diff --git a/gcc/testsuite/g++.dg/tm/pr45940-3.C b/gcc/testsuite/g++.dg/tm/pr45940-3.C
index e55e935..e89b5a2 100644
--- a/gcc/testsuite/g++.dg/tm/pr45940-3.C
+++ b/gcc/testsuite/g++.dg/tm/pr45940-3.C
@@ -61,7 +61,7 @@ shared_count * c;
 int main()
 {
   if ( i == 0) {
-    __transaction [[atomic]] {
+    __transaction_atomic {
      shared_count sc;
     }
   }
diff --git a/gcc/testsuite/g++.dg/tm/pr45940-4.C b/gcc/testsuite/g++.dg/tm/pr45940-4.C
index dae0cb8..02bb9b1 100644
--- a/gcc/testsuite/g++.dg/tm/pr45940-4.C
+++ b/gcc/testsuite/g++.dg/tm/pr45940-4.C
@@ -61,7 +61,7 @@ shared_count * c;
 int main()
 {
   if ( i == 0) {
-    __transaction [[atomic]] {
+    __transaction_atomic {
      shared_count sc;
     }
   }
diff --git a/gcc/testsuite/g++.dg/tm/pr45940.C b/gcc/testsuite/g++.dg/tm/pr45940.C
index f8122e8..f9573f4 100644
--- a/gcc/testsuite/g++.dg/tm/pr45940.C
+++ b/gcc/testsuite/g++.dg/tm/pr45940.C
@@ -24,7 +24,7 @@ public:
 sp_counted_base *base;
 
 void here(){
-  __transaction[[atomic]] {
+  __transaction_atomic {
     base->release();
   }
 }
diff --git a/gcc/testsuite/g++.dg/tm/pr46270.C b/gcc/testsuite/g++.dg/tm/pr46270.C
index 6407188..d7c97f8 100644
--- a/gcc/testsuite/g++.dg/tm/pr46270.C
+++ b/gcc/testsuite/g++.dg/tm/pr46270.C
@@ -17,7 +17,7 @@ static std::list<std::list<Game::BuildProject>::iterator> erasableBuildProjects;
 
 static void *buildProjectSyncStepConcurrently(int id, int localTeam)
 {
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     std::list<std::list<Game::BuildProject>::iterator>::iterator it
       = erasableBuildProjects.begin();
     game.buildProjects.erase( (std::list<Game::BuildProject> 
diff --git a/gcc/testsuite/g++.dg/tm/pr46300.C b/gcc/testsuite/g++.dg/tm/pr46300.C
index 3108b87..c67ae51 100644
--- a/gcc/testsuite/g++.dg/tm/pr46300.C
+++ b/gcc/testsuite/g++.dg/tm/pr46300.C
@@ -2,7 +2,7 @@
 // { dg-options "-fgnu-tm" }
 
 void foo(){
-    __transaction {
+    __transaction_atomic {
         throw 5;
     }
 }
diff --git a/gcc/testsuite/g++.dg/tm/pr46567.C b/gcc/testsuite/g++.dg/tm/pr46567.C
index 4b77650..e1d1fb9 100644
--- a/gcc/testsuite/g++.dg/tm/pr46567.C
+++ b/gcc/testsuite/g++.dg/tm/pr46567.C
@@ -2664,7 +2664,7 @@ static std::list<std::list<Game::BuildProject>::iterator>
 erasableBuildProjects;
 void *buildProjectSyncStepConcurrently(int id, int localTeam)
 {
-        __transaction [[relaxed]] {
+        __transaction_relaxed {
                 std::list<std::list<Game::BuildProject>::iterator>::iterator it
 = erasableBuildProjects.begin();
 		foobarit();
diff --git a/gcc/testsuite/g++.dg/tm/pr46653.C b/gcc/testsuite/g++.dg/tm/pr46653.C
index 2239ad0..f8f3a1d 100644
--- a/gcc/testsuite/g++.dg/tm/pr46653.C
+++ b/gcc/testsuite/g++.dg/tm/pr46653.C
@@ -11,7 +11,7 @@ public:
 shared_count * c;
 int main()
 {
-  __transaction [[atomic]] {
+  __transaction_atomic {
     shared_count sc;
   }
   return 0;
diff --git a/gcc/testsuite/g++.dg/tm/pr47530.C b/gcc/testsuite/g++.dg/tm/pr47530.C
index 226989d..9ea2cd1 100644
--- a/gcc/testsuite/g++.dg/tm/pr47530.C
+++ b/gcc/testsuite/g++.dg/tm/pr47530.C
@@ -55,7 +55,7 @@ LinkedList::LinkedList() : head(new LLNode(-1, 0)) { }
 
 void LinkedList::insert(int val)
 {
-  __transaction [[atomic]] {
+  __transaction_atomic {
     LLNode* prev = head;
     LLNode* curr = head->get_next();
 
diff --git a/gcc/testsuite/g++.dg/tm/pr47554.C b/gcc/testsuite/g++.dg/tm/pr47554.C
index 77c3519..28841bb 100644
--- a/gcc/testsuite/g++.dg/tm/pr47554.C
+++ b/gcc/testsuite/g++.dg/tm/pr47554.C
@@ -20,7 +20,7 @@ class list
 static list buildProjects;
 static void build()
 {
-	__transaction [[relaxed]]
+	__transaction_relaxed
 	{
 		buildProjects._M_get_Tp_allocator();
 	}
diff --git a/gcc/testsuite/gcc.dg/tm/20091013.c b/gcc/testsuite/gcc.dg/tm/20091013.c
index bc9f478..945ec12 100644
--- a/gcc/testsuite/gcc.dg/tm/20091013.c
+++ b/gcc/testsuite/gcc.dg/tm/20091013.c
@@ -6,7 +6,7 @@ int g,i;
 
 f()
 {
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     for (i=0; i < 10; ++i)
       ringo(g);  
   }
diff --git a/gcc/testsuite/gcc.dg/tm/20091221.c b/gcc/testsuite/gcc.dg/tm/20091221.c
index 1d21323..1d75d15 100644
--- a/gcc/testsuite/gcc.dg/tm/20091221.c
+++ b/gcc/testsuite/gcc.dg/tm/20091221.c
@@ -6,7 +6,7 @@ extern void virgin () __attribute__((transaction_pure));
 
 foo()
 {
-	__transaction {
+	__transaction_atomic {
 	    virgin(i);
 	}
 }
diff --git a/gcc/testsuite/gcc.dg/tm/20100125.c b/gcc/testsuite/gcc.dg/tm/20100125.c
index 0544cc5..3f1dd10 100644
--- a/gcc/testsuite/gcc.dg/tm/20100125.c
+++ b/gcc/testsuite/gcc.dg/tm/20100125.c
@@ -7,7 +7,7 @@ int trxn;
 
 void set_remove(int * val)
 {
-  __transaction {
+  __transaction_atomic {
       trxn = 5;
   }
   george();
diff --git a/gcc/testsuite/gcc.dg/tm/20100609.c b/gcc/testsuite/gcc.dg/tm/20100609.c
index 68e556f..91223e9 100644
--- a/gcc/testsuite/gcc.dg/tm/20100609.c
+++ b/gcc/testsuite/gcc.dg/tm/20100609.c
@@ -7,7 +7,7 @@ int later;
 
 void MyFunc()
 {
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
         funcNoReturn();
         later=8;
   }
diff --git a/gcc/testsuite/gcc.dg/tm/20100615.c b/gcc/testsuite/gcc.dg/tm/20100615.c
index 022d5f5..3d9e468 100644
--- a/gcc/testsuite/gcc.dg/tm/20100615.c
+++ b/gcc/testsuite/gcc.dg/tm/20100615.c
@@ -28,7 +28,7 @@ static node_t *new_node(node_t *next)
 static node_t *set_new()
 {
   node_t *min, *max;
-  __transaction [[atomic]] {
+  __transaction_atomic {
     max = new_node(NULL);
     min = new_node(max);
   }
diff --git a/gcc/testsuite/gcc.dg/tm/20110216.c b/gcc/testsuite/gcc.dg/tm/20110216.c
index 25f038b..22edae0 100644
--- a/gcc/testsuite/gcc.dg/tm/20110216.c
+++ b/gcc/testsuite/gcc.dg/tm/20110216.c
@@ -6,7 +6,7 @@ int george;
 __attribute__((transaction_callable))
 void q1()
 {
-  __transaction [[atomic]] {
+  __transaction_atomic {
       george=999;
   }
   q1();
diff --git a/gcc/testsuite/gcc.dg/tm/alias-1.c b/gcc/testsuite/gcc.dg/tm/alias-1.c
index 1b3cd53..4a913da 100644
--- a/gcc/testsuite/gcc.dg/tm/alias-1.c
+++ b/gcc/testsuite/gcc.dg/tm/alias-1.c
@@ -30,7 +30,7 @@ ptrcast ui;
 
 void f(void)
 {
-  __transaction {
+  __transaction_atomic {
     ui = TM_LOAD  (&mystruct);
     mystruct = (struct mystruct_type *) ui;
     ui = TM_LOAD  (&someptr);
diff --git a/gcc/testsuite/gcc.dg/tm/data-1.c b/gcc/testsuite/gcc.dg/tm/data-1.c
index cceae95..1606147 100644
--- a/gcc/testsuite/gcc.dg/tm/data-1.c
+++ b/gcc/testsuite/gcc.dg/tm/data-1.c
@@ -24,7 +24,7 @@ static long double gld;
 
 void f(void)
 {
-  __transaction {
+  __transaction_atomic {
     gc++;
     gsc++;
     guc++;
diff --git a/gcc/testsuite/gcc.dg/tm/data-2.c b/gcc/testsuite/gcc.dg/tm/data-2.c
index 3f3611e..3e2a604 100644
--- a/gcc/testsuite/gcc.dg/tm/data-2.c
+++ b/gcc/testsuite/gcc.dg/tm/data-2.c
@@ -16,7 +16,7 @@ void f(void)
   struct S l;
   fill(&l);
 
-  __transaction {
+  __transaction_atomic {
     g = l;
   }
 }
diff --git a/gcc/testsuite/gcc.dg/tm/debug-1.c b/gcc/testsuite/gcc.dg/tm/debug-1.c
index 035bddf..94e1b3b 100644
--- a/gcc/testsuite/gcc.dg/tm/debug-1.c
+++ b/gcc/testsuite/gcc.dg/tm/debug-1.c
@@ -11,7 +11,7 @@ testing(){
 
 main() {
         b = 9898;
-	__transaction [[relaxed]] {
+	__transaction_relaxed {
 	        z = c;
 		a = 888;
 		testing();
diff --git a/gcc/testsuite/gcc.dg/tm/indirect-1.c b/gcc/testsuite/gcc.dg/tm/indirect-1.c
index fe8798f..eade848 100644
--- a/gcc/testsuite/gcc.dg/tm/indirect-1.c
+++ b/gcc/testsuite/gcc.dg/tm/indirect-1.c
@@ -3,7 +3,7 @@
 
 void foo(void (*fn)(void))
 {
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     fn();
   }
 }
diff --git a/gcc/testsuite/gcc.dg/tm/ipa-1.c b/gcc/testsuite/gcc.dg/tm/ipa-1.c
index 7c433b1..ec1cdca 100644
--- a/gcc/testsuite/gcc.dg/tm/ipa-1.c
+++ b/gcc/testsuite/gcc.dg/tm/ipa-1.c
@@ -5,7 +5,7 @@ void foo(void) __attribute__((transaction_safe));
 
 void bar(void)
 {
-  __transaction {
+  __transaction_atomic {
     foo();
   }
 }
diff --git a/gcc/testsuite/gcc.dg/tm/ipa-2.c b/gcc/testsuite/gcc.dg/tm/ipa-2.c
index 57606f6..e7a02cb 100644
--- a/gcc/testsuite/gcc.dg/tm/ipa-2.c
+++ b/gcc/testsuite/gcc.dg/tm/ipa-2.c
@@ -5,7 +5,7 @@ void foo(void);
 
 void bar(void)
 {
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     foo();
   }
 }
diff --git a/gcc/testsuite/gcc.dg/tm/irrevocable-1.c b/gcc/testsuite/gcc.dg/tm/irrevocable-1.c
index 2408d0c..60f6291 100644
--- a/gcc/testsuite/gcc.dg/tm/irrevocable-1.c
+++ b/gcc/testsuite/gcc.dg/tm/irrevocable-1.c
@@ -8,7 +8,7 @@ extern crap() __attribute__((transaction_unsafe));
 
 foo()
 {
-	__transaction [[relaxed]] {
+	__transaction_relaxed {
 		global++;
 		crap();
 		george++;
diff --git a/gcc/testsuite/gcc.dg/tm/irrevocable-2.c b/gcc/testsuite/gcc.dg/tm/irrevocable-2.c
index 83ad435..17ac8a5 100644
--- a/gcc/testsuite/gcc.dg/tm/irrevocable-2.c
+++ b/gcc/testsuite/gcc.dg/tm/irrevocable-2.c
@@ -9,7 +9,7 @@ int george;
 
 foo()
 {
-	__transaction [[relaxed]] {
+	__transaction_relaxed {
 		global++;
 		__builtin__ITM_changeTransactionMode (0);
 		george++;
diff --git a/gcc/testsuite/gcc.dg/tm/irrevocable-3.c b/gcc/testsuite/gcc.dg/tm/irrevocable-3.c
index 46602e2..c085479 100644
--- a/gcc/testsuite/gcc.dg/tm/irrevocable-3.c
+++ b/gcc/testsuite/gcc.dg/tm/irrevocable-3.c
@@ -5,7 +5,7 @@ extern void bar(void) __attribute__((transaction_callable));
 
 foo()
 {
-	__transaction [[relaxed]] {
+	__transaction_relaxed {
 		bar();
 	}
 }
diff --git a/gcc/testsuite/gcc.dg/tm/irrevocable-4.c b/gcc/testsuite/gcc.dg/tm/irrevocable-4.c
index fea2e96..ee759b8 100644
--- a/gcc/testsuite/gcc.dg/tm/irrevocable-4.c
+++ b/gcc/testsuite/gcc.dg/tm/irrevocable-4.c
@@ -7,7 +7,7 @@ void xyz(void) __attribute__((transaction_wrap (orig)));
 
 foo()
 {
-	__transaction [[relaxed]] {
+	__transaction_relaxed {
 		orig();
 	}
 }
diff --git a/gcc/testsuite/gcc.dg/tm/irrevocable-5.c b/gcc/testsuite/gcc.dg/tm/irrevocable-5.c
index e3bd7fc..155879f 100644
--- a/gcc/testsuite/gcc.dg/tm/irrevocable-5.c
+++ b/gcc/testsuite/gcc.dg/tm/irrevocable-5.c
@@ -10,7 +10,7 @@ void danger(void) __attribute__((transaction_unsafe));
 void wildthing()
 {
   /* All blocks should be propagated as irrevocable.  */
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     if (a)
       foo();
     else
diff --git a/gcc/testsuite/gcc.dg/tm/irrevocable-6.c b/gcc/testsuite/gcc.dg/tm/irrevocable-6.c
index 6518bf0..2399131 100644
--- a/gcc/testsuite/gcc.dg/tm/irrevocable-6.c
+++ b/gcc/testsuite/gcc.dg/tm/irrevocable-6.c
@@ -10,7 +10,7 @@ void danger(void) __attribute__((transaction_unsafe));
 void wildthing()
 {
   /* All blocks should be propagated as irrevocable.  */
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     if (eee) {
       if (a)
 	foo();
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-1.c b/gcc/testsuite/gcc.dg/tm/memopt-1.c
index ebef0fe..06d4f64 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-1.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-1.c
@@ -8,7 +8,7 @@ int i;
 
 f()
 {
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     g = 666;
     george();
     if (i == 9)
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-10.c b/gcc/testsuite/gcc.dg/tm/memopt-10.c
index 0266646..5caa6b5 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-10.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-10.c
@@ -10,7 +10,7 @@ int f()
 
   p = malloc (sizeof (*p) * 100);
 
-  __transaction {
+  __transaction_atomic {
     /* p[5] is thread private, but not transaction local since the
        malloc is outside of the transaction.  We can use the logging
        functions for this.  */
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-11.c b/gcc/testsuite/gcc.dg/tm/memopt-11.c
index 316e0e1..07972a4 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-11.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-11.c
@@ -11,7 +11,7 @@ int f()
   p = malloc (sizeof (*p) * 100);
   foo(p[5]);
 
-  __transaction {
+  __transaction_atomic {
     /* p[5] is thread private, however the SSA_NAME that holds the
        address dominates the entire transaction (transaction
        invariant) so we can use a save/restore pair.  */
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-12.c b/gcc/testsuite/gcc.dg/tm/memopt-12.c
index f37f9bb..5520ece 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-12.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-12.c
@@ -11,7 +11,7 @@ int f()
   int *p1, *p2, *p3;
 
   p1 = malloc (sizeof (*p1)*5000);
-  __transaction {
+  __transaction_atomic {
     *p1 = 0;
 
     p2 = malloc (sizeof (*p2)*6000);
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-13.c b/gcc/testsuite/gcc.dg/tm/memopt-13.c
index 9b7e5e5..6e93b7f 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-13.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-13.c
@@ -7,7 +7,7 @@ extern struct large function (void) __attribute__((transaction_safe));
 
 void f()
 {
-  __transaction {
+  __transaction_atomic {
       large_global = function();
   }
 }
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-15.c b/gcc/testsuite/gcc.dg/tm/memopt-15.c
index 00a6f05..975c794 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-15.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-15.c
@@ -15,7 +15,7 @@ vectype f()
 
   p = malloc (sizeof (*p) * 100);
 
-  __transaction {
+  __transaction_atomic {
     /* p[5] is thread private, but not transaction local since the
        malloc is outside of the transaction.  We can use the logging
        functions for this.  */
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-2.c b/gcc/testsuite/gcc.dg/tm/memopt-2.c
index be946a2..08aa9ac 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-2.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-2.c
@@ -5,7 +5,7 @@ char c;
 
 void f(void)
 {
-  __transaction {
+  __transaction_atomic {
     ++c;
   }
 }
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-3.c b/gcc/testsuite/gcc.dg/tm/memopt-3.c
index a4d95ab..7733716 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-3.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-3.c
@@ -8,7 +8,7 @@ int f()
 {
   int i = readint();
   struct large lala = { 0 };
-  __transaction {
+  __transaction_atomic {
     lala.x[i] = 666;
     if (test())
       __transaction_cancel;
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-4.c b/gcc/testsuite/gcc.dg/tm/memopt-4.c
index c316eef..58f8fd0 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-4.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-4.c
@@ -11,7 +11,7 @@ int f()
 {
   int i = readint();
   struct large lala = bark();
-  __transaction {
+  __transaction_atomic {
     lala.x[55] = 666;
     if (test())
       __transaction_cancel;
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-5.c b/gcc/testsuite/gcc.dg/tm/memopt-5.c
index a766af6..53f7f12 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-5.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-5.c
@@ -11,7 +11,7 @@ int f()
 {
   int i = readint();
   struct large lala = bark();
-  __transaction {
+  __transaction_atomic {
     lala.x[55] = 666;
     if (test())
       __transaction_cancel;
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-6.c b/gcc/testsuite/gcc.dg/tm/memopt-6.c
index 5bd0918..f434373 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-6.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-6.c
@@ -10,7 +10,7 @@ int f()
 {
   int i = readint();
   struct large lala = bark();
-  __transaction {
+  __transaction_atomic {
     lala.x[55] = 666;
     lala = lacopy;		/* Aggregate instrumentation.  */
   }
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-7.c b/gcc/testsuite/gcc.dg/tm/memopt-7.c
index f282a63..d3a478b 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-7.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-7.c
@@ -11,7 +11,7 @@ int f()
 {
   struct large lala;
   struct large lacopy = foobie();
-  __transaction {
+  __transaction_atomic {
     lala = lacopy;
   }
   return lala.x[asdf];
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-8.c b/gcc/testsuite/gcc.dg/tm/memopt-8.c
index dc752e6..10320e7 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-8.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-8.c
@@ -9,7 +9,7 @@ int f()
 {
   int *p;
 
-  __transaction {
+  __transaction_atomic {
     p = malloc (sizeof (*p) * 100);
     escape (p);
 
diff --git a/gcc/testsuite/gcc.dg/tm/memopt-9.c b/gcc/testsuite/gcc.dg/tm/memopt-9.c
index 90692ff..0c34f20 100644
--- a/gcc/testsuite/gcc.dg/tm/memopt-9.c
+++ b/gcc/testsuite/gcc.dg/tm/memopt-9.c
@@ -11,7 +11,7 @@ int f()
   int *p;
   struct large *lp;
 
-  __transaction {
+  __transaction_atomic {
     p = malloc (sizeof (*p) * 100);
     lp = malloc (sizeof (*lp) * 100);
 
diff --git a/gcc/testsuite/gcc.dg/tm/memset-2.c b/gcc/testsuite/gcc.dg/tm/memset-2.c
index 1b54901..743ada1 100644
--- a/gcc/testsuite/gcc.dg/tm/memset-2.c
+++ b/gcc/testsuite/gcc.dg/tm/memset-2.c
@@ -7,7 +7,7 @@ void *memset(void *s, int c, __SIZE_TYPE__);
 
 int main()
 {
-  __transaction [[atomic]] {
+  __transaction_atomic {
     memset(array, 'b', sizeof(4));
   }
   return 0;
diff --git a/gcc/testsuite/gcc.dg/tm/memset.c b/gcc/testsuite/gcc.dg/tm/memset.c
index 33d3ce5..59a1c40 100644
--- a/gcc/testsuite/gcc.dg/tm/memset.c
+++ b/gcc/testsuite/gcc.dg/tm/memset.c
@@ -15,7 +15,7 @@ void *my_memset()
 int main()  
 {  
 
-	__transaction [[atomic]] {  
+	__transaction_atomic {  
 		my_memset();  
 	}  
 	return 0;  
diff --git a/gcc/testsuite/gcc.dg/tm/nested-1.c b/gcc/testsuite/gcc.dg/tm/nested-1.c
index 7e8ec9c..cf500d1 100644
--- a/gcc/testsuite/gcc.dg/tm/nested-1.c
+++ b/gcc/testsuite/gcc.dg/tm/nested-1.c
@@ -4,13 +4,13 @@
 extern int foo(int) __attribute__((transaction_safe));
 void bar(void)
 {
-  __transaction {
+  __transaction_atomic {
     if (foo(1))
-      __transaction {
+      __transaction_atomic {
         if (foo(2))
-          __transaction {
+          __transaction_atomic {
             if (foo(3))
-              __transaction {
+              __transaction_atomic {
                 if (foo(4))
                   foo(5);
                 else
diff --git a/gcc/testsuite/gcc.dg/tm/nested-2.c b/gcc/testsuite/gcc.dg/tm/nested-2.c
index e7b81ff..205ca8d 100644
--- a/gcc/testsuite/gcc.dg/tm/nested-2.c
+++ b/gcc/testsuite/gcc.dg/tm/nested-2.c
@@ -3,7 +3,7 @@
 
 void foobar(void)
 {
-    __transaction {
+    __transaction_atomic {
        foobar();
     }
 }
@@ -13,7 +13,7 @@ void doit(void) __attribute__((transaction_safe));
 __attribute__((transaction_callable))
 void callable(void)
 {
-  __transaction {
+  __transaction_atomic {
     doit();
   }
 }
diff --git a/gcc/testsuite/gcc.dg/tm/opt-1.c b/gcc/testsuite/gcc.dg/tm/opt-1.c
index ce18eb2..87a8c72 100644
--- a/gcc/testsuite/gcc.dg/tm/opt-1.c
+++ b/gcc/testsuite/gcc.dg/tm/opt-1.c
@@ -20,7 +20,7 @@ void *hello(void *arg)
   int tmp = p->id;
   int tmp3;
   printf ("Thread reads %d.\n", tmp);
-  __transaction
+  __transaction_atomic
     {
       int tmp2 = gvar;
       usleep ((int) (10.0*rand()/(10+1.0))/100);
diff --git a/gcc/testsuite/gcc.dg/tm/opt-2.c b/gcc/testsuite/gcc.dg/tm/opt-2.c
index 98bba14..d9e2b8a 100644
--- a/gcc/testsuite/gcc.dg/tm/opt-2.c
+++ b/gcc/testsuite/gcc.dg/tm/opt-2.c
@@ -5,7 +5,7 @@ int foo(int *arr, int v)
 {
   int r = 0;
   int i;
-  __transaction {
+  __transaction_atomic {
     for (i = 0; i < 10; ++i)
       if (arr[i] < 27)
 	r += arr[i] += v;
diff --git a/gcc/testsuite/gcc.dg/tm/pr45985.c b/gcc/testsuite/gcc.dg/tm/pr45985.c
index decd268..c811840 100644
--- a/gcc/testsuite/gcc.dg/tm/pr45985.c
+++ b/gcc/testsuite/gcc.dg/tm/pr45985.c
@@ -7,7 +7,7 @@ void illegal();
 static int a = 0;
 void func()
 {
-  __transaction [[relaxed]] {
+  __transaction_relaxed {
     if( a == 0)
       illegal();
   }
diff --git a/gcc/testsuite/gcc.dg/tm/pr46654.c b/gcc/testsuite/gcc.dg/tm/pr46654.c
index 57929b7..bb63b68 100644
--- a/gcc/testsuite/gcc.dg/tm/pr46654.c
+++ b/gcc/testsuite/gcc.dg/tm/pr46654.c
@@ -6,7 +6,7 @@ extern void baz(int);
 int y;
 void foo(volatile int x)
 {
-  __transaction {
+  __transaction_atomic {
     x = 5; /* { dg-error "invalid volatile use of 'x' inside transaction" } */
     x += y;
     y++;
@@ -19,7 +19,7 @@ volatile int i = 0;
 
 void george()
 {
-  __transaction [[atomic]] {
+  __transaction_atomic {
    if (i == 2) /* { dg-error "invalid volatile use of 'i' inside transaction" } */
      i = 1;
   }
diff --git a/gcc/testsuite/gcc.dg/tm/pr47690.c b/gcc/testsuite/gcc.dg/tm/pr47690.c
index ffb0798..d18e2e1 100644
--- a/gcc/testsuite/gcc.dg/tm/pr47690.c
+++ b/gcc/testsuite/gcc.dg/tm/pr47690.c
@@ -5,7 +5,7 @@ int george;
 
 void q1()
 {
-  __transaction [[atomic]] {
+  __transaction_atomic {
       george=999;
   }
   q1();
diff --git a/gcc/testsuite/gcc.dg/tm/pr47905.c b/gcc/testsuite/gcc.dg/tm/pr47905.c
index fac82db..362ed48 100644
--- a/gcc/testsuite/gcc.dg/tm/pr47905.c
+++ b/gcc/testsuite/gcc.dg/tm/pr47905.c
@@ -6,7 +6,7 @@ void funcB();
 
 void *thread()
 {
-        __transaction [[relaxed]]
+        __transaction_relaxed
         {
                 funcA();
         };
diff --git a/gcc/testsuite/gcc.dg/tm/props-1.c b/gcc/testsuite/gcc.dg/tm/props-1.c
index 73cb9cc..8969024 100644
--- a/gcc/testsuite/gcc.dg/tm/props-1.c
+++ b/gcc/testsuite/gcc.dg/tm/props-1.c
@@ -5,7 +5,7 @@ int global;
 
 foo(int local)
 {
-  __transaction {
+  __transaction_atomic {
     local++;
     if (++global == 10)
       __transaction_cancel;
diff --git a/gcc/testsuite/gcc.dg/tm/props-2.c b/gcc/testsuite/gcc.dg/tm/props-2.c
index d85e784..9c97e3b 100644
--- a/gcc/testsuite/gcc.dg/tm/props-2.c
+++ b/gcc/testsuite/gcc.dg/tm/props-2.c
@@ -9,7 +9,7 @@ int george;
 extern crap() __attribute__((transaction_unsafe));
 
 foo(){
-    __transaction [[relaxed]] {
+    __transaction_relaxed {
         global++;
         crap();
         george++;
diff --git a/gcc/testsuite/gcc.dg/tm/props-3.c b/gcc/testsuite/gcc.dg/tm/props-3.c
index 189dd37..48f2230 100644
--- a/gcc/testsuite/gcc.dg/tm/props-3.c
+++ b/gcc/testsuite/gcc.dg/tm/props-3.c
@@ -6,7 +6,7 @@
 void (*indirect)(void);
 
 foo(){
-    __transaction [[relaxed]] {
+    __transaction_relaxed {
       (*indirect)();
     }
 }
diff --git a/gcc/testsuite/gcc.dg/tm/unsafe.c b/gcc/testsuite/gcc.dg/tm/unsafe.c
index b27e73b..824368a 100644
--- a/gcc/testsuite/gcc.dg/tm/unsafe.c
+++ b/gcc/testsuite/gcc.dg/tm/unsafe.c
@@ -9,5 +9,5 @@ static int func ()
 
 int main()
 {
-  __transaction { return func(); } /* { dg-error "unsafe function call .func. " } */
+  __transaction_atomic { return func(); } /* { dg-error "unsafe function call .func. " } */
 }
diff --git a/gcc/testsuite/gcc.dg/tm/wrap-3.c b/gcc/testsuite/gcc.dg/tm/wrap-3.c
index e8d2650..8be4040 100644
--- a/gcc/testsuite/gcc.dg/tm/wrap-3.c
+++ b/gcc/testsuite/gcc.dg/tm/wrap-3.c
@@ -7,7 +7,7 @@ void *p;
 
 void foo() 
 {
-  __transaction [[relaxed]] { free (p); }
+  __transaction_relaxed { free (p); }
 }
 
 /* { dg-final { scan-tree-dump-times "free" 0 "optimized" } } */
diff --git a/gcc/testsuite/gcc.dg/tm/wrap-4.c b/gcc/testsuite/gcc.dg/tm/wrap-4.c
index 9346366..9e1e70c 100644
--- a/gcc/testsuite/gcc.dg/tm/wrap-4.c
+++ b/gcc/testsuite/gcc.dg/tm/wrap-4.c
@@ -8,7 +8,7 @@ static void tootsie_roll () { bark(); }
 
 void foo()
 {
-  __transaction [[relaxed]] { candy(); }
+  __transaction_relaxed { candy(); }
 }
 
 /* { dg-final { scan-tree-dump-times "candy" 0 "optimized" } } */
diff --git a/gcc/tree-pretty-print.c b/gcc/tree-pretty-print.c
index fafefab..caaaec0 100644
--- a/gcc/tree-pretty-print.c
+++ b/gcc/tree-pretty-print.c
@@ -2254,11 +2254,12 @@ dump_generic_node (pretty_printer *buffer, tree node, int spc, int flags,
       break;
 
     case TRANSACTION_EXPR:
-      pp_string (buffer, "__transaction");
       if (TRANSACTION_EXPR_OUTER (node))
-	pp_string (buffer, " [[outer]]");
+	pp_string (buffer, "__transaction_atomic [[outer]]");
       else if (TRANSACTION_EXPR_RELAXED (node))
-	pp_string (buffer, " [[relaxed]]");
+	pp_string (buffer, "__transaction_relaxed");
+      else
+        pp_string (buffer, "__transaction_atomic");
       if (!(flags & TDF_SLIM) && TRANSACTION_EXPR_BODY (node))
         {
           newline_and_indent (buffer, spc);