is it possible to modify or even destroy constant memory in cuda?
sorry for my silly question as beginner.. is it possible to modify or even destroy constant memory in cuda? can i do it with [font=monospace][size=2]cudaDeviceReset ?[/size][/font]
sorry for my silly question as beginner.. is it possible to modify or even destroy constant memory in cuda? can i do it with [font=monospace]cudaDeviceReset ?[/font]

#1
Posted 05/05/2012 08:10 AM   
You can modify contents of constant memory using [url="http://developer.download.nvidia.com/compute/cuda/4_2/rel/toolkit/docs/online/group__CUDART__MEMORY_gf268fa2004636b6926fdcd3189152a14.html#gf268fa2004636b6926fdcd3189152a14"]cudaMemcpyToSymbol()[/url].
You can modify contents of constant memory using cudaMemcpyToSymbol().

Always check return codes of CUDA calls for errors. Do not use __syncthreads() in conditional code unless the condition is guaranteed to evaluate identically for all threads of each block. Run your program under cuda-memcheck to detect stray memory accesses. If your kernel dies for larger problem sizes, it might exceed the runtime limit and trigger the watchdog timer.

#2
Posted 05/08/2012 07:32 AM   
[quote name='tera' date='08 May 2012 - 04:32 PM' timestamp='1336462355' post='1405544']
You can modify contents of constant memory using [url="http://developer.download.nvidia.com/compute/cuda/4_2/rel/toolkit/docs/online/group__CUDART__MEMORY_gf268fa2004636b6926fdcd3189152a14.html#gf268fa2004636b6926fdcd3189152a14"]cudaMemcpyToSymbol()[/url].
[/quote]

i still dont get it.. if i have already copy the memory from host to device
is it possible if I set it to be zero memory again and then use it later for another different kernel?
[quote name='tera' date='08 May 2012 - 04:32 PM' timestamp='1336462355' post='1405544']

You can modify contents of constant memory using cudaMemcpyToSymbol().





i still dont get it.. if i have already copy the memory from host to device

is it possible if I set it to be zero memory again and then use it later for another different kernel?

#3
Posted 06/28/2012 06:40 AM   
"The global, constant, and texture memory spaces are persistent across kernel launches by the same application." (Programming guide)

This is: your host code can do whatever it likes with the constant memory (i.e. modifying it), but you can not do any dynamic allocation/freeing.

example:

alloc constant mem
memcpytosymbol //fill constant with some values
launch kernel
memcpytosymbol //fill constant with other values
launch another kernel
"The global, constant, and texture memory spaces are persistent across kernel launches by the same application." (Programming guide)



This is: your host code can do whatever it likes with the constant memory (i.e. modifying it), but you can not do any dynamic allocation/freeing.



example:



alloc constant mem

memcpytosymbol //fill constant with some values

launch kernel

memcpytosymbol //fill constant with other values

launch another kernel

#4
Posted 06/28/2012 07:17 AM   
[quote name='MKasper' date='28 June 2012 - 04:17 PM' timestamp='1340867831' post='1427513']
"The global, constant, and texture memory spaces are persistent across kernel launches by the same application." (Programming guide)

This is: your host code can do whatever it likes with the constant memory (i.e. modifying it), but you can not do any dynamic allocation/freeing.

example:

alloc constant mem
memcpytosymbol //fill constant with some values
launch kernel
memcpytosymbol //fill constant with other values
launch another kernel
[/quote]

so it means that the memory will be the all the time either i use it or not, doesnt it?
thanks for enlightenment


[quote name='MKasper' date='28 June 2012 - 04:17 PM' timestamp='1340867831' post='1427513']

"The global, constant, and texture memory spaces are persistent across kernel launches by the same application." (Programming guide)



This is: your host code can do whatever it likes with the constant memory (i.e. modifying it), but you can not do any dynamic allocation/freeing.



example:



alloc constant mem

memcpytosymbol //fill constant with some values

launch kernel

memcpytosymbol //fill constant with other values

launch another kernel





so it means that the memory will be the all the time either i use it or not, doesnt it?

thanks for enlightenment




#5
Posted 06/28/2012 07:26 AM   
[quote name='xnov' date='28 June 2012 - 09:26 AM' timestamp='1340868377' post='1427515']
so it means that the memory will be the all the time either i use it or not, doesnt it?
thanks for enlightenment
[/quote]

Yes the constant memory is always 64 KB. All reads from constant memory are cached.
[quote name='xnov' date='28 June 2012 - 09:26 AM' timestamp='1340868377' post='1427515']

so it means that the memory will be the all the time either i use it or not, doesnt it?

thanks for enlightenment





Yes the constant memory is always 64 KB. All reads from constant memory are cached.

#6
Posted 06/28/2012 09:04 AM   
[quote name='wanderine' date='28 June 2012 - 11:04 AM' timestamp='1340874265' post='1427536']
Yes the constant memory is always 64 KB. All reads from constant memory are cached.
[/quote]

But constant memory is local to each .cu module. So if your application consists of several kernels distributed over multiple .cu files, each one has its own 64 KB.

When doing a kernel call the CUDA API automatically takes the necessary steps to make that module's 64kb page of memory cached.

While multiple .cu modules may share other global device variables through extern declarations, this is not possible with constant memory.

Christian
[quote name='wanderine' date='28 June 2012 - 11:04 AM' timestamp='1340874265' post='1427536']

Yes the constant memory is always 64 KB. All reads from constant memory are cached.





But constant memory is local to each .cu module. So if your application consists of several kernels distributed over multiple .cu files, each one has its own 64 KB.



When doing a kernel call the CUDA API automatically takes the necessary steps to make that module's 64kb page of memory cached.



While multiple .cu modules may share other global device variables through extern declarations, this is not possible with constant memory.



Christian

#7
Posted 06/28/2012 09:27 AM   
Scroll To Top