dylanyunlong/project_6
1
0
Fork 0
Code Issues Pull Requests Actions Projects Releases Wiki Activity
Files
main
project_6/cccl_upstream/thrust/testing/binary_search_descending.cu

122 lines
7.1 KiB
Plaintext
Raw Permalink Normal View History

[INFRA] Import NVIDIA/CCCL upstream as optimization reference library CCCL (CUDA C++ Core Libraries) provides: - CUB: device/block/warp-level GPU primitives (reduce, scan, sort, topk) - Thrust: high-level parallel algorithms (transform_reduce, sort, scan) - libcudacxx: CUDA C++ standard library (atomics, barriers, memory) - cudax: experimental features (memory resources, allocators) - Tuning policies: per-SM hardware-specific algorithm parameters Competition optimization vectors mapped to CCCL: - Output TPS (83% weight): warp_reduce, block_reduce, device_topk - Input TPS (14% weight): device_scan, block_load, prefetch - Cache TPS (3% weight): prefix caching strategy patterns - Memory (0.9 util): pooled/cached/buddy allocators Source: https://github.com/NVIDIA/cccl (shallow clone, HEAD only) License: Apache-2.0
2026-07-30 09:35:51 +00:00
#include <thrust/binary_search.h>
#include <thrust/functional.h>
#include <thrust/sequence.h>
#include <thrust/sort.h>
#include <unittest/unittest.h>
//////////////////////
// Scalar Functions //
//////////////////////
template <class Vector>
void TestScalarLowerBoundDescendingSimple()
{
using T = typename Vector::value_type;
Vector vec{8, 7, 5, 2, 0};
ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::lower_bound(vec.begin(), vec.end(), T{0}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::lower_bound(vec.begin(), vec.end(), T{1}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::lower_bound(vec.begin(), vec.end(), T{2}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::lower_bound(vec.begin(), vec.end(), T{3}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::lower_bound(vec.begin(), vec.end(), T{4}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::lower_bound(vec.begin(), vec.end(), T{5}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::lower_bound(vec.begin(), vec.end(), T{6}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::lower_bound(vec.begin(), vec.end(), T{7}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::lower_bound(vec.begin(), vec.end(), T{8}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::lower_bound(vec.begin(), vec.end(), T{9}, ::cuda::std::greater<T>()));
}
DECLARE_VECTOR_UNITTEST(TestScalarLowerBoundDescendingSimple);
template <class Vector>
void TestScalarUpperBoundDescendingSimple()
{
using T = typename Vector::value_type;
Vector vec{8, 7, 5, 2, 0};
ASSERT_EQUAL_QUIET(vec.begin() + 5, thrust::upper_bound(vec.begin(), vec.end(), T{0}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::upper_bound(vec.begin(), vec.end(), T{1}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 4, thrust::upper_bound(vec.begin(), vec.end(), T{2}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::upper_bound(vec.begin(), vec.end(), T{3}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::upper_bound(vec.begin(), vec.end(), T{4}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 3, thrust::upper_bound(vec.begin(), vec.end(), T{5}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::upper_bound(vec.begin(), vec.end(), T{6}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 2, thrust::upper_bound(vec.begin(), vec.end(), T{7}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 1, thrust::upper_bound(vec.begin(), vec.end(), T{8}, ::cuda::std::greater<T>()));
ASSERT_EQUAL_QUIET(vec.begin() + 0, thrust::upper_bound(vec.begin(), vec.end(), T{9}, ::cuda::std::greater<T>()));
}
DECLARE_VECTOR_UNITTEST(TestScalarUpperBoundDescendingSimple);
template <class Vector>
void TestScalarBinarySearchDescendingSimple()
{
using T = typename Vector::value_type;
Vector vec{8, 7, 5, 2, 0};
ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{0}, ::cuda::std::greater<T>()));
ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{1}, ::cuda::std::greater<T>()));
ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{2}, ::cuda::std::greater<T>()));
ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{3}, ::cuda::std::greater<T>()));
ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{4}, ::cuda::std::greater<T>()));
ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{5}, ::cuda::std::greater<T>()));
ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{6}, ::cuda::std::greater<T>()));
ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{7}, ::cuda::std::greater<T>()));
ASSERT_EQUAL(true, thrust::binary_search(vec.begin(), vec.end(), T{8}, ::cuda::std::greater<T>()));
ASSERT_EQUAL(false, thrust::binary_search(vec.begin(), vec.end(), T{9}, ::cuda::std::greater<T>()));
}
DECLARE_VECTOR_UNITTEST(TestScalarBinarySearchDescendingSimple);
template <class Vector>
void TestScalarEqualRangeDescendingSimple()
{
using T = typename Vector::value_type;
Vector vec{8, 7, 5, 2, 0};
ASSERT_EQUAL_QUIET(vec.begin() + 4,
thrust::equal_range(vec.begin(), vec.end(), T{0}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 4,
thrust::equal_range(vec.begin(), vec.end(), T{1}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 3,
thrust::equal_range(vec.begin(), vec.end(), T{2}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 3,
thrust::equal_range(vec.begin(), vec.end(), T{3}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 3,
thrust::equal_range(vec.begin(), vec.end(), T{4}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 2,
thrust::equal_range(vec.begin(), vec.end(), T{5}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 2,
thrust::equal_range(vec.begin(), vec.end(), T{6}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 1,
thrust::equal_range(vec.begin(), vec.end(), T{7}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 0,
thrust::equal_range(vec.begin(), vec.end(), T{8}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 0,
thrust::equal_range(vec.begin(), vec.end(), T{9}, ::cuda::std::greater<T>()).first);
ASSERT_EQUAL_QUIET(vec.begin() + 5,
thrust::equal_range(vec.begin(), vec.end(), T{0}, ::cuda::std::greater<T>()).second);
ASSERT_EQUAL_QUIET(vec.begin() + 4,
thrust::equal_range(vec.begin(), vec.end(), T{1}, ::cuda::std::greater<T>()).second);
ASSERT_EQUAL_QUIET(vec.begin() + 4,
thrust::equal_range(vec.begin(), vec.end(), T{2}, ::cuda::std::greater<T>()).second);
ASSERT_EQUAL_QUIET(vec.begin() + 3,
thrust::equal_range(vec.begin(), vec.end(), T{3}, ::cuda::std::greater<T>()).second);
ASSERT_EQUAL_QUIET(vec.begin() + 3,
thrust::equal_range(vec.begin(), vec.end(), T{4}, ::cuda::std::greater<T>()).second);
ASSERT_EQUAL_QUIET(vec.begin() + 3,
thrust::equal_range(vec.begin(), vec.end(), T{5}, ::cuda::std::greater<T>()).second);
ASSERT_EQUAL_QUIET(vec.begin() + 2,
thrust::equal_range(vec.begin(), vec.end(), T{6}, ::cuda::std::greater<T>()).second);
ASSERT_EQUAL_QUIET(vec.begin() + 2,
thrust::equal_range(vec.begin(), vec.end(), T{7}, ::cuda::std::greater<T>()).second);
ASSERT_EQUAL_QUIET(vec.begin() + 1,
thrust::equal_range(vec.begin(), vec.end(), T{8}, ::cuda::std::greater<T>()).second);
ASSERT_EQUAL_QUIET(vec.begin() + 0,
thrust::equal_range(vec.begin(), vec.end(), T{9}, ::cuda::std::greater<T>()).second);
}
DECLARE_VECTOR_UNITTEST(TestScalarEqualRangeDescendingSimple);
Reference in New Issue Copy Permalink