cuda内存总结
1.shared memory
__shared__ 聲明為共享內(nèi)存,將會(huì)保存在共享內(nèi)存中
?2.constant memory
__constant__ 聲明為常量?jī)?nèi)存,將會(huì)保存在常量?jī)?nèi)存中,常量?jī)?nèi)存是只讀內(nèi)存,聲明時(shí)要靜態(tài)的分配空間
將數(shù)據(jù)從CPU拷貝到常量?jī)?nèi)存中時(shí)用cudaMemcpyToSymbol,例如cudaMemcpyToSymbol( s, temp_s,sizeof(Sphere) * SPHERES)
常量?jī)?nèi)存帶來(lái)性能提升的原因:
1.對(duì)常量?jī)?nèi)存的單次讀操作可以廣播到臨近線程,將節(jié)約15次讀操作
2.常量?jī)?nèi)存的數(shù)據(jù)將緩存起來(lái),對(duì)相同地址的連續(xù)讀操作將不會(huì)產(chǎn)生額外的內(nèi)存通信量
當(dāng)處理常量?jī)?nèi)存時(shí),NVIDIA硬件將單次內(nèi)存讀操作廣播到每個(gè)半線程束(線程束中線程的一半),如果半線程束中的每個(gè)線程都從常量?jī)?nèi)存的相同地址上讀取數(shù)據(jù),那么使用常量?jī)?nèi)存產(chǎn)生的內(nèi)存流量將會(huì)是使用全局內(nèi)存的1/16。但是當(dāng)所有的線程讀取不同的地址時(shí),會(huì)降低性能,因?yàn)槿舭刖€程束中的16個(gè)線程訪問(wèn)常量?jī)?nèi)存中的不同數(shù)據(jù)時(shí),這16次不同的讀取操作將會(huì)被串行化,從而需要16倍的時(shí)間來(lái)發(fā)出請(qǐng)求,但在全局內(nèi)存中將會(huì)同時(shí)發(fā)出請(qǐng)求。
轉(zhuǎn)載于:https://www.cnblogs.com/shrimp-can/p/5052305.html
總結(jié)
- 上一篇: Oracle初始化参数之memory_t
- 下一篇: 基于verilog贪吃蛇游戏设计