CudaMemory.cu 10.5 KB
Newer Older
incardon's avatar
incardon committed
1
#include "config.h"
incardon's avatar
incardon committed
2
3
4
#include <cstddef>
#include "CudaMemory.cuh"
#include "cuda_macro.h"
5
#include <cstring>
incardon's avatar
incardon committed
6

7
8
#define CUDA_EVENT 0x1201

Pietro Incardona's avatar
Pietro Incardona committed
9
10
11
12
13
14
15
16
17
18
19
/*! \brief Move the memory into device
 *
 * \return true if the memory is correctly flushed
 *
 */
bool CudaMemory::flush()
{
	if (hm != NULL && dm != NULL)
	{
		//! copy from host to device memory

incardon's avatar
incardon committed
20
21
22
23
24
		#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
		CUDA_SAFE_CALL(cudaMemcpy(dm,hm,sz,cudaMemcpyHostToDevice));
		#else
		memcpy(dm,hm,sz);
		#endif
Pietro Incardona's avatar
Pietro Incardona committed
25
26
27
28
29
	}
	
	return true;
}

incardon's avatar
incardon committed
30
31
32
33
34
35
36
37
38
39
40
/*! \brief Allocate a chunk of memory
 *
 * Allocate a chunk of memory
 *
 * \param sz size of the chunk of memory to allocate in byte
 *
 */
bool CudaMemory::allocate(size_t sz)
{
	//! Allocate the device memory
	if (dm == NULL)
incardon's avatar
incardon committed
41
42
43
44
	{
		#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
		CUDA_SAFE_CALL(cudaMalloc(&dm,sz));
		#else
incardon's avatar
incardon committed
45
		if (sz != 0)
incardon's avatar
incardon committed
46
47
48
49
50
51
		{
			dm = new unsigned char[sz];
			#ifdef GARBAGE_INJECTOR
			memset(dm,0xFF,sz);
			#endif
		}
incardon's avatar
incardon committed
52
53
		#endif
	}
incardon's avatar
incardon committed
54
55
56
57
58
59
60
61
	else
	{
		if (sz != this->sz)
		{
			std::cout << __FILE__ << ":" << __LINE__ << " error FATAL: using allocate to resize the memory, please use resize." << std::endl;
			return false;
		}
	}
62
63
64

	this->sz = sz;

incardon's avatar
incardon committed
65
66
67
68
#ifdef FILL_CUDA_MEMORY_WITH_MINUS_ONE
	CUDA_SAFE_CALL(cudaMemset(dm,-1,sz))
#endif

69
	return true;
incardon's avatar
incardon committed
70
71
72
73
74
75
76
77
78
}

/*! \brief destroy a chunk of memory
 *
 * Destroy a chunk of memory
 *
 */
void CudaMemory::destroy()
{
incardon's avatar
incardon committed
79
80
81
	if (dm != NULL)
	{
		//! Release the allocated memory
incardon's avatar
incardon committed
82
		#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
83
		CUDA_SAFE_CALL(cudaFree(dm));
incardon's avatar
incardon committed
84
85
86
		#else
		delete [] (unsigned char *)dm;
		#endif
incardon's avatar
incardon committed
87
		dm = NULL;
incardon's avatar
incardon committed
88
89
90
91
92
	}

	if (hm != NULL)
	{
		//! we invalidate hm
incardon's avatar
incardon committed
93
		#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
94
		CUDA_SAFE_CALL(cudaFreeHost(hm));
incardon's avatar
incardon committed
95
96
97
		#else
		delete [] (unsigned char *)hm;
		#endif
incardon's avatar
incardon committed
98
		hm = NULL;
incardon's avatar
incardon committed
99
	}
Pietro Incardona's avatar
Pietro Incardona committed
100
101
	
	sz = 0;
incardon's avatar
incardon committed
102
103
}

incardon's avatar
incardon committed
104
105
106
107
108
109
110
111
112
113
/*! \brief copy memory from device to device
 *
 * \param external device pointer
 * \param start source starting point (where it start to copy)
 * \param stop end point
 * \param offset where to copy in the device pointer
 *
 */
