     ASSERT_EQUAL_QUIET(d_result, d_input.end());
 }
 DECLARE_VARIABLE_UNITTEST(TestForEach);
+
+
+template <typename T>
+void TestForEachN(const size_t n)
+{
+    const size_t output_size = std::min((size_t) 10, 2 * n);
+    
+    thrust::host_vector<T> h_input = unittest::random_integers<T>(n);
+
+    for(size_t i = 0; i < n; i++)
+        h_input[i] =  ((size_t) h_input[i]) % output_size;
+    
+    thrust::device_vector<T> d_input = h_input;
+
+    thrust::host_vector<T>   h_output(output_size, (T) 0);
+    thrust::device_vector<T> d_output(output_size, (T) 0);
+
+    mark_present_for_each<T> h_f;
+    mark_present_for_each<T> d_f;
+    h_f.ptr = &h_output[0];
+    d_f.ptr = (&d_output[0]).get();
+    
+    typename thrust::host_vector<T>::iterator h_result =
+      thrust::for_each_n(h_input.begin(), h_input.size(), h_f);
+
+    typename thrust::device_vector<T>::iterator d_result =
+      thrust::for_each_n(d_input.begin(), d_input.size(), d_f);
+
+    ASSERT_EQUAL(h_output, d_output);
+    ASSERT_EQUAL_QUIET(h_result, h_input.end());
+    ASSERT_EQUAL_QUIET(d_result, d_input.end());
+}
+DECLARE_VARIABLE_UNITTEST(TestForEachN);
 
 
 template <size_t N> __host__ __device__ void f   (int * x) { int temp = *x; f<N - 1>(x + 1); *x = temp;};

    thrust::for_each(input.begin(), input.end(), CopyFunctorWithManyRegisters<N>());
}
DECLARE_UNITTEST(TestForEachLargeRegisterFootprint);


template <typename T, unsigned int N>


    ASSERT_EQUAL_QUIET(h_data, d_data);
}


void TestForEachWithLargeTypes(void)
{
    _TestForEachWithLargeTypes<int,    1>();
