[EPIC] Clarify support for CCCL headers with host-only translation units

[jrhemstad]

Is this a duplicate?

• I confirmed there appear to be no duplicate issues for this request and that I agree to the Code of Conduct

Area

General CCCL

Is your feature request related to a problem? Please describe.

As a user of CCCL code, there are certain headers and features that I'd like to be able to use independent of compiling with nvcc. In other words, I'd like to know which headers and functionality can be used with only a host compiler (i.e., no nvcc).

I'd expect CCCL to clearly document guarantees which headers/features should work in host-only TUs. Furthermore, I'd expect CCCL to thoroughly test these guarantees as part of its CI testing.

Describe the solution you'd like

Tasks

• Document guarantees about host-only support for libcudacxx headers
• Ensure CI testing for supported libcudacxx headers in host-only TUs #1658
• Document guarantees about ABI consistency for cuda:: types between host and device TUs
• Document guarantees about host-only support for Thrust headers
• Document (lack of) guarantees about ABI consistency for thrust:: types between host and device TUs
• Ensure CI testing for supported thrust headers in host-only TUs (Various and sundry fixes for Thrust's CPP backends. #1722)
• Document (lack of) guarantees about host-only support for CUB headers

Additional context

Off the top of my head, here's a quick summary of the status quo:

• Every cuda/ header should be able to be included in a host-only TU and compile succesfully. The ABI should always be consistent between host and device TUs.
• No cub/ header supports inclusion in host-only TUs and this will never change
• No thrust/ header is supported in host-only TUs when THRUST_DEVICE_SYSTEM=CUDA. Likewise, thrust:: types are not guaranteed to have a consistent ABI between host and device TUs
• thrust/ headers are supported in host-only TUs when THRUST_DEVICE_SYSTEM={CPP, OMP, TBB}

Related:

• [FEA]: Make Thrust types have a consistent ABI between host and device code #1262

[bernhardmgruber]

Here is the wording in the current documentation of libcu++:

• std:: / <*>: This is your host compiler's Standard Library that works in __host__ code only, although you
  can use the --expt-relaxed-constexpr flag to use any constexpr functions in __device__ code.
  libcu++ does not replace or interfere with host compiler's Standard Library.
• cuda::std:: / <cuda/std/*>: Conforming implementations of facilities from the Standard Library that work in
  __host__ __device__ code.
• cuda:: / <cuda/*>: Conforming extensions to the Standard Library that work in __host__ __device__ code.
• cuda::device / <cuda/device/*>: Conforming extensions to the Standard Library that work only in
  __device__ code.
• cuda::ptx: C++ convenience wrappers for inline PTX (only usable in __device__ code).

I discussed with @wmaxey that "works in __host__ code" implies that it can be compiled by a non-CUDA compiler like g++. We should add a corresponding clarifying comment to the documentation of libcu++.