2012-06-22 114 views
1

我是CUDA编程的新手,并且由于性能原因,我主要使用每块的共享内存。我的程序现在的结构是这样的,我使用一个内核来加载共享内存和另一个内核来读取预加载的共享内存。但据我所知,共享内存不能在两个不同的内核之间持续存在。持久GPU共享内存

我有两个解决方案;我不确定第一个,第二个可能会很慢。

第一个解决方案:我使用的是一个内核,而不是使用两个内核。加载共享内存后,内核可能会等待来自主机的输入,然后执行该操作,然后将该值返回给主机。我不确定内核是否可以等待来自主机的信号。

第二种解决方案:加载共享内存后,将共享内存值复制到全局内存中。当下一个内核启动时,将全局内存中的值复制回共享内存,然后执行操作。

请评论两种解决方案的可行性。

+0

为什么启动内核后需要等待主机输入? – pQB

+2

这听起来像你可能试图将共享内存压缩到一个并非真正意义上的用法。如果您的目标是计算能力> = 2.0的设备,我会说,首先不要担心使用共享内存。相反,使用'cudaFuncSetCacheConfig(MyKernel,cudaFuncCachePreferL1)'设置L1缓存的首选项。当您的算法运行时,在CUDA分析器中运行您的应用程序并检查它是否受内存限制。如果是内存绑定,那么考虑是否有任何方法可以通过使用共享内存或更有效地使用L1来提高性能。 –

回答

2

我会使用你提出的第一个解决方案的一个变种:正如你已经怀疑,你不能等待内核中的主机输入 - 但你可以在一个点同步你的内核。只需调用“__syncthreads();”在你的数据加载到共享内存之后在你的内核中。

我不明白你的第二个解决方案:你为什么要将数据复制到共享内存中,只是将它复制回第一个内核中的全局内存?或者这第一个内核是否也计算了一些内容在这种情况下,我想它不会帮助将初步结果存储在共享内存中,我宁愿将它们直接存储在全局内存中(但这可能取决于算法)。