TL; DR: на основании предоставленного вами кода в обоих случаях вашего конкретного использования кажется, что cudaSetDevice() заменяет контекст наверху стека.
Давайте изменим ваш код немного, а затем посмотрим, что мы можем сделать о влиянии каждого вызова API в вашем коде на стек контекста:
$ cat t1759.cu
#include <cuda.h>
#include <cuda_runtime_api.h>
#include <cassert>
#include <iostream>
void check(int j, CUcontext ctx1, CUcontext ctx2){
CUcontext ctx0;
int i = 0;
while (true) {
auto status = cuCtxPopCurrent(&ctx0);
if (status != CUDA_SUCCESS) { break; }
if (ctx0 == ctx1) std::cout << j << ":Next context on stack (" << i++ << ") is ctx1:" << (void*) ctx0 << '\n';
else if (ctx0 == ctx2) std::cout << j << ":Next context on stack (" << i++ << ") is ctx2:" << (void*) ctx0 << '\n';
else std::cout << j << ":Next context on stack (" << i++ << ") is unknown:" << (void*) ctx0 << '\n';
}
}
void runtest(int i)
{
CUcontext ctx1, primary = NULL;
cuInit(0);
auto dstatus = cuCtxCreate(&ctx1, 0, 0); // checkpoint 1
assert (dstatus == CUDA_SUCCESS);
if (i == 1) {check(i,ctx1,primary); return;}// checkpoint 1
dstatus = cuCtxPushCurrent(ctx1); // checkpoint 2
assert (dstatus == CUDA_SUCCESS);
if (i == 2) {check(i,ctx1,primary); return;}// checkpoint 2
auto rstatus = cudaSetDevice(0); // checkpoint 3
assert (rstatus == cudaSuccess);
if (i == 3) {check(i,ctx1,primary); return;}// checkpoint 3
void* ptr1;
void* ptr2;
rstatus = cudaMalloc(&ptr1, 1024); // checkpoint 4
assert (rstatus == cudaSuccess);
if (i == 4) {check(i,ctx1,primary); return;}// checkpoint 4
dstatus = cuCtxGetCurrent(&primary); // checkpoint 5
assert (dstatus == CUDA_SUCCESS);
assert(primary != ctx1);
if (i == 5) {check(i,ctx1,primary); return;}// checkpoint 5
dstatus = cuCtxPushCurrent(ctx1); // checkpoint 6
assert (dstatus == CUDA_SUCCESS);
if (i == 6) {check(i,ctx1,primary); return;}// checkpoint 6
rstatus = cudaMalloc(&ptr2, 1024); // checkpoint 7
assert (rstatus == cudaSuccess);
if (i == 7) {check(i,ctx1,primary); return;}// checkpoint 7
rstatus = cudaSetDevice(0); // checkpoint 8
assert (rstatus == cudaSuccess);
if (i == 8) {check(i,ctx1,primary); return;}// checkpoint 8
return;
}
int main(){
for (int i = 1; i < 9; i++){
cudaDeviceReset();
runtest(i);}
}
$ nvcc -o t1759 t1759.cu -lcuda -std=c++11
$ ./t1759
1:Next context on stack (0) is ctx1:0x11087e0
2:Next context on stack (0) is ctx1:0x1741160
2:Next context on stack (1) is ctx1:0x1741160
3:Next context on stack (0) is unknown:0x10dc520
3:Next context on stack (1) is ctx1:0x1c5aa70
4:Next context on stack (0) is unknown:0x10dc520
4:Next context on stack (1) is ctx1:0x23eaa00
5:Next context on stack (0) is ctx2:0x10dc520
5:Next context on stack (1) is ctx1:0x32caf30
6:Next context on stack (0) is ctx1:0x3a44ed0
6:Next context on stack (1) is ctx2:0x10dc520
6:Next context on stack (2) is ctx1:0x3a44ed0
7:Next context on stack (0) is ctx1:0x41cfd90
7:Next context on stack (1) is ctx2:0x10dc520
7:Next context on stack (2) is ctx1:0x41cfd90
8:Next context on stack (0) is ctx2:0x10dc520
8:Next context on stack (1) is ctx2:0x10dc520
8:Next context on stack (2) is ctx1:0x4959c70
$
Исходя из вышеизложенного, по мере того, как мы выполняем каждый вызов API в вашем коде:
1.
auto dstatus = cuCtxCreate(&ctx1, 0, 0); // checkpoint 1
1:Next context on stack (0) is ctx1:0x11087e0
Создание контекста также помещает вновь созданный контекст в стек, как упоминалось здесь .
2.
dstatus = cuCtxPushCurrent(ctx1); // checkpoint 2
2:Next context on stack (0) is ctx1:0x1741160
2:Next context on stack (1) is ctx1:0x1741160
Неудивительно, что нажатие того же контекста в стеке создает для него другую запись стека.
3.
auto rstatus = cudaSetDevice(0); // checkpoint 3
3:Next context on stack (0) is unknown:0x10dc520
3:Next context on stack (1) is ctx1:0x1c5aa70
Вызов cudaSetDevice() имеет заменил вершину стека «неизвестным» контекстом. (На данный момент неизвестно, потому что мы не получили значение дескриптора «другого» контекста).
4.
rstatus = cudaMalloc(&ptr1, 1024); // checkpoint 4
4:Next context on stack (0) is unknown:0x10dc520
4:Next context on stack (1) is ctx1:0x23eaa00
Никакой разницы в конфигурации стека из-за этого вызова.
5.
dstatus = cuCtxGetCurrent(&primary); // checkpoint 5
5:Next context on stack (0) is ctx2:0x10dc520
5:Next context on stack (1) is ctx1:0x32caf30
Нет разницы в конфигурации стека из-за этого вызова, но теперь мы знаем, что верхняя часть контекста стека является текущим контекстом (и мы можем предположить, что это первичный контекст. ).
6.
dstatus = cuCtxPushCurrent(ctx1); // checkpoint 6
6:Next context on stack (0) is ctx1:0x3a44ed0
6:Next context on stack (1) is ctx2:0x10dc520
6:Next context on stack (2) is ctx1:0x3a44ed0
Здесь нет ничего удивительного. Мы помещаем ctx1 в стек, и поэтому стек имеет 3 записи, первая из которых представляет собой контекст, созданный API драйвера, а следующие две записи совпадают с конфигурацией стека из шага 5, просто перемещены на одно место в стеке. .
7.
rstatus = cudaMalloc(&ptr2, 1024); // checkpoint 7
7:Next context on stack (0) is ctx1:0x41cfd90
7:Next context on stack (1) is ctx2:0x10dc520
7:Next context on stack (2) is ctx1:0x41cfd90
Опять же, этот вызов не влияет на конфигурацию стека.
8.
rstatus = cudaSetDevice(0); // checkpoint 8
8:Next context on stack (0) is ctx2:0x10dc520
8:Next context on stack (1) is ctx2:0x10dc520
8:Next context on stack (2) is ctx1:0x4959c70
И снова мы видите, что поведение здесь таково, что при вызове cudaSetDevice() заменяется верхний контекст стека основным контекстом.
Вывод, который я сделал из вашего тестового кода, заключается в том, что я вижу нет несогласованность в поведении вызова cudaSetDevice() при смешивании с различными вызовами API среды выполнения и драйверов, как в вашем коде.
С моей точки зрения, такая парадигма программирования - безумие. Я не могу представить, зачем вам смешивать таким образом API драйвера и код API среды выполнения.