   ASSERT_EQUAL(v[3], 3);
   ASSERT_EQUAL(v[4], 4);
 
-  tabulate_kernel<<<1,1>>>(v.begin(), v.end(), -_1);
+  tabulate_kernel<<<1,1>>>(exec, v.begin(), v.end(), -_1);
 
   ASSERT_EQUAL(v[0],  0);
   ASSERT_EQUAL(v[1], -1);

#include <thrust/execution_policy.h>


template<typename Iterator, typename Function>
__global__
void tabulate_kernel(Iterator first, Iterator last, Function f)
{
  thrust::tabulate(thrust::seq, first, last, f);
}


void TestTabulateDeviceSeq()
{
  typedef thrust::device_vector<int> Vector;
  using namespace thrust::placeholders;

  ASSERT_EQUAL(v[3], 27);
  ASSERT_EQUAL(v[4], 64);
}

void TestTabulateDeviceSeq()
{
  TestTabulateDevice(thrust::seq);
}
DECLARE_UNITTEST(TestTabulateDeviceSeq);

void TestTabulateDeviceDevice()
{
  TestTabulateDevice(thrust::device);
}
DECLARE_UNITTEST(TestTabulateDeviceDevice);

