https://gcc.gnu.org/bugzilla/show_bug.cgi?id=122281
Tobias Burnus <burnus at gcc dot gnu.org> changed:
What |Removed |Added
----------------------------------------------------------------------------
CC| |burnus at gcc dot gnu.org
Keywords| |wrong-code
--- Comment #3 from Tobias Burnus <burnus at gcc dot gnu.org> ---
(In reply to Benjamin Schulz from comment #1)
> Created attachment 62558 [details]
> archiv.tar.gz
>
> above, i postet a precompiled .ii file. Here is the full source code.
>
> I dont know why this suddenly happens...
The file seems identical to PR 122280's first attachment, but this time the
file to test is sparsetests.cpp
If I compile it as:
mpicxx -std=c++23 -fopenmp -I . sparsetests.cpp -g -O3
-foffload-options=nvptx-none=-march=sm_80
using the git version of g++ and CUDA 13 + Nvidia RTX A1000 6GB Laptop.
I get the following segfault (host code):
Thread 20 "a.out" received signal SIGSEGV, Segmentation fault.
[Switching to Thread 0x7fffed2fa6c0 (LWP 3041701)]
0x0000000000406c18 in _ZNK9DataBlockIdE8sparsityEv._omp_fn.2(void) () at
/dev/shm/z/datablock.h:703
703 #pragma omp atomic update
(gdb) list
698 #pragma omp parallel for simd shared(count)
699 for(size_t i=0; i<dpdatalength; i++)
700 {
701 if(dpdata[i]==0)
702 {
703 #pragma omp atomic update
704 count++;
705 }
with:
(gdb) p i
$1 = 61
(gdb) p dpdatalength
$2 = 64
(gdb) p dpdata[i]
$3 = 0
(gdb) p count
$4 = 0
* * *
Compiling with -O0 solves this - but it re-appears with -O1.
That's in line with valgrind:
==3042608== Thread 9:
==3042608== Invalid read of size 8
==3042608== at 0x4025C1: DataBlock<double>::sparsity() const [clone
._omp_fn.2] (datablock.h:703)
__attribute__((omp declare target))
void DataBlock<double>::sparsity (struct .omp_data_s.156 & restrict
.omp_data_i)
{
...
size_t * _29(D);
...
<bb 8> [local count: 620085900]:
[datablock.h:703:25 discrim 3] __atomic_fetch_add_8 (_29(D), 1, 0);
* * *
Compare this to -O0:
<bb 10> :
[datablock.h:703:25] _32 = .omp_data_i_8(D)->count;
[datablock.h:703:25 discrim 1] __atomic_fetch_add_8 (_32, 1, 0);
* * *
I wonder what removes the '= .omp_data_i_8(D)->count;' assignment with -O1 ?!?