Make stream and memory resource explicit in fixed_capacity_map APIs

[PointKernel]

Description

closes #9695

This PR removes the default CUDA stream and default memory resource from cuda::experimental::cuco::fixed_capacity_map, following the cuda::buffer design agreed on during the #7705 review discussion:

• All constructors now take cuda::stream_ref as the first argument and the memory resource as the second, matching cuda::buffer's argument order. The helper that silently picked the current device's default memory pool is removed.
• All host bulk APIs (clear, clear_async, insert, insert_async, contains, contains_async) take the stream as the first parameter, with no default.
• The internal __open_addressing_impl follows the same stream-first, memory-resource-second order.

Checklist

• New or existing tests cover these changes.
• The documentation is up to date with these changes.

[coderabbitai]

No actionable comments were generated in the recent review. 🎉

ℹ️ Recent review info

⚙️ Run configuration

Configuration used: Path: .coderabbit.yaml

Review profile: CHILL

Plan: Enterprise

Run ID: e7408540-f783-47c2-af8d-98c982686db7

📥 Commits

Reviewing files that changed from the base of the PR and between 7126ca5 and 30ffeb3.

📒 Files selected for processing (1)

• cudax/test/cuco/fixed_capacity_map/test_key_sentinel.cu

🚧 Files skipped from review as they are similar to previous changes (1)

• cudax/test/cuco/fixed_capacity_map/test_key_sentinel.cu

---

📝 Walkthrough

Summary by CodeRabbit

• Breaking Changes

  • Updated fixed_capacity_map and its underlying open-addressing implementation to require explicit cuda::stream_ref (and memory resource) for stream-executed operations.
  • Reordered stream parameters so cuda::stream_ref comes first, and removed default stream arguments from clear, clear_async, insert, insert_async, contains, and contains_async.

• Tests

  • Updated fixed_capacity_map tests to construct maps with an explicit CUDA stream and device default memory pool, and to use stream-aware insert/contains/clear APIs with device-side verification buffers.

Walkthrough

fixed_capacity_map and its backend now require explicit stream and memory-resource arguments, with stream-aware methods reordered to take the stream first. Tests were updated to construct explicit streams and device-default memory pools and pass them through map construction and operations.

Changes

Stream parameter reordering

Layer / File(s)	Summary
open_addressing_impl signature reordering cudax/include/cuda/experimental/__cuco/detail/open_addressing/open_addressing_impl.cuh	Constructors and insert/insert_async/contains_async reorder ::cuda::stream_ref and _MemoryResource to the front of parameter lists.
fixed_capacity_map wrapper API update cudax/include/cuda/experimental/__cuco/fixed_capacity_map.cuh	Constructors and clear, clear_async, insert, insert_async, contains, contains_async require an explicit leading stream and forward to the reordered backend calls; unused device/driver includes are removed.
Test updates for explicit stream/memory pool cudax/test/cuco/fixed_capacity_map/test_capacity.cu, cudax/test/cuco/fixed_capacity_map/test_insert_and_contains.cu, cudax/test/cuco/fixed_capacity_map/test_key_sentinel.cu, cudax/test/cuco/fixed_capacity_map/test_shared_memory.cu	Tests add memory-pool/stream includes, construct a cuda::stream and device-default memory pool, and pass them into map construction and stream-aware insert/contains/clear calls.

Assessment against linked issues

Objective	Addressed	Explanation
[#9695] Make MR and stream explicit for fixed capacity map	✅
[#9695] Update fixed-capacity map call sites to use the explicit stream/MR API	✅

Out-of-scope changes

None found.

---

_{Comment @coderabbitai help to get the list of available commands.}

[coderabbitai]

🧹 Nitpick comments (1)

cudax/test/cuco/fixed_capacity_map/test_shared_memory.cu (1)

84-89: 🩺 Stability & Availability | 🔵 Trivial | 💤 Low value

suggestion: prefer stream.sync() over cudaDeviceSynchronize() now that an explicit stream is used for the kernel launch — keeps synchronization scoped to the stream this test actually cares about.

Proposed fix

   insert_shmem_kernel<<<grid_size, block_size, 0, stream.get()>>>(
     map.ref(),
     cuda::transform_iterator(cuda::counting_iterator<int>{0}, iota_pair<fixed_capacity_map_512_type::value_type>{}),
     num_keys);
-  REQUIRE(cudaDeviceSynchronize() == cudaSuccess);
+  stream.sync();

Source: Path instructions

---

ℹ️ Review info

⚙️ Run configuration

Configuration used: Path: .coderabbit.yaml

Review profile: CHILL

Plan: Enterprise

Run ID: 16036a28-a5bc-42e5-bad4-e3eebd92a1a1

📥 Commits

Reviewing files that changed from the base of the PR and between 43db4b1 and 355d423.

📒 Files selected for processing (6)

• cudax/include/cuda/experimental/__cuco/detail/open_addressing/open_addressing_impl.cuh
• cudax/include/cuda/experimental/__cuco/fixed_capacity_map.cuh
• cudax/test/cuco/fixed_capacity_map/test_capacity.cu
• cudax/test/cuco/fixed_capacity_map/test_insert_and_contains.cu
• cudax/test/cuco/fixed_capacity_map/test_key_sentinel.cu
• cudax/test/cuco/fixed_capacity_map/test_shared_memory.cu

[github-actions]

🥳 CI Workflow Results

🟩 Finished in 1h 08m: Pass: 100%/55 | Total: 5h 06m | Max: 1h 08m | Hits: 91%/39400

See results here.