问如何防止thrust::reduce_by_key写入可分页内存？

EN

#include <thrust/system/cuda/vector.h>
#include <thrust/host_vector.h>
#include <thrust/pair.h>
#include <iostream>
#include <thrust/device_vector.h>
#include <thrust/reduce.h>
#include <thrust/random.h>

int main(void)
{
  int N = 20;
  thrust::default_random_engine rng;
  thrust::uniform_int_distribution<int> dist(10, 99);

  // initialize data
  thrust::device_vector<int> array(N);
  for (size_t i = 0; i < array.size(); i++)
    array[i] = dist(rng);

  // allocate storage for sums and indices
  thrust::device_vector<int> sums(N);
  thrust::device_vector<int> indices(N);

  // make a pinned memory location for the returned pair of iterators
  typedef thrust::device_vector<int>::iterator  dIter;
  thrust::pair<dIter, dIter>*  new_end;

  const unsigned int bytes =  sizeof(thrust::pair<dIter, dIter>);
  cudaMallocHost((void**)&new_end, bytes);

  for(int i = 0 ; i< 20; i++){   // you can see in the profiler each operator writes 4 bytes to pageable memory

        *new_end = thrust::reduce_by_key
            (thrust::make_counting_iterator(0),
             thrust::make_counting_iterator(N),
             array.begin(),
             indices.begin(),
             sums.begin(),
             thrust::equal_to<int>(),
             thrust::plus<int>());
  }
  std::cout << "done \n";
  return 0;
}