Thrust device_malloc and device_new

Question

What are the advantages for using Thrust device_malloc instead of the normal cudaMalloc and what does device_new do?

For device_malloc it seems the only reason to use it is that it's just a bit cleaner.

The device_new documentation says:

"device_new implements the placement new operator for types resident in device memory. device_new calls T's null constructor on a array of objects in device memory. No memory is allocated by this function."

Which I don't understand...

score 1 · Answer 1 · answered Apr 18 '13 at 11:15

So I think I found out one good use for device_new

It's basically a better way of initialising an object and copying it to the device, while holding a pointer to it on host.

so instead of doing:

Particle *dev_p;
cudaMalloc((void**)&(dev_p), sizeof(Particle));
cudaMemcpy(dev_p, &p, sizeof(Particle), cudaMemcpyHostToDevice);

test2<<<1,1>>>(dev_p);

I can just do:

thrust::device_ptr<Particle> p = thrust::device_new<Particle>(1);
test2<<<1,1>>>(thrust::raw_pointer_cast(p));

score 1 · Accepted Answer · edited Apr 19 '13 at 08:20

1

device_malloc returns the proper type of object if you plan on using Thrust for other things. There is normally no reason to use cudaMalloc if you are using Thrust. Encapsulating CUDA calls makes it easier and usually cleaner. The same thing goes for C++ and STL containers versus C-style arrays and malloc.

For device_new, you should read the following line of the documentation:

 template<typename T>
 device_ptr<T> thrust::device_new (device_ptr< void > p, const size_t n = 1)

p: A device_ptr to a region of device memory into which to construct one or many Ts.

Basically, this function can be used if memory has already been allocated. Only the default constructor will be called, and this will return a device_pointer casted to T's type.

On the other hand, the following method allocates memory and returns a device_ptr<T>:

template<typename T >
device_ptr<T> thrust::device_new (const size_t n = 1)

edited Apr 19 '13 at 08:20

Ashwin Nanjappa

76,204
83
211
292

answered Apr 18 '13 at 11:17

BenC

8,729
3
49
68

Cheers, I wonder why no one pointed out that I can use thrust device_new when [I posted a question about copying an object to device memory](http://stackoverflow.com/questions/16024087/copy-an-object-to-device) ! I understand that I still have to copy the fields inside manually, but the first step (copying the object itself) better be done with thrust.. – SpaceMonkey Apr 18 '13 at 11:21
Well, everyone does not use Thrust. For simple things, you do not need it, but it makes your code much clearer and readable for non-CUDA developers. It is a good thing to know how to do things without Thrust though, else you may be lost once you start debugging your code. – BenC Apr 18 '13 at 11:24

Thrust device_malloc and device_new

2 Answers2