Implement C API for device for

[gevtushenko]

To expose algorithms like transform, fill, copy, etc, we have to implement a C version of the cub::DeviceFor::ForEachN algorithm. Overall, development can be informed by C version of the reduce API. A few key points:

• We can ignore cub::DeviceFor::Bulk, cub::DeviceFor::ForEachCopy*, and cub::DeviceFor::ForEach for now and limit this issue to cub::DeviceFor::ForEachN
• We can ignore cub::DeviceFor::ForEachN overload that requieres temporary storage
• We can limit offset types to always be std::int64_t
• We can limit the tuning policy to be always set to 256 threads per block and 2 items per thread

Following the points above, I suggest the following API:

struct cccl_device_for_build_result_t
{
  int cc;
  void* cubin;
  size_t cubin_size;
  CUlibrary library;
  CUkernel static_kernel;
};

extern "C" CCCL_C_API CUresult cccl_device_for_build(
  cccl_device_for_build_result_t* build,
  cccl_iterator_t d_data,
  cccl_op_t op,
  int cc_major,
  int cc_minor,
  const char* cub_path,
  const char* thrust_path,
  const char* libcudacxx_path,
  const char* ctk_path) noexcept;

extern "C" CCCL_C_API CUresult cccl_device_for(
  cccl_device_for_build_result_t build,
  cccl_iterator_t d_data,
  int64_t num_items,
  cccl_op_t op,
  CUstream stream) noexcept;

extern "C" CCCL_C_API CUresult cccl_device_for_cleanup(cccl_device_for_build_result_t* bld_ptr);

Tests of this API can be inspired by cub::DeviceFor::ForEachN test and C reduction test. C library should be enabled in the all-dev preset. To build the tests, it's sufficient to write: cmake --build . --target cccl.c.test.reduce in cuda12.5-gcc13 devcontainer.

[gevtushenko]

Closed by #2378