test(autoware_tensorrt_plugins): add reference kernel tests

[mojomex]

Summary

Adds CUDA kernel vs. known-good CPU reference tests for autoware_tensorrt_plugins:

• verifies argsort output against a CPU stable-sort reference
• verifies unique values, inverse indices, and counts against a CPU reference

Compatibility note

While validating the test, the unique counts check exposed an existing sentinel write bug, so this PR includes the minimal fix needed for the new test to pass: write the final range sentinel to range_ptr + num_out, not range_ptr + num_out * sizeof(int64_t) (range_ptr is an int64_t*).

See

• fix(autoware_tensorrt_plugins): correct CustomUnique count offset #12502

for a standalone version of that fix.

Validation

colcon build --packages-up-to autoware_tensorrt_plugins && \
colcon test-result --delete-yes &&
colcon test --packages-select autoware_tensorrt_plugins --event-handlers console_cohesion+ && \
colcon test-result --verbose

Look out for:

[==========] Running 2 tests from 1 test suite.
1: [----------] Global test environment set-up.
1: [----------] 2 tests from ReferenceKernelsTest
1: [ RUN      ] ReferenceKernelsTest.ArgsortMatchesCpuReference
1: [       OK ] ReferenceKernelsTest.ArgsortMatchesCpuReference (277 ms)
1: [ RUN      ] ReferenceKernelsTest.UniqueMatchesCpuReference
1: [       OK ] ReferenceKernelsTest.UniqueMatchesCpuReference (1 ms)
1: [----------] 2 tests from ReferenceKernelsTest (278 ms total)
1: 
1: [----------] Global test environment tear-down
1: [==========] 2 tests from 1 test suite ran. (278 ms total)
1: [  PASSED  ] 2 tests.

[github-actions]

Thank you for contributing to the Autoware project!

🚧 If your pull request is in progress, switch it to draft mode.

Please ensure:

• You've checked our contribution guidelines.
• Your PR follows our pull request guidelines.
• All required CI checks pass before marking the PR ready for review.

[codecov]

Codecov Report

✅ All modified and coverable lines are covered by tests.
✅ Project coverage is 0.22%. Comparing base (54af299) to head (5002ac5).

❗ There is a different number of reports uploaded between BASE (54af299) and HEAD (5002ac5). Click for more details.

HEAD has 1 upload less than BASE

Flag	BASE (54af299)	HEAD (5002ac5)
daily	1	0

Additional details and impacted files

@@             Coverage Diff             @@
##             main   #12561       +/-   ##
===========================================
- Coverage   18.64%    0.22%   -18.42%     
===========================================
  Files        1918       97     -1821     
  Lines      131362     3500   -127862     
  Branches    44502       25    -44477     
===========================================
- Hits        24489        8    -24481     
+ Misses      86760     3491    -83269     
+ Partials    20113        1    -20112     

Flag	Coverage Δ
daily	?
full-suite	0.22% <ø> (-18.42%)	⬇️

☔ View full report in Codecov by Sentry.
📢 Have feedback on the report? Share it here.

🚀 New features to boost your workflow:

• ❄️ Test Analytics: Detect flaky tests, report on failures, and find test suite problems.

[manato]

@mojomex
Thank you very much for improving the unit tests. In terms of test strictness, I left small suggestions. I would appreciate it if you consider them!

[manato]

Suggested change

	copy_to_device(input_d.get(), input);
	ASSERT_EQ(
	argsort(
	input_d.get(), output_d.get(), workspace_d.get(), input.size(),
	get_argsort_workspace_size(input.size()), stream.get()),
	cudaSuccess);
	cudaEvent_t copy_event;
	cudaEventCreate(&copy_event);
	copy_to_device(input_d.get(), input);
	cudaEventRecord(copy_event, 0); // record event on the default stream
	cudaStreamWaitEvent(stream.get(), copy_event, cudaEventWaitDefault);
	ASSERT_EQ(
	argsort(
	input_d.get(), output_d.get(), workspace_d.get(), input.size(),
	get_argsort_workspace_size(input.size()), stream.get()),
	cudaSuccess);

According to the memcpy synchronous behavior:

For transfers from pageable host memory to device memory, a stream sync is performed before the copy is initiated. The function will return once the pageable buffer has been copied to the staging memory for DMA transfer to device memory, but the DMA to final destination may not have completed.

Since input is pageable host memory, it would be better to insert explicit synchronization before operating on input_d.

Alternatively, we can skip this kind of explicit synchronization if copy_to_device takes a CUDA stream as an argument and performs cudaMemcpyAsync inside.

[manato]

same as the the case of argsort-_kernel_test.cpp. better to insert explicit sync