CUDA 4.0 RC — много потоков хоста на один GPU — поведение cudaStreamQuery и cudaStreamSynchronize

Я написал код, который использует множество потоков хоста (OpenMP) на один GPU. Каждый поток имеет свой собственный поток CUDA для упорядочения запросов. Это очень похоже на приведенный ниже код:

#pragma omp parallel for num_threads(STREAM_NUMBER)
for (int sid = 0; sid < STREAM_NUMBER; sid++) {
    cudaStream_t stream;
    cudaStreamCreate(&stream);

    while (hasJob()) {

        //... code to prepare job - dData, hData, dataSize etc

        cudaError_t streamStatus = cudaStreamQuery(stream);
        if (streamStatus == cudaSuccess) {
             cudaMemcpyAsync(dData, hData, dataSize, cudaMemcpyHostToDevice, stream);
             doTheJob<<<gridDim, blockDim, smSize, stream>>>(dData, dataSize);
        else {
             CUDA_CHECK(streamStatus);
        }
        cudaStreamSynchronize(stream);
    }
    cudaStreamDestroy(stream);
}

И все было хорошо, пока я не получил много мелких работ. В таком случае время от времени cudaStreamQuery возвращает cudaErrorNotReady, что для меня неожиданно, т.к. я использую cudaStreamSynchronize. До сих пор я думал, что cudaStreamQuery всегда будет возвращать cudaSuccess, если он вызывается после cudaStreamSynchronize. К сожалению, оказалось, что cudaStreamSynchronize может завершиться, даже если cudaStreamQuery все еще возвращает cudaErrorNotReady.

Я изменил код на следующий, и все работает правильно.

#pragma omp parallel for num_threads(STREAM_NUMBER)
for (int sid = 0; sid < STREAM_NUMBER; sid++) {
    cudaStream_t stream;
    cudaStreamCreate(&stream);

    while (hasJob()) {

        //... code to prepare job - dData, hData, dataSize etc

        cudaError_t streamStatus;
        while ((streamStatus = cudaStreamQuery(stream)) == cudaErrorNotReady) {
             cudaStreamSynchronize();
        }
        if (streamStatus == cudaSuccess) {
             cudaMemcpyAsync(dData, hData, dataSize, cudaMemcpyHostToDevice, stream);
             doTheJob<<<gridDim, blockDim, smSize, stream>>>(dData, dataSize);
        else {
             CUDA_CHECK(streamStatus);
        }
        cudaStreamSynchronize(stream);
    }
    cudaStreamDestroy(stream);
}

Итак, мой вопрос .... это ошибка или функция?

РЕДАКТИРОВАТЬ: это похоже на JAVA

synchronize {
    while(waitCondition) {
         wait();
    }
}