void CudaMemory::deviceToDevice(void * ptr, size_t start, size_t stop, size_t offset)
{
incardon's avatar
incardon committed
114
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
115
	CUDA_SAFE_CALL(cudaMemcpy(((unsigned char *)dm)+offset,((unsigned char *)ptr)+start,(stop-start),cudaMemcpyDeviceToDevice));
incardon's avatar
incardon committed
116
117
118
	#else
	memcpy(((unsigned char *)dm)+offset,((unsigned char *)ptr)+start,(stop-start));
	#endif
incardon's avatar
incardon committed
119
120
}

121
/*! \brief Allocate the host buffer
incardon's avatar
incardon committed
122
 *
123
 * Allocate the host buffer
incardon's avatar
incardon committed
124
125
 *
 */
Pietro Incardona's avatar
Pietro Incardona committed
126
void CudaMemory::allocate_host(size_t sz) const
127
128
{
	if (hm == NULL)
incardon's avatar
incardon committed
129
	{
incardon's avatar
incardon committed
130
		#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
131
		CUDA_SAFE_CALL(cudaHostAlloc(&hm,sz,cudaHostAllocMapped))
incardon's avatar
incardon committed
132
133
		#else
		hm = new unsigned char[sz];
incardon's avatar
incardon committed
134
135
136
		#ifdef GARBAGE_INJECTOR
		memset(hm,0xFF,sz);
		#endif
incardon's avatar
incardon committed
137
		#endif
incardon's avatar
incardon committed
138
	}
139
140
141
142
143
144
145
146
147
}

/*! \brief copy the data from a pointer
 *
 * copy the data from a pointer
 *
 *	\param ptr
 *	\return true if success
 */
Pietro Incardona's avatar
Pietro Incardona committed
148
bool CudaMemory::copyFromPointer(const void * ptr)
incardon's avatar
incardon committed
149
150
151
{
	// check if we have a host buffer, if not allocate it

152
153
154
	allocate_host(sz);

	// get the device pointer
incardon's avatar
incardon committed
155

156
	void * dvp;
incardon's avatar
incardon committed
157
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
158
	CUDA_SAFE_CALL(cudaHostGetDevicePointer(&dvp,hm,0));
incardon's avatar
incardon committed
159

160
	// memory copy
incardon's avatar
incardon committed
161

Pietro Incardona's avatar
Pietro Incardona committed
162
	memcpy(dvp,ptr,sz);
incardon's avatar
incardon committed
163
164
165
	#else
	memcpy(hm,ptr,sz);
	#endif
166
167

	return true;
incardon's avatar
incardon committed
168
169
}

170
171
172
173
174
175
176
177
/*! \brief copy from device to device
 *
 * copy a piece of memory from device to device
 *
 * \param CudaMemory from where to copy
 *
 * \return true is success
 */
Pietro Incardona's avatar
Pietro Incardona committed
178
bool CudaMemory::copyDeviceToDevice(const CudaMemory & m)
incardon's avatar
incardon committed
179
{
180
181
182
183
184
185
186
	//! The source buffer is too big to copy it

	if (m.sz > sz)
	{
		std::cerr << "Error " << __LINE__ << __FILE__ << ": source buffer is too big to copy";
		return false;
	}
incardon's avatar
incardon committed
187

188
	//! Copy the memory
incardon's avatar
incardon committed
189
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
Pietro Incardona's avatar
Pietro Incardona committed
190
	CUDA_SAFE_CALL(cudaMemcpy(dm,m.dm,m.sz,cudaMemcpyDeviceToDevice));
incardon's avatar
incardon committed
191
192
193
	#else
	memcpy(dm,m.dm,m.sz);
	#endif
194
195

	return true;
incardon's avatar
incardon committed
196
197
}

198
199
200
201
202
203
204
/*! \brief copy from memory
 *
 * copy from memory
 *
 * \param m a memory interface
 *
 */
Pietro Incardona's avatar
Pietro Incardona committed
205
bool CudaMemory::copy(const memory & m)
incardon's avatar
incardon committed
206
207
{
	//! Here we try to cast memory into OpenFPMwdeviceCudaMemory
Pietro Incardona's avatar
Pietro Incardona committed
208
	const CudaMemory * ofpm = dynamic_cast<const CudaMemory *>(&m);
incardon's avatar
incardon committed
209
210
211
212
213
214
215

	//! if we fail we get the pointer and simply copy from the pointer

	if (ofpm == NULL)
	{
		// copy the memory from device to host and from host to device

216
		return copyFromPointer(m.getPointer());
incardon's avatar
incardon committed
217
218
219
220
221
	}
	else
	{
		// they are the same memory type, use cuda/thrust buffer copy

222
		return copyDeviceToDevice(*ofpm);
incardon's avatar
incardon committed
223
224
225
	}
}

