Skip to content
This repository has been archived by the owner on Mar 21, 2024. It is now read-only.

Commit

Permalink
Add abstractions that use memory accessible from both hosts and devices.
Browse files Browse the repository at this point in the history
- `thrust::universal_vector`.
- `thrust::universal_ptr`.
- `thrust::universal_allocator`.

Change all backend fancy pointer and reference types to be aliases.

Substantially refactor `thrust::reference`.

Fix a bug that allowed `thrust::reference`s to const objects to be swapped:
https://godbolt.org/z/r9G4nY

Introduce a new `thrust::tagged_reference` type that breaks the circular
template argument dependency between `thrust::pointer` and `thrust::reference`.
  • Loading branch information
alliepiper committed Nov 30, 2020
1 parent 79c72ce commit b778e2a
Show file tree
Hide file tree
Showing 59 changed files with 1,702 additions and 2,801 deletions.
9 changes: 5 additions & 4 deletions CHANGELOG.md
Original file line number Diff line number Diff line change
Expand Up @@ -12,8 +12,8 @@ The new `thrust::shuffle` algorithm has been tweaked to improve the randomness
of the output.

Our CMake package and build system continue to see improvements with
improved `add_subdirectory` support, installation rules, status messages, and
other features that make CUB easier to use from CMake projects.
better `add_subdirectory` support, installation rules, status messages, and
other features that make Thrust easier to use from CMake projects.

The release includes several other bugfixes and modernizations, and received
updates from 12 contributors.
Expand Down Expand Up @@ -72,11 +72,12 @@ updates from 12 contributors.
- Github's `thrust/cub` repository is now `NVIDIA/cub`
- Development has moved from the `master` branch to the `main` branch.

# Thrust 1.10.0 (NVIDIA HPC SDK 20.9)
# Thrust 1.10.0 (NVIDIA HPC SDK 20.9, CUDA Toolkit 11.2)

## Summary

Thrust 1.10.0 is the major release accompanying the NVIDIA HPC SDK 20.9 release.
Thrust 1.10.0 is the major release accompanying the NVIDIA HPC SDK 20.9 release
and the CUDA Toolkit 11.2 release.
It drops support for C++03, GCC < 5, Clang < 6, and MSVC < 2017.
It also overhauls CMake support.
Finally, we now have a Code of Conduct for contributors:
Expand Down
7 changes: 4 additions & 3 deletions examples/sorting_aos_vs_soa.cu
Original file line number Diff line number Diff line change
@@ -1,3 +1,4 @@
#include <thrust/host_vector.h>
#include <thrust/device_vector.h>
#include <thrust/sort.h>
#include <thrust/random.h>
Expand All @@ -7,7 +8,7 @@

// This examples compares sorting performance using Array of Structures (AoS)
// and Structure of Arrays (SoA) data layout. Legacy applications will often
// store data in C/C++ structs, such as MyStruct defined below. Although
// store data in C/C++ structs, such as MyStruct defined below. Although
// Thrust can process array of structs, it is typically less efficient than
// the equivalent structure of arrays layout. In this particular example,
// the optimized SoA approach is approximately *five times faster* than the
Expand Down Expand Up @@ -57,7 +58,7 @@ int main(void)
{
size_t N = 2 * 1024 * 1024;

// Sort Key-Value pairs using Array of Structures (AoS) storage
// Sort Key-Value pairs using Array of Structures (AoS) storage
{
thrust::device_vector<MyStruct> structures(N);

Expand All @@ -71,7 +72,7 @@ int main(void)
std::cout << "AoS sort took " << 1e3 * t.elapsed() << " milliseconds" << std::endl;
}

// Sort Key-Value pairs using Structure of Arrays (SoA) storage
// Sort Key-Value pairs using Structure of Arrays (SoA) storage
{
thrust::device_vector<int> keys(N);
thrust::device_vector<float> values(N);
Expand Down
1 change: 1 addition & 0 deletions examples/transform_input_output_iterator.cu
Original file line number Diff line number Diff line change
@@ -1,3 +1,4 @@
#include <thrust/host_vector.h>
#include <thrust/device_vector.h>
#include <thrust/functional.h>
#include <thrust/gather.h>
Expand Down
141 changes: 0 additions & 141 deletions testing/cuda/managed_memory_pointer.cu

This file was deleted.

7 changes: 7 additions & 0 deletions testing/functional_placeholders_bitwise.cu
Original file line number Diff line number Diff line change
Expand Up @@ -24,6 +24,13 @@ template<typename T, typename U, typename Allocator>
typename Allocator::template rebind<U>::other> type;
};

template<typename T, typename U, typename Allocator>
struct rebind_vector<thrust::universal_vector<T, Allocator>, U>
{
typedef thrust::universal_vector<U,
typename Allocator::template rebind<U>::other> type;
};

#define BINARY_FUNCTIONAL_PLACEHOLDERS_TEST(name, op, reference_functor, type_list) \
template<typename Vector> \
struct TestFunctionalPlaceholders##name \
Expand Down
7 changes: 7 additions & 0 deletions testing/functional_placeholders_logical.cu
Original file line number Diff line number Diff line change
Expand Up @@ -23,6 +23,13 @@ template<typename T, typename U, typename Allocator>
typename Allocator::template rebind<U>::other> type;
};

template<typename T, typename U, typename Allocator>
struct rebind_vector<thrust::universal_vector<T, Allocator>, U>
{
typedef thrust::universal_vector<U,
typename Allocator::template rebind<U>::other> type;
};

#define BINARY_FUNCTIONAL_PLACEHOLDERS_TEST(name, reference_operator, functor) \
template<typename Vector> \
void TestFunctionalPlaceholders##name(void) \
Expand Down
7 changes: 7 additions & 0 deletions testing/functional_placeholders_relational.cu
Original file line number Diff line number Diff line change
Expand Up @@ -23,6 +23,13 @@ template<typename T, typename U, typename Allocator>
typename Allocator::template rebind<U>::other> type;
};

template<typename T, typename U, typename Allocator>
struct rebind_vector<thrust::universal_vector<T, Allocator>, U>
{
typedef thrust::universal_vector<U,
typename Allocator::template rebind<U>::other> type;
};

#define BINARY_FUNCTIONAL_PLACEHOLDERS_TEST(name, reference_operator, functor) \
template<typename Vector> \
void TestFunctionalPlaceholdersBinary##name(void) \
Expand Down
Loading

0 comments on commit b778e2a

Please sign in to comment.