天天看點

CUDA實踐指南(八)

優化CUDA應用程式:

在每輪應用程式并行化完成後,開發人員可以轉向優化實施以提高性能。 由于可以考慮許多可能的優化,對應用程式的需求有充分的了解可以幫助盡可能平滑地實作流程。 但是,與APOD整體一樣,程式優化是一個疊代過程(找出優化的機會,應用和測試優化,驗證實作的加速并重複),這意味着程式員不需要花費大量 在看到很好的加速之前記住大部分可能的優化政策。 相反,政策可以在學習時逐漸應用。

優化可以應用于各個層面,從重疊的資料傳輸和計算一直到精細調整浮點運算序列。 可用的分析工具對于指導此過程非常有用,因為它們可以幫助為開發人員的優化工作提供次佳的操作方法。

性能名額:

在嘗試優化CUDA代碼時,需要知道如何準确衡量性能并了解帶寬在性能測量中的作用。 本章讨論如何使用CPU定時器和CUDA事件正确測量性能。 然後探讨帶寬如何影響性能名額以及如何減輕它帶來的一些挑戰。

Timing:

CUDA調用和核心執行可以使用CPU或GPU定時器來定時。

用CPU計時器:

任何CPU計時器都可以用來測量CUDA調用或核心執行所用的時間。 各種CPU時序方法的細節超出了本文檔的範圍,但開發人員應始終注意其定時調用提供的分辨率。

在使用CPU定時器時,記住許多CUDA API函數是異步的,這一點很重要。 也就是說,他們在完成工作之前将控制權傳回給調用CPU線程。 所有核心啟動都是異步的,因為記憶體複制函數的名稱上帶有Async字尾。 是以,要準确測量某個特定調用或CUDA調用序列的運作時間,需要在啟動和停止CPU計時器之前立即調用cudaDeviceSynchronize()來同步CPU線程與GPU。 cudaDeviceSynchronize()阻塞調用CPU線程,直到線程先前發出的所有CUDA調用都完成為止。

盡管也可以使CPU線程與GPU上的特定流或事件同步,但這些同步功能不适用于除預設流以外的流中的計時代碼。 cudaStreamSynchronize()阻塞CPU線程,直到先前發送到給定流中的所有CUDA調用都完成為止。 cudaEventSynchronize()阻塞,直到GPU記錄特定流中的給定事件。 由于驅動程式可能會交錯執行來自其他非預設流的CUDA調用,是以其他流中的調用可能會包含在該時序中。

由于預設流0在裝置上表現出序列化行為(預設流中的操作隻有在任何流中的所有前面的調用都完成之後才能開始;并且任何流中的後續操作都不能開始,直到完成為止) 這些功能可以在預設流中可靠地用于定時。

CPU到GPU同步點意味着GPU處理流水線中存在停頓,是以應謹慎使用,以盡量減少其對性能的影響。

用GPU計時器:

CUDA事件API提供建立和銷毀事件,記錄事件(通過時間戳)以及将時間戳差異轉換為浮點值(以毫秒為機關)的調用。 如何使用CUDA事件計時代碼說明了它們的使用。

這裡使用cudaEventRecord()将開始和停止事件放入預設流,即流0中。裝置将在事件到達流中時記錄該事件的時間戳。 cudaEventElapsedTime()函數傳回記錄開始和停止事件之間的時間間隔。 該值以毫秒為機關表示并具有大約半微秒的分辨率。 與本清單中的其他調用一樣,它們的具體操作,參數和傳回值在CUDA工具包參考手冊中進行了描述。 請注意,時序是在GPU時鐘上測量的,是以時序分辨率與作業系統無關。

繼續閱讀