On Sun, 1 Sep 2024, Prathamesh Kulkarni wrote:
> Hi,
> For the following test:
> #include <arm_sve.h>
>
> int main()
> {
> svint32_t x;
> #pragma omp target map(x)
> x;
> return 0;
> }
>
> compiling with -fopenmp -foffload=nvptx-none results in following ICE:
>
> t_sve.c: In function 'main':
> t_sve.c:6:11: internal compiler error: Segmentation fault
> 6 | #pragma omp target map(x)
> | ^~~
> 0x228ed13 internal_error(char const*, ...)
> ../../gcc/gcc/diagnostic-global-context.cc:491
> 0xfcf68f crash_signal
> ../../gcc/gcc/toplev.cc:321
> 0xc17d9c omp_add_variable
> ../../gcc/gcc/gimplify.cc:7811
that's not on trunk head? Anyway, I think that instead
/* When adding a variable-sized variable, we have to handle all sorts
of additional bits of data: the pointer replacement variable, and
the parameters of the type. */
if (DECL_SIZE (decl) && TREE_CODE (DECL_SIZE (decl)) != INTEGER_CST)
should instead be checking for !POLY_INT_CST_P (DECl_SIZE (decl))
Richard.
> 0xc17d9c omp_add_variable
> ../../gcc/gcc/gimplify.cc:7752
> 0xc4176b gimplify_scan_omp_clauses
> ../../gcc/gcc/gimplify.cc:12881
> 0xc46d53 gimplify_omp_workshare
> ../../gcc/gcc/gimplify.cc:17139
> 0xc23383 gimplify_expr(tree_node**, gimple**, gimple**, bool (*)(tree_node*),
> int)
> ../../gcc/gcc/gimplify.cc:18668
> 0xc27f53 gimplify_stmt(tree_node**, gimple**)
> ../../gcc/gcc/gimplify.cc:7646
> 0xc24ef7 gimplify_statement_list
> ../../gcc/gcc/gimplify.cc:2250
> 0xc24ef7 gimplify_expr(tree_node**, gimple**, gimple**, bool (*)(tree_node*),
> int)
> ../../gcc/gcc/gimplify.cc:18565
> 0xc27f53 gimplify_stmt(tree_node**, gimple**)
> ../../gcc/gcc/gimplify.cc:7646
> 0xc289d3 gimplify_bind_expr
> ../../gcc/gcc/gimplify.cc:1642
> 0xc24b9b gimplify_expr(tree_node**, gimple**, gimple**, bool (*)(tree_node*),
> int)
> ../../gcc/gcc/gimplify.cc:18315
> 0xc27f53 gimplify_stmt(tree_node**, gimple**)
> ../../gcc/gcc/gimplify.cc:7646
> 0xc24ef7 gimplify_statement_list
> ../../gcc/gcc/gimplify.cc:2250
> 0xc24ef7 gimplify_expr(tree_node**, gimple**, gimple**, bool (*)(tree_node*),
> int)
> ../../gcc/gcc/gimplify.cc:18565
> 0xc27f53 gimplify_stmt(tree_node**, gimple**)
> ../../gcc/gcc/gimplify.cc:7646
> 0xc2aadb gimplify_body(tree_node*, bool)
> ../../gcc/gcc/gimplify.cc:19393
> 0xc2b05f gimplify_function_tree(tree_node*)
> ../../gcc/gcc/gimplify.cc:19594
> 0xa0e47f cgraph_node::analyze()
> ../../gcc/gcc/cgraphunit.cc:687
>
> The attached patch fixes the issue by checking if variable is VLA vector,
> and emits an error in that case since no accel currently supports VLA vectors.
> Does the patch look OK ?
>
> Signed-off-by: Prathamesh Kulkarni <prathame...@nvidia.com>
>
> Thanks,
> Prathamesh
>
>
--
Richard Biener <rguent...@suse.de>
SUSE Software Solutions Germany GmbH,
Frankenstrasse 146, 90461 Nuernberg, Germany;
GF: Ivo Totev, Andrew McDonald, Werner Knoblich; (HRB 36809, AG Nuernberg)
