Creazione del contesto cuda e associazione delle risorse nelle applicazioni API di runtime

Question

Desidero capire come viene creato un contesto cuda e associato a un kernel nelle applicazioni API di runtime cuda?

So che è fatto sotto il cofano dalle API del driver. Ma mi piacerebbe capire la cronologia della creazione.

Per cominciare so che cudaRegisterFatBinary è la prima chiamata di cuda api e registra un file fatbin con il runtime. È seguito da una manciata di API di registrazione della funzione cuda che chiamano cuModuleLoad nel livello driver. Ma se l'applicazione API di runtime Cuda invoca cudaMalloc, come viene fornito il puntatore a questa funzione associata al contesto, che credo avrebbe dovuto essere stato creato in precedenza. Come si ottiene un handle per questo contesto già creato e si associano le future chiamate API di runtime con esso? Per favore demistificare i meccanismi interni

Per citare la documentazione di NVIDIA su questo

CUDA API runtime chiamate operano sul CUDA CUcontext driver API che è legato al thread host corrente.

Se non esiste CUDA driver API CUcontext vincolato alla corrente filo al momento di una chiamata API CUDA runtime che richiede un CUcontext allora il CUDA runtime implicitamente creare un nuovo CUcontext prima di eseguire la chiamata.

Se il CUDA runtime crea un CUcontext allora il CUcontext sarà creato utilizzando i parametri specificati dalle API CUDA runtime funzioni cudaSetDevice, cudaSetValidDevices, cudaSetDeviceFlags, cudaGLSetGLDevice, cudaD3D9SetDirect3DDevice, cudaD3D10SetDirect3DDevice e cudaD3D11SetDirect3DDevice. Notare che queste funzioni avranno esito negativo con cudaErrorSetOnActiveProcess se sono chiamate quando un CUcontext è associato al thread host corrente.

La durata di un CUcontext è gestita da un meccanismo di conteggio riferimento . Il conteggio dei riferimenti di un CUcontext è inizialmente impostato su 0, e viene incrementato da cuCtxAttach e decrementato da cuCtxDetach.

Se un CUcontext viene creato dal CUDA runtime, quindi il runtime CUDA sarà diminuire il conteggio di riferimento di tale CUcontext nella funzione cudaThreadExit. Se un CUcontext viene creato dall'API driver CUDA (o viene creato da un'istanza separata della libreria API di Runtime CUDA), quindi il Runtime CUDA non incrementerà o decrementerà il conteggio di riferimento di CUcontext.

Tutto lo stato dell'API di runtime CUDA (ad esempio, indirizzi di variabili globali e valori ) viaggia con il relativo CUcontext sottostante. In particolare, se un CUcontext viene spostato da un thread a un altro (utilizzando cuCtxPopCurrent e cuCtxPushCurrent), allora tutto lo stato dell'API di runtime di CUDA si sposterà a anche tale thread.

Ma quello che non capisco è come cuda runtime crea il contesto? quali chiamate API vengono utilizzate per questo? Il compilatore nvcc inserisce alcune chiamate API per farlo in fase di compilazione o è fatto interamente in fase di runtime? Se il primo è vero, quali API vengono utilizzate per questa gestione del contesto? Più tardi è vero come è fatto esattamente?

Se un contesto è associato a un thread host, come possiamo accedere a questo contesto? È automaticamente associato a tutte le variabili e ai riferimenti ai puntatori trattati dalla discussione?

in che modo un caricamento del modulo viene eseguito nel contesto?

Answer 1

Il runtime CUDA mantiene un elenco globale di moduli da caricare e aggiunge a tale elenco ogni volta che una DLL o un .so che utilizza il runtime CUDA viene caricato nel processo. Ma i moduli non vengono effettivamente caricati fino a quando non viene creato un dispositivo.

La creazione del contesto e l'inizializzazione vengono eseguite "pigramente" dal runtime CUDA - ogni volta che si chiama una funzione come cudaMemcpy(), si verifica se CUDA è stato inizializzato e, in caso contrario, crea un contesto (sul dispositivo precedentemente specificato da cudaSetDevice(), o il dispositivo predefinito se cudaSetDevice() non è mai stato chiamato) e carica tutti i moduli. Il contesto è associato a quel thread della CPU da quel momento in poi, finché non viene modificato da cudaSetDevice().

È possibile utilizzare le funzioni di gestione del contesto/thread dall'API del driver, ad esempio cuCtxPopCurrent()/cuCtxPushCurrent(), per utilizzare il contesto da un thread diverso.

È possibile chiamare cudaFree (0); per forzare questa inizializzazione pigra a verificarsi.

Consiglio vivamente di farlo al momento dell'inizializzazione dell'applicazione, per evitare condizioni di gara e comportamento indefinito. Vai avanti e enumerare e inizializzare i dispositivi il prima possibile nella tua app; una volta fatto, in CUDA 4.0 è possibile chiamare cudaSetDevice() da qualsiasi thread della CPU e selezionerà il contesto corrispondente creato dal codice di inizializzazione.