   {
     for (size_t n = (1 << 20); n < ((size_t) 1 << 40); n *= 2)
     {
-      // TODO ideally we'd use the fallback_allocator in the device_vector too
-      //thrust::device_vector<int, fallback_allocator> d_vec(n);
+      // TODO ideally we'd use the fallback_allocator in the vector too
+      //thrust::cuda::vector<int, fallback_allocator> d_vec(n);
 
       std::cout << "attempting to sort " << n << " values" << std::endl;
 

      {
        kernel<<<100,256>>>(raw_ptr, n); // generate unsorted values

        thrust::device_ptr<int> begin = thrust::device_pointer_cast(raw_ptr);
        thrust::device_ptr<int> end   = begin + n;

        thrust::sort(thrust::retag<my_tag>(begin), thrust::retag<my_tag>(end));

        assert(thrust::is_sorted(begin, end));


// XXX ideally this variable is declared thread_local
fallback_allocator g_allocator;

// create a tag derived from cuda::tag for distinguishing
// our overloads of get_temporary_buffer and return_temporary_buffer
struct my_tag : thrust::cuda::tag {};


// overload get_temporary_buffer on my_tag
