On 9/2/20 12:44 PM, Jakub Jelinek wrote: > On Wed, Sep 02, 2020 at 12:22:28PM +0200, Tom de Vries wrote: >> And test-case passes on x86_64 with this patch (obviously, in >> combination with trigger patch above). >> >> Jakub, WDYT? > > I guess the normal answer would be use libatomic, but it isn't ported for > nvptx.
Ah, I was not aware of that one, filed https://gcc.gnu.org/bugzilla/show_bug.cgi?id=96898 to look into that. > I guess at least temporarily this is ok, though I'm wondering why > you need __sync_*_16 rather than __atomic_*_16, That's what omp-expand.c uses in expand_omp_atomic_pipeline: BUILT_IN_SYNC_VAL_COMPARE_AND_SWAP_N . Thanks, - Tom > or perhaps both __sync_* and > __atomic_*. > > What happens if you try > unsigned __int128 v; > #pragma omp declare target (v) > int > main () > { > #pragma omp target > { > __atomic_add_fetch (&v, 1, __ATOMIC_RELAXED); > __atomic_fetch_add (&v, 1, __ATOMIC_RELAXED); > unsigned __int128v exp = 2; > __atomic_compare_exchange_n (&v, &expected, 7, 0, __ATOMIC_RELEASE, > __ATOMIC_ACQUIRE); > } > } > etc. (see some gcc.dg/atomic* tests, ditto for __sync_*)? > I guess better not to throw everything into one test, because not every > target supports them all (e.g. I think x86_64 doesn't really do 128-bit > atomic loads because the cmpxchg16b insn are not appropriate for .rodata > locations). > > Jakub >
