This is the mail archive of the gcc-patches@gcc.gnu.org mailing list for the GCC project.


Index Nav: [Date Index] [Subject Index] [Author Index] [Thread Index]
Message Nav: [Date Prev] [Date Next] [Thread Prev] [Thread Next]
Other format: [Raw text]

Re: [PATCH] error on missing LTO symbols


On Fri, Dec 14, 2018 at 10:21:35AM +0100, Tom de Vries wrote:
> Build and reg-tested on x86_64 with nvptx accelerator.
> 
> 2018-12-13  Tom de Vries  <tdevries@suse.de>
> 
> 	* lto-cgraph.c (verify_node_partition): New function.
> 	(input_overwrite_node, input_varpool_node): Use verify_node_partition.
> 
> 	* testsuite/libgomp.c-c++-common/function-not-offloaded.c: New test.
> 	* testsuite/libgomp.c-c++-common/variable-not-offloaded.c: New test.

> +      if (TREE_CODE (decl) == FUNCTION_DECL
> +	  || TREE_CODE (decl) == VAR_DECL)
> +	error_at (DECL_SOURCE_LOCATION (decl),
> +		  "%s %qs has been referenced in offloaded code but"
> +		  " hasn't been marked to be included in the offloaded code",
> +		  TREE_CODE (decl) == FUNCTION_DECL ? "function" : "variable",
> +		  name);

This is translation unfriendly.  Please just do:
      if (TREE_CODE (decl) == FUNCTION_DECL)
	error_at (...);
      else if (VAR_P (decl))
	error_at (...);
      else
	gcc_unreachable ();
(also note VAR_P).  And, please use hasn%'t instead of hasn't.

> @@ -1153,9 +1184,8 @@ input_overwrite_node (struct lto_file_decl_data *file_data,
>    node->resolution = bp_unpack_enum (bp, ld_plugin_symbol_resolution,
>  				     LDPR_NUM_KNOWN);
>    node->split_part = bp_unpack_value (bp, 1);
> -  gcc_assert (flag_ltrans
> -	      || (!node->in_other_partition
> -		  && !node->used_from_other_partition));
> +  verify_node_partition (node->decl, node->name (), node->in_other_partition,
> +			 node->used_from_other_partition);
>  }

Why are you passing all these arguments to that function (especially calling
node->name () even when you don't know if it will be needed or not)?
Doesn't both cgraph_node and varpool_node inherit from symtab_node, which
has all these 3 fields as well as name () method?
So, I think if verify_node_partition takes a symtab_node *node argument
and the callers just both do
  verify_node_partition (node);
it should work fine.  I'd make it inline too.

> --- /dev/null
> +++ b/libgomp/testsuite/libgomp.c-c++-common/function-not-offloaded.c
> @@ -0,0 +1,21 @@
> +/* { dg-do link } */
> +/* { dg-excess-errors "lto1, mkoffload and lto-wrapper fatal errors" } */
> +
> +#pragma omp declare target
> +int results[2000];
> +#pragma omp end declare target
> +
> +void __attribute__((noinline, noclone))
> +baz (int i) /* { dg-error "function 'baz' has been referenced in offloaded code but hasn't been marked to be included in the offloaded code" } */
> +{
> +  results[i]++;
> +}
> +
> +int
> +main ()
> +{
> +#pragma omp target
> +#pragma omp for
> +  for (int i = 0; i < 2000; i++)
> +    baz (i);
> +}

Note, this will be well defined in OpenMP 5.0, just the support isn't there.
The spec says:
"If a function (C, C++, Fortran) or subroutine (Fortran) is referenced in a target construct then
that function or subroutine is treated as if its name had appeared in a to clause on a
declare target directive."
and
"If a function is referenced in a function that appears as a list item in a to clause on a
declare target directive then the name of the referenced function is treated as if it had
appeared in a to clause on a declare target directive.

If a variable with static storage duration or a function (except lambda for C++) is referenced in the
initializer expression list of a variable with static storage duration that appears as a list item in a to
clause on a declare target directive then the name of the referenced variable or function is
treated as if it had appeared in a to clause on a declare target directive."

so for functions it should work through implicit propagation of the
"omp declare target" attribute.

Can't find a restriction I'd expect to see that if a declaration of a
function or variable is marked declare target then the definition has to be
as well, will talk to omp-lang.

In any case, for the above testcase it might be better to split it into two
sources with dg-auxiliary-source, declare baz in the source with main and
define in another file (where declare instead of define the variable).

> diff --git a/libgomp/testsuite/libgomp.c-c++-common/variable-not-offloaded.c b/libgomp/testsuite/libgomp.c-c++-common/variable-not-offloaded.c
> new file mode 100644
> index 00000000000..c2e1d57adea
> --- /dev/null
> +++ b/libgomp/testsuite/libgomp.c-c++-common/variable-not-offloaded.c
> @@ -0,0 +1,21 @@
> +/* { dg-do link } */
> +/* { dg-excess-errors "lto1, mkoffload and lto-wrapper fatal errors" } */
> +
> +int results[2000]; /* { dg-error "variable 'results' has been referenced in offloaded code but hasn't been marked to be included in the offloaded code" } */
> +
> +#pragma omp declare target
> +void  __attribute__((noinline, noclone))
> +baz (int i)
> +{
> +  results[i]++;
> +}
> +#pragma omp end declare target
> +
> +int
> +main ()
> +{
> +#pragma omp target
> +#pragma omp for
> +  for (int i = 0; i < 2000; i++)
> +    baz (i);
> +}

Perhaps just simplify the testcases too, the variable can be just
called var and be a scalar, baz can be renamed to foo, drop the i argument,
drop the #pragma omp for and loop from there too (in both tests).
So simply just call a function from target region that increments a
variable.

Otherwise LGTM.

	Jakub


Index Nav: [Date Index] [Subject Index] [Author Index] [Thread Index]
Message Nav: [Date Prev] [Date Next] [Thread Prev] [Thread Next]