Fix device_delete and enable the test
Two types added to device_delete_allocator:
* value_type - allocators must have it, see is_allocator and
allocator_traits;
* system_type - required for destroy_range to run for_each_n on device;
MODIFY
test/test_device_delete.cpp
test/test_device_delete.cpp
@@ -29,6 +29,16 @@ struct Foo
{
}
+ __host__ __device__
+ ~Foo(void)
+ {
+#if defined(__CUDA_ARCH__) || defined(__HIP_DEVICE_COMPILE__)
+ // __device__ overload
+ if(set_me_upon_destruction != 0)
+ *set_me_upon_destruction = true;
+#endif
+ }
+
bool* set_me_upon_destruction;
};
@@ -44,7 +54,7 @@ TEST(DeviceDelete, TestDeviceDeleteDestructorInvocation)
ASSERT_EQ(false, destructor_flag[0]);
- // TODO: Known failure
- //thrust::device_delete(foo_ptr);
- //ASSERT_EQ(true, destructor_flag[0]);
+ thrust::device_delete(foo_ptr);
+
+ ASSERT_EQ(true, destructor_flag[0]);
}

MODIFY
thrust/detail/device_delete.inl
thrust/detail/device_delete.inl
@@ -29,7 +29,12 @@ namespace detail
{
// define an empty allocator class to use below
-struct device_delete_allocator {};
+template<typename T>
+struct device_delete_allocator
+{
+ typedef T value_type;
+ typedef thrust::device_system_tag system_type;
+};
}
@@ -38,10 +43,9 @@ template<typename T>
const size_t n)
{
// we can use device_allocator to destroy the range
- thrust::detail::device_delete_allocator a;
+ thrust::detail::device_delete_allocator<T> a;
thrust::detail::destroy_range(a, ptr, n);
thrust::device_free(ptr);
} // end device_delete()
} // end thrust
-

Fix device_delete and enable the test

Two types added to device_delete_allocator:
* value_type - allocators must have it, see is_allocator and
  allocator_traits;
* system_type - required for destroy_range to run for_each_n on device;