📅 2011-Apr-04 ⬩ ✍️ Ashwin Nanjappa ⬩ 🏷️ cuda, thrust, zipiterator ⬩ 📚 Archive
Thrust is a CUDA library of some of the most basic parallel algorithms that can be applied on data. It is written using C++ templates, looks like STL and thus consists of only header files!
One of the coolest yet confusing feature of Thrust is the zip_iterator. Most Thrust functions accept one or two vectors. The zip_iterator is useful when you want to apply the function on more than one or two vectors. It can be understood with a simple example.
Consider an application where each integer key is associated with a Foo object. In old-school serial applications, you would place the key and Foo together in a structure and use a vector of that structure. For parallel applications, such information is easier to manipulate if they are represented separately as a vector of keys and a vector of Foo objects:
#include <thrust/device_vector.h> typedef thrust::device_vector<int> IntVec; typedef thrust::device_vector<Foo> FooVec; IntVec keyVec;FooVec fooVec;
These keys and their corresponding Foo objects can be sorted easily using
thrust::sort_by_key. This function takes two vectors: a key vector and a value vector. It sorts the keys in the key vector and also the value in the value vector corresponding to the location of each key:
#include <thrust/sort.h> // Assuming keyVec and fooVec are of equal size ... thrust::sort_by_key( keyVec.begin(), keyVec.end(), fooVec.begin() );
What if the application has two or more vectors that need to be sorted along with the key vector? For example:
#include <thrust/device_vector.h> typedef thrust::device_vector<int> IntVec; typedef thrust::device_vector<Foo> FooVec; typedef thrust::device_vector<Bar> BarVec; IntVec keyVec; FooVec fooVec;BarVec barVec;
This is where zip_iterator is useful! It can be used to convert multiple vectors into tuples, so that they can be manipulated as one element. Sorting both the Foo vector and the Bar vector along with the key vector is now easy:
#include <thrust/sort.h> // Assuming keyVec, fooVec and barVec are of equal size ... thrust::sort_by_key( keyVec.begin(), keyVec.end(), thrust::make_zip_iterator( make_tuple( fooVec.begin(), barVec.begin() ) ) );
That is it! Any number of vectors can be zipped up like this into a tuple and can be used with any of Thrust’s functions! 😊
Tried with: Thrust 1.3.0 and CUDA 3.2