在OpenCL內核中可以讀寫競爭條件是否導致數據損壞？

考慮一對讀取和寫入相同內存位置的OpenCL內核。作爲一個簡單的例子，請考慮以下的OpenCL程序：在OpenCL內核中可以讀寫競爭條件是否導致數據損壞？

__kernel void k1(__global int * a) 
{ 
    a[0] = 2*a[1]; 
} 

__kernel void k2(__global int * a) 
{ 
    a[1] = a[0]-1; 
}

如果許多線程啓動，運行諸多每個內核中，所產生的全局存儲器的狀態是不確定性。

這仍然可能允許編寫接受任何可能的內核操作順序的異步算法。

但是，這要求對全局GPU內存的讀寫是原子。

我的問題是

是這個保證是真的在任何當前GPGPU的硬件？
如果這被OpenCL標準視爲未定義行爲？如果是這樣，那麼常見的實現（特別是CUDA工具包中包含的）會做什麼？
如何測試這種擔憂？

來源

2015-03-18 Patrick Sanan

爲什麼你有兩個同名的內核？無論如何，內存讀取不是原子的。 – 2015-03-19 12:47:50

謝謝，錯字固定。你能更精確嗎？「內存讀取不是原子」是什麼意思？ OpenCL標準不保證它們是原子的嗎？他們在任何現有的硬件上都不是原子的？你能指出某種參考嗎？ – 2015-03-19 14:19:22

如果你排隊你的內核命令到作爲一個有序隊列創建一個單一的命令隊列（即沒有指定CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE創建時），那麼只有一個內核命令將執行一次。這意味着在不同的內核實例之間不會有任何這樣的問題（儘管如果他們正在訪問相同的內存位置，那麼在單個內核實例中的工作項之間仍可能存在競爭條件）。

如果您正在使用亂序隊列或多個命令隊列，那麼您可能確實存在爭用條件。不能保證你的load-modify-store序列是一個原子操作，這會導致未定義的行爲。

根據你實際想要處理的內核，你可以使用OpenCL的內置原子函數，它允許你在原子中執行一組特定的讀 - 修改 - 寫操作方式。

來源

2015-03-18 21:33:15 jprice

我的問題不是load-modify-store序列是原子，而是讀寫自己是原子。 – 2015-04-14 13:23:32

在OpenCL內核中可以讀寫競爭條件是否導致數據損壞？

回答

相關問題