226
227
228
229
230
231
232
/*! \brief Get the size of the allocated memory
 *
 * Get the size of the allocated memory
 *
 * \return the size of the allocated memory
 *
 */
incardon's avatar
incardon committed
233

Pietro Incardona's avatar
Pietro Incardona committed
234
size_t CudaMemory::size() const
incardon's avatar
incardon committed
235
{
236
	return sz;
incardon's avatar
incardon committed
237
238
}

Pietro Incardona's avatar
Pietro Incardona committed
239

240
241
242
243
244
245
246
247
248
249
/*! \brief Resize the allocated memory
 *
 * Resize the allocated memory, if request is smaller than the allocated memory
 * is not resized
 *
 * \param sz size
 * \return true if the resize operation complete correctly
 *
 */

incardon's avatar
incardon committed
250
251
bool CudaMemory::resize(size_t sz)
{
incardon's avatar
incardon committed
252
	// if the allocated memory is enough, do not resize
incardon's avatar
incardon committed
253
	if (sz <= CudaMemory::size())
incardon's avatar
incardon committed
254
	{return true;}
incardon's avatar
incardon committed
255

256
257
	//! Allocate the device memory if not done yet

incardon's avatar
incardon committed
258
259
	if (CudaMemory::size() == 0)
	{return allocate(sz);}
260

261
	//! Create a new buffer, if sz is bigger than the actual size
incardon's avatar
incardon committed
262
	void * thm = NULL;
incardon's avatar
incardon committed
263

264
	//! Create a new buffer, if sz is bigger than the actual size
incardon's avatar
incardon committed
265
	void * tdm = NULL;
266

incardon's avatar
incardon committed
267
268
269
	if (dm != NULL)
	{
		if (this->sz < sz)
incardon's avatar
incardon committed
270
		{
incardon's avatar
incardon committed
271
			#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
272
			CUDA_SAFE_CALL(cudaMalloc(&tdm,sz));
incardon's avatar
incardon committed
273
274
			#else
			tdm = new unsigned char [sz];
incardon's avatar
incardon committed
275
276
277
			#ifdef GARBAGE_INJECTOR
			memset(tdm,0xFF,sz);
			#endif
incardon's avatar
incardon committed
278
			#endif
incardon's avatar
incardon committed
279

incardon's avatar
incardon committed
280
#ifdef FILL_CUDA_MEMORY_WITH_MINUS_ONE
incardon's avatar
incardon committed
281
			#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
282
			CUDA_SAFE_CALL(cudaMemset(tdm,-1,sz));
incardon's avatar
incardon committed
283
284
285
			#else
			memset(tdm,-1,sz);
			#endif
incardon's avatar
incardon committed
286
287
288
#endif
		}

incardon's avatar
incardon committed
289
		//! copy from the old buffer to the new one
incardon's avatar
incardon committed
290
		#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
291
		CUDA_SAFE_CALL(cudaMemcpy(tdm,dm,CudaMemory::size(),cudaMemcpyDeviceToDevice));
incardon's avatar
incardon committed
292
293
294
		#else
		memcpy(tdm,dm,CudaMemory::size());
		#endif
incardon's avatar
incardon committed
295
	}
296

incardon's avatar
incardon committed
297
298
299
	if (hm != NULL)
	{
		if (this->sz < sz)
incardon's avatar
incardon committed
300
301
		{
			#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
302
			CUDA_SAFE_CALL(cudaHostAlloc(&thm,sz,cudaHostAllocMapped));
incardon's avatar
incardon committed
303
304
			#else
			thm = new unsigned char [sz];
incardon's avatar
incardon committed
305
306
307
			#ifdef GARBAGE_INJECTOR
			memset(thm,0xFF,sz);
			#endif
incardon's avatar
incardon committed
308
309
			#endif
		}
incardon's avatar
incardon committed
310
311

		//! copy from the old buffer to the new one
incardon's avatar
incardon committed
312
		#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
313
		CUDA_SAFE_CALL(cudaMemcpy(thm,hm,CudaMemory::size(),cudaMemcpyHostToHost));
incardon's avatar
incardon committed
314
315
316
		#else
		memcpy(thm,hm,CudaMemory::size());
		#endif
incardon's avatar
incardon committed
317
	}
318
319
320
321
322

	//! free the old buffer

	destroy();

incardon's avatar
incardon committed
323
324
325
	dm = tdm;
	hm = thm;

326
327
328
329
330
331
332
333
334
	//! change to the new buffer

	this->sz = sz;

	return true;
}

/*! \brief Return a readable pointer with your data
 *
335
 * \return a readable pointer with your data
336
337
338
339
340
 *
 */

void * CudaMemory::getPointer()
{
341
	// allocate an host memory if not allocated
incardon's avatar
incardon committed
342
343
344
	if (hm == NULL)
		allocate_host(sz);

345
	return hm;
incardon's avatar
incardon committed
346
}
Pietro Incardona's avatar
Pietro Incardona committed
347
348
349

/*! \brief Return a readable pointer with your data
 *
350
 * \return a readable pointer with your data
Pietro Incardona's avatar
Pietro Incardona committed
351
352
353
 *
 */

354
void CudaMemory::deviceToHost()
Pietro Incardona's avatar
Pietro Incardona committed
355
{
356
	// allocate an host memory if not allocated
Pietro Incardona's avatar
Pietro Incardona committed
357
358
359
	if (hm == NULL)
		allocate_host(sz);

360
	//! copy from device to host memory
incardon's avatar
incardon committed
361
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
362
	CUDA_SAFE_CALL(cudaMemcpy(hm,dm,sz,cudaMemcpyDeviceToHost));
incardon's avatar
incardon committed
363
364
365
	#else
	memcpy(hm,dm,sz);
	#endif
366
367
}

incardon's avatar
incardon committed
368
369
370
371
372
373
374
375
376
377
378
379
380
381
382
/*! \brief It transfer to device memory from the host of another memory
 *
 * \param mem the other memory object
 *
 */
void CudaMemory::deviceToHost(CudaMemory & mem)
{
	// allocate an host memory if not allocated
	if (mem.hm == NULL)
		mem.allocate_host(sz);

	if (mem.sz > sz)
	{resize(mem.sz);}

	//! copy from device to host memory
incardon's avatar
incardon committed
383
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
384
	CUDA_SAFE_CALL(cudaMemcpy(mem.hm,dm,mem.sz,cudaMemcpyDeviceToHost));
incardon's avatar
incardon committed
385
386
387
	#else
	memcpy(mem.hm,dm,mem.sz);
	#endif
incardon's avatar
incardon committed
388
389
}

incardon's avatar
incardon committed
390
391
392
393
394
395
396
397
398
399
400
401
402
403
404
/*! \brief It transfer to device memory from the host of another memory
 *
 * \param mem the other memory object
 *
 */
void CudaMemory::hostToDevice(CudaMemory & mem)
{
	// allocate an host memory if not allocated
	if (mem.hm == NULL)
		mem.allocate_host(sz);

	if (mem.sz > sz)
	{resize(mem.sz);}

	//! copy from device to host memory
incardon's avatar
incardon committed
405
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
406
	CUDA_SAFE_CALL(cudaMemcpy(dm,mem.hm,mem.sz,cudaMemcpyHostToDevice));
incardon's avatar
incardon committed
407
408
409
	#else
	memcpy(dm,mem.hm,mem.sz);
	#endif
incardon's avatar
incardon committed
410
411
412
413
414
415
416
417
418
}

void CudaMemory::hostToDevice(size_t start, size_t stop)
{
	// allocate an host memory if not allocated
	if (hm == NULL)
		allocate_host(sz);

	//! copy from device to host memory
incardon's avatar
incardon committed
419
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
420
	CUDA_SAFE_CALL(cudaMemcpy(((unsigned char *)dm)+start,((unsigned char *)hm)+start,(stop-start),cudaMemcpyHostToDevice));
incardon's avatar
incardon committed
421
422
423
	#else
	memcpy(((unsigned char *)dm)+start,((unsigned char *)hm)+start,(stop-start));
	#endif
incardon's avatar
incardon committed
424
425
}

incardon's avatar
incardon committed
426
427
428
429
430
431
432
433
434
435
436
437
/*! \brief Return a readable pointer with your data
 *
 * \return a readable pointer with your data
 *
 */
void CudaMemory::deviceToHost(size_t start, size_t stop)
{
	// allocate an host memory if not allocated
	if (hm == NULL)
		allocate_host(sz);

	//! copy from device to host memory
incardon's avatar
incardon committed
438
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
439
	CUDA_SAFE_CALL(cudaMemcpy(((unsigned char *)hm)+start,((unsigned char *)dm)+start,(stop-start),cudaMemcpyDeviceToHost));
incardon's avatar
incardon committed
440
441
442
	#else
	memcpy(((unsigned char *)hm)+start,((unsigned char *)dm)+start,(stop-start));
	#endif
incardon's avatar
incardon committed
443
444
445
446
}



447
448
449
450
451
452
453
454
455
456
457
/*! \brief Return a readable pointer with your data
 *
 * \return a readable pointer with your data
 *
 */

const void * CudaMemory::getPointer() const
{
	// allocate an host memory if not allocated
	if (hm == NULL)
		allocate_host(sz);
Pietro Incardona's avatar
Pietro Incardona committed
458
459
460

	return hm;
}
461

incardon's avatar
incardon committed
462
463
464
465
466
467
/*! \brief fill host and device memory with the selected byte
 *
 *
 */
void CudaMemory::fill(unsigned char c)
{
incardon's avatar
incardon committed
468
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
incardon's avatar
incardon committed
469
	CUDA_SAFE_CALL(cudaMemset(dm,c,size()));
incardon's avatar
incardon committed
470
471
472
	#else
	memset(dm,c,size());
	#endif
incardon's avatar
incardon committed
473
474
	if (hm != NULL)
	{memset(hm,c,size());}
incardon's avatar
incardon committed
475
476
}

477
478
479
480
481
482
/*! \brief Return the CUDA device pointer
 *
 * \return CUDA device pointer
 *
 */
void * CudaMemory::getDevicePointer()
incardon's avatar
incardon committed
483
484
485
486
487
488
489
490
491
492
493
{
	return dm;
}

/*! \brief Return a readable pointer with your data
 *
 * \return a readable pointer with your data
 *
 */

void CudaMemory::hostToDevice()
494
495
496
497
498
499
{
	// allocate an host memory if not allocated
	if (hm == NULL)
		allocate_host(sz);

	//! copy from device to host memory
incardon's avatar
incardon committed
500
	#if defined(CUDA_GPU) && !defined(CUDA_ON_CPU)
501
	CUDA_SAFE_CALL(cudaMemcpy(dm,hm,sz,cudaMemcpyHostToDevice));
incardon's avatar
incardon committed
502
503
504
	#else
	memcpy(dm,hm,sz);
	#endif
505
506
507
}


incardon's avatar
incardon committed
508
509
510
511
512
513
514
515
516
/*! \brief Swap the memory
 *
 * \param mem memory to swap
 *
 */
void CudaMemory::swap(CudaMemory & mem)
{
	size_t sz_tmp;
	void * dm_tmp;
incardon's avatar
incardon committed
517
//	long int ref_cnt_tmp;
incardon's avatar
incardon committed
518
519
520
521
522
523
524
	bool is_hm_sync_tmp;
	void * hm_tmp;

	hm_tmp = hm;
	is_hm_sync_tmp = is_hm_sync;
	sz_tmp = sz;
	dm_tmp = dm;
incardon's avatar
incardon committed
525
//	ref_cnt_tmp = ref_cnt;
incardon's avatar
incardon committed
526
527
528
529
530
531
532
533
534
535
536

	hm = mem.hm;
	is_hm_sync = mem.is_hm_sync;
	sz = mem.sz;
	dm = mem.dm;
	ref_cnt = mem.ref_cnt;

	mem.hm = hm_tmp;
	mem.is_hm_sync = is_hm_sync_tmp;
	mem.sz = sz_tmp;
	mem.dm = dm_tmp;
incardon's avatar
incardon committed
537
//	mem.ref_cnt = ref_cnt_tmp;
incardon's avatar
incardon committed
538
}