public inbox for gcc-patches@gcc.gnu.org
 help / color / mirror / Atom feed
* [committed][nvptx] Fix boolean type test in write_fn_proto
@ 2020-09-09 12:34 Tom de Vries
  0 siblings, 0 replies; only message in thread
From: Tom de Vries @ 2020-09-09 12:34 UTC (permalink / raw)
  To: gcc-patches

Hi,

When running this libgomp testcase for nvptx accelerator:
...
/* { dg-do run } */
__uint128_t v;
int main () {
  #pragma omp target
  {
    __uint128_t exp = 2;
    __atomic_compare_exchange_n (&v, &exp, 7, false, __ATOMIC_RELEASE,
				 __ATOMIC_ACQUIRE);
  }
}
...
we run into this assert in write_fn_proto:
...
913             gcc_assert (type == boolean_type_node);
...

This happens when doing some special-handling code for
__atomic_compare_exchange_1/2/4/8/16.  The function decls have a parameter
called weak of type bool, which is skipped when writing the decl because
the corresponding libatomic functions do not have that parameter.  The assert
is there to verify that we skip the correct parameter.

However, we assert because we have different type of bools:
...
(gdb) call debug_generic_expr (type)
_Bool
(gdb) call debug_generic_expr (global_trees[TI_BOOLEAN_TYPE])
bool
...

Fix this by checking for TREE_CODE (type) == BOOLEAN_TYPE instead.

Tested libgomp on x86_64-linux with nvptx accelerator.

Likewise, tested that the test-case above does not ICE anymore.

Committed to trunk.

Thanks,
- Tom

[nvptx] Fix boolean type test in write_fn_proto

gcc/ChangeLog:

	PR target/96991
	* config/nvptx/nvptx.c (write_fn_proto): Fix boolean type check.

---
 gcc/config/nvptx/nvptx.c | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/gcc/config/nvptx/nvptx.c b/gcc/config/nvptx/nvptx.c
index 39d0275493a..6f393dfea01 100644
--- a/gcc/config/nvptx/nvptx.c
+++ b/gcc/config/nvptx/nvptx.c
@@ -910,7 +910,7 @@ write_fn_proto (std::stringstream &s, bool is_defn,
       if (not_atomic_weak_arg)
 	argno = write_arg_type (s, -1, argno, type, prototyped);
       else
-	gcc_assert (type == boolean_type_node);
+	gcc_assert (TREE_CODE (type) == BOOLEAN_TYPE);
     }
 
   if (stdarg_p (fntype))

^ permalink raw reply	[flat|nested] only message in thread

only message in thread, other threads:[~2020-09-09 12:34 UTC | newest]

Thread overview: (only message) (download: mbox.gz / follow: Atom feed)
-- links below jump to the message on this page --
2020-09-09 12:34 [committed][nvptx] Fix boolean type test in write_fn_proto Tom de Vries

This is a public inbox, see mirroring instructions
for how to clone and mirror all data and code used for this inbox;
as well as URLs for read-only IMAP folder(s) and NNTP newsgroup(s).