AlexVlx wrote:

> @efriedma-quic was kind enough to have a call where we discussed this a bit 
> more. I'll update tomorrow with a potential way forward, for the group's 
> consideration.

Following up, here's a possible approach to making progress, broken down in 
phases, (@efriedma-quic can correct me if I am misrepresenting any of these):

1. Have what is proposed here as an initial step, with the addition that we 
issue warnings on unguarded uses of builtins / ASM (similar to what 
`__builtin_available` / `@available` do), and we clean-up non-extern functions 
that become unreachable as a consequence of predicate expansion (i.e. `foo` can 
only be called from within this module, and it was only being called from a 
predicate guarded block, which was removed);
2. Add attribute based checking for predicate guarded areas:
     - Functions can be annotated either with the existing `target` attribute 
or with a new `target_can_invoke` (name up for bike-shedding) attribute;
     - Within a predicate guarded scope, if we encounter contradictions, e.g. 
we call a `target("gfx9000")` function, or a 
`target_can_invoke(builtin_only_on_gfx9000)`, within a 
`__builtin_amdgcn_processor_is("gfx8999")`, that is an error
     - This should reward users that go through the effort of annotating their 
functions, making it much harder to write bugs
     - I'm not entirely sure how to do this well yet (nested guarded regions, 
where to track the currently active guard etc.), and it probably needs a bit 
more design, hence why it's a different phase
     - It is a pre-requisite for any attempt at making these general, rather 
than target specific
3. In relation with generalisation, if we go in that direction (i.e. other 
targets are interested / we think there's merit into hoisting these into 
generic Clang builtins), we will have to look at whether or not we want a 
different IR representation (possibly / probably along the lines of what has 
been discussed here), for cases where a target must run some potentially 
disruptive optimisations before and cannot just do the expansion right after 
Clang.

https://github.com/llvm/llvm-project/pull/134016
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits

Reply via email to