CUDA кубические текстуры карты

Ответ 1

ОК - мне удалось решить проблему самостоятельно, хотя решение не так просто, как использование другой функции CUDA.

Чтобы связать ссылку на текстуру CUDA с любой текстурой, будь она получена из OpenGL или D3D, необходимо предоставить массив CUDA, который сопоставляется с ресурсом, используя cuGraphicsSubResourceGetMappedArray для его получения. Как я упоминал в вопросе, это просто в случае одномерной или двухмерной текстуры. Но с другими доступными типами это сложнее.

В любой момент нам нужен массив CUDA, к которому привязана ссылка. То же самое происходит с текстурой кубической карты. Но в таком случае массив должен быть трехмерным. Проблема в том, что API-интерфейс драйвера CUDA предоставляет только вышеупомянутую функцию для извлечения одного слоя из такого ресурса текстуры и сопоставляет его с одним двумерным массивом. Чтобы получить то, что мы хотим, мы должны сделать 3D-массив, содержащий все слои (или лица в случае карты куба).

Прежде всего, мы должны получить массивы для каждого слоя/лица, используя указанную выше функцию. Следующим шагом будет создание 3D-массива путем вызова cuArray3DCreate, который будет снабжен соответствующим набором параметров (размер/количество слоев, уровень деталь, формат данных, количество каналов на тексель и некоторые флаги). Затем мы должны скопировать массивы слоев в 3D с серией вызовов cuMemcpy3D, по одному для каждого массива уровня/лица.

Наконец, мы установили нашу целевую ссылку на текстуУ CUDA cuTexRefSetArray, которая была загружена с помощью 3D-массива, который мы создали и скопировали. Внутри кода устройства мы создаем ссылку с соответствующим типом и режимом текстуры (float4 и cube map) и примерем texCubemap.

Ниже я помещаю фрагмент функции, который делает все это, доступное по всей длине в CIRT Repository (файл cirt_server.c, функция cirtTexImage3D).

//...
if (result)
{
    // Create a 3D array...
    CUDA_ARRAY3D_DESCRIPTOR layeredTextureDescr;
    layeredTextureDescr.Width = w;
    layeredTextureDescr.Height = h;
    layeredTextureDescr.Depth = d;
    layeredTextureDescr.Format = map_type_to_format(type);
    layeredTextureDescr.NumChannels = format == CIRT_RGB ? CIRT_RGBA : format;
    layeredTextureDescr.Flags = map_target_to_flags(target);

    if (result) result = LogCUDADriverCall(cuArray3DCreate(&hTexRefArray, &layeredTextureDescr),
        FUN_NAME(": cuArray3DCreate_tex3D"), __FILE_LINE__);

    // Copy the acquired layer/face arrays into the collective 3D one...
    CUDA_MEMCPY3D layerCopyDescr;
    layerCopyDescr.srcMemoryType = CU_MEMORYTYPE_ARRAY;
    layerCopyDescr.srcXInBytes = 0;
    layerCopyDescr.srcZ = 0;
    layerCopyDescr.srcY = 0;
    layerCopyDescr.srcLOD = 0;

    layerCopyDescr.dstMemoryType = CU_MEMORYTYPE_ARRAY;
    layerCopyDescr.dstLOD = 0;

    layerCopyDescr.WidthInBytes = layeredTextureDescr.NumChannels * w;
    layerCopyDescr.Height = h;
    layerCopyDescr.Depth = target == CIRT_TEXTURE_CUBE_MAP ? 1 : d;
    layerCopyDescr.dstArray = hTexRefArray;

    for (i = 0; i < num_layers; ++i)
    {
        layer = ((num_layers == 6) ? CU_CUBEMAP_FACE_POSITIVE_X + i : i);
        layerCopyDescr.dstXInBytes = 0;
        layerCopyDescr.dstY = 0;
        layerCopyDescr.dstZ = i;
        layerCopyDescr.srcArray = hLayres[i];

        if (result) result = LogCUDADriverCall(cuMemcpy3D(&layerCopyDescr), 
            FUN_NAME(": cuMemcpy3D_tex3D"), __FILE_LINE__);
    }

    // Finally bind the 3D array with texture reference...
    if (result) LogCUDADriverCall(cuTexRefSetArray(hTexRef, hTexRefArray, CU_TRSA_OVERRIDE_FORMAT),
        FUN_NAME(": cuTexRefSetArray_tex3D"), __FILE_LINE__);

    if (hLayres)
        free(hLayres);

    if (result)
        current->m_oTextureManager.m_cuTextureRes[current->m_oTextureManager.m_nTexCount++] = hTexResource;
}
//...

Я проверил его с картами куба только сейчас, но он должен отлично работать с 3D-текстурой.

Ответ 2

Я не очень хорошо знаком с CUDA, но у меня есть некоторый опыт работы с OpenGL и DirectX, и я также знаком с API-интерфейсами, библиотеками и конвейерами 3D Graphics Rendering и возможностью установки и использования этих API.

Когда я смотрю на ваш вопрос (ы):

Как работать с текстурами карты куба OpenGL в CUDA?

И вы продолжаете объяснять это следующим образом:

Когда вы хотите использовать текстуры OpenGL в ядре CUDA, одна из вещей, которые нужно сделать, - это получить массив CUDA из зарегистрированного изображения и отображаемого ресурса, в данном случае текстуры. В API-интерфейсе драйвера выполняется вызов cuGraphicsSubResourceGetMappedArray, который в случае 2D-текстуры не является проблемой. Но когда речь идет о вышеупомянутой карте куба, третьим параметром этой функции требуется переименование лица (например, CU_CUBEMAP_FACE_POSITIVE_X). Таким образом, возникают некоторые вопросы - когда кто-то передает такое перечисление, то возвращенный массив текстур будет содержать только данные этой конкретной грани, верно? Затем, как использовать текстуру куба в целом, выполнить отображение куба, а также:

color = texCube(cubeMap, x, y, z);

Или это невозможно сделать в ядре CUDA, и нужно использовать 2D-текстуры с правильными вычислениями и выборкой в ​​коде пользователя?

Я пошел на сайт CUDA для их API SDK и программных документов. И нашел эту функцию cuGraphicsSubResourceGetMappedArray()

CUresult cuGraphicsSubResourceGetMappedArray ( CUarray* pArray,
                                               CUgraphicsResource resource, 
                                               unsigned int arrayIndex,
                                               unsigned int mipLevel ) 

Получить массив, с помощью которого можно получить доступ к подресурсу графического ресурса с графикой.

Параметры

• pArray - возвращаемый массив, через который можно получить доступ к ресурсу ресурса
• resource - переназначенный ресурс для доступа
• arrayIndex - Индекс массива для текстур массива или индекс лица куба-карты, как определено CUarray_cubemap_face для текстур кубатуры для субресурса для доступа к
• mipLevel - уровень Mipmap для подресурса для доступа

Возвращает

• CUDA_SUCCESS, CUDA_ERROR_DEINITIALIZED, CUDA_ERROR_NOT_INITIALIZED,
• CUDA_ERROR_INVALID_CONTEXT, CUDA_ERROR_INVALID_VALUE,
• CUDA_ERROR_INVALID_HANDLE, CUDA_ERROR_NOT_MAPPED,
• CUDA_ERROR_NOT_MAPPED_AS_ARRAY

Описание

Возвращает в * pArray массив, через который можно получить доступ к подресурсу ресурса ресурсов графического ресурса, который соответствует массиву indexIndex массива и mipLap уровня mipmap. Значение, установленное в * pArray, может меняться каждый раз при отображении этого ресурса.

Если resource не является texture, тогда к нему нельзя получить доступ через array и CUDA_ERROR_NOT_MAPPED_AS_ARRAY. Если arrayIndex не является допустимым array index для resource, возвращается CUDA_ERROR_INVALID_VALUE. Если mipLevel не является допустимым mipmap level для resource, возвращается CUDA_ERROR_INVALID_VALUE. Если ресурс не mapped, возвращается CUDA_ERROR_NOT_MAPPED.

Примечание: Обратите внимание, что эта функция также может возвращать коды ошибок из предыдущих асинхронных запусков.

См. также:

cuGraphicsResourceGetMappedPointer

Подробнее: http://docs.nvidia.com/cuda/cuda-driver-api/index.html#ixzz4ic22V4DzСледуйте за нами: @GPUCcomputing on Twitter | NVIDIA на Facebook

Этот метод функции был найден в NVidia CUDA DriverAPI, а не в RuntimeAPI. При понимании аппаратного обеспечения с возможностями CUDA существует разница между программируемыми конвейерами Host и Device, которые можно найти здесь: http://docs.nvidia.com/cuda/cuda-c-best-practices-guide/index.html#axzz4ic6tFjXR

2. Гетерогенные вычисления

Программирование CUDA включает запуск кода на двух разных платформах одновременно: хост-система с одним или несколькими процессорами и один или несколько графических процессоров NVIDIA с поддержкой CUDA.

Хотя графические процессоры NVIDIA часто связаны с графикой, они также являются мощными арифметическими механизмами, способными параллельно запускать тысячи легких потоков. Эта возможность делает их хорошо подходящими для вычислений, которые могут использовать параллельное выполнение.

Однако устройство основано на совершенно другом дизайне от хост-системы, и важно понять эти различия и как они определяют производительность приложений CUDA, чтобы эффективно использовать CUDA.

• 2.1. Различия между хостом и устройствомОсновные отличия в модели потоков и в отдельных физических памяти:
  • Ресурсы Threading- Конвейеры выполнения на хост-системах могут поддерживать ограниченное количество одновременных потоков. Серверы с четырьмя шестнадцатеричными процессорами сегодня могут одновременно запускать только 24 потока (или 48, если процессоры поддерживают Hyper-Threading.) Для сравнения, самый маленький исполняемый модуль parallelism на устройстве CUDA содержит 32 потока (называемых основанием потоки). Современные графические процессоры NVIDIA могут поддерживать до 1536 активных потоков одновременно на мультипроцессор (см. "Функции и спецификации Руководства по программированию CUDA C" ). На графических процессорах с 16 многопроцессорными системами это приводит к более чем 24 000 одновременно активных потоков.
  • Темы. Нити на процессоре обычно являются супертяжелыми объектами. Операционная система должна обменивать потоки на каналах выполнения CPU и выключать их, чтобы обеспечить возможность многопоточности. Контекстные коммутаторы (при замене двух потоков) поэтому медленны и дороги. Для сравнения, потоки на графических процессорах чрезвычайно легки. В типичной системе тысячи потоков помещаются в очередь для работы (в искажениях по 32 потока каждый). Если графический процессор должен ждать один переплет нитей, он просто начинает выполнять работу над другим. Поскольку отдельные регистры распределяются по всем активным потокам, при переключении между потоками графического процессора не происходит перекосов регистров или другого состояния. Ресурсы остаются выделенными для каждого потока, пока не завершат его выполнение. Короче говоря, ядра процессора предназначены для минимизации латентности для одного или двух потоков за каждый раз, тогда как графические процессоры предназначены для обработки большого количества параллельных, легких потоков, чтобы максимизировать пропускную способность.
  • ОЗУ. Хост-система и устройство имеют свои собственные прикрепленные физические запоминающие устройства. Поскольку память хоста и устройства разделена шиной PCI Express (PCIe), элементы в памяти хоста иногда должны передаваться по шине в память устройства или наоборот, как описано в разделе "Что работает на устройстве с включенным CUDA"?

Это основные аппаратные различия между хостами CPU и устройствами GPU в отношении параллельного программирования. Другие различия обсуждаются, поскольку они возникают в другом месте этого документа. Приложения, составленные с учетом этих различий, могут рассматривать хост и устройство вместе как сплоченную гетерогенную систему, в которой каждый блок обработки используется для выполнения той работы, которую он лучше всего выполняет: последовательная работа с хостом и параллельная работа на устройстве.

Подробнее: http://docs.nvidia.com/cuda/cuda-c-best-practices-guide/index.html#ixzz4ic8ch2fqСледуйте за нами: @GPUCcomputing on Twitter | NVIDIA на Facebook

Теперь, зная, что для библиотек API CUDA есть два разных API, мы должны понимать разницу между двумя найденными здесь: Разница между драйверами и API-интерфейсами во время выполнения

1. Разница между драйверами и API-интерфейсами времени выполнения

API-интерфейс драйвера и времени выполнения очень похожи и могут использоваться по большей части взаимозаменяемо. Тем не менее, есть несколько ключевых различий, которые стоит отметить между ними.

Сложность и контроль

API среды выполнения упрощает управление кодом устройства, обеспечивая неявную инициализацию, управление контекстом и управление модулем. Это приводит к более простому коду, но также не имеет уровня контроля, который имеет API-интерфейс драйвера.

Для сравнения, API-интерфейс драйвера предлагает более мелкозернистый контроль, особенно в контексте контекста и загрузки модулей. Запуск ядра намного сложнее реализовать, так как параметры выполнения и параметры ядра должны быть указаны с явными вызовами функций. Однако, в отличие от среды выполнения, где все ядра автоматически загружаются во время инициализации и остаются загруженными до тех пор, пока программа запускается, с API-интерфейсом драйвера можно только сохранить модули, которые в настоящее время загружены или даже динамически перезагружать модули. API-интерфейс драйвера также не зависит от языка, поскольку он имеет дело только с кубиновыми объектами.

Управление контекстом

Контекстное управление может быть выполнено с помощью API-интерфейса драйвера, но не отображается в API-интерфейсе. Вместо этого API времени выполнения сам определяет, какой контекст использовать для потока: если контекст стал текущим для вызывающего потока через API-интерфейс драйвера, среда выполнения будет использовать его, но если такого контекста нет, он использует "первичный контекст". Первичные контексты создаются по мере необходимости, по одному на каждое устройство на процесс, подсчитываются по ссылке и затем уничтожаются, когда больше нет ссылок на них. В течение одного процесса все пользователи API среды выполнения будут совместно использовать основной контекст, если только контекст не стал актуальным для каждого потока. Контекст, который использует среда выполнения, то есть текущий контекст или первичный контекст, может быть синхронизирован с cudaDeviceSynchronize() и уничтожен cudaDeviceReset().

Использование API среды выполнения с первичными контекстами имеет свои компромиссы. Это может вызвать проблемы для пользователей, которые пишут плагины для более крупных пакетов программного обеспечения, например, потому что, если все подключаемые модули работают в одном процессе, все они будут использовать контекст, но, скорее всего, не смогут общаться друг с другом. Итак, если один из них называет cudaDeviceReset() после завершения всей своей работы CUDA, другие плагины будут терпеть неудачу, потому что контекст, который они использовали, был уничтожен без их ведома. Чтобы избежать этой проблемы, клиенты CUDA могут использовать API-интерфейс драйвера для создания и установки текущего контекста, а затем использовать API среды выполнения для работы с ним. Однако контексты могут потреблять значительные ресурсы, такие как память устройства, дополнительные потоки хоста и затраты на производительность при переключении контекста на устройство. Это совместное использование контекста во время выполнения важно при использовании API-интерфейса драйвера в сочетании с библиотеками, построенными на API-интерфейсе выполнения, например cuBLAS или cuFFT.

Подробнее: http://docs.nvidia.com/cuda/cuda-driver-api/index.html#ixzz4icCoAXb7Следуйте за нами: @GPUCcomputing on Twitter | NVIDIA на Facebook

Так как это происходит в DriverAPI, он обладает большей гибкостью контроля над программистом, но также требует большей ответственности за управление тем, где библиотека RuntimeAPI делает вещи более автоматическими, но дает вам меньше контроля.

Это видно, поскольку вы упомянули, что работаете со своими Kernels, но из описания их реализации функции

 CUresult cuGraphicsSubResourceGetMappedArray ( CUarray* pArray,
                                                CUgraphicsResource resource, 
                                                unsigned int arrayIndex,
                                                unsigned int mipLevel )

В документации мне сообщается, что первым параметром, который выполняет эта функция, является возвращаемый массив, через который можно получить доступ к ресурсу ресурса. Второй параметр этой функции - сам отображаемый графический ресурс. Третий параметр, в котором я считаю, является параметром, о котором вы говорили, где он является перечисляемым типом лица, и затем вы спрашивали: когда кто-то передает такое перечисление, тогда возвращаемый массив текстур будет содержать только данные этой конкретной грани, правильно? Из того, что я собираю и понимаю из документации, является то, что это значение индекса для array вашего ресурса карты куба.

Что видно из их документации:

arrayIndex - Индекс массива для текстур массива или индекс лица куба-карты, как определено CUarray_cubemap_face для текстур куба-карт для подресурса для доступа к

Подробнее: http://docs.nvidia.com/cuda/cuda-driver-api/index.html#ixzz4icHnwe9vСледуйте за нами: @GPUCcomputing on Twitter | NVIDIA на Facebook

который является unsigned int или местом индекса в текстурах, которые составляют cube map, типичная карта куба будет иметь 6 faces или самое большее 12, если отображаться как внутри, так и снаружи куба, Поэтому, если мы посмотрим на карту куба, а также на текстуры и их связь с псевдокодом, мы увидим, что:

// Texture
struct Texture {
    unsigned pixelsWidth;
    unsigned pixelsHeight;        
    // Other Texture member variables or fields here.
};

// Only interested in the actual size of the texture `width by height`
// where these would be used to map this texture to one of the 6 faces
// of a cube:

struct CubeMap {
    Texture face[6];
    // face[0] = frontFace
    // face[1] = backFace
    // face[2] = leftFace
    // face[3] = rightFace
    // face[4] = topFace
    // face[5] = bottomFace
};

Объект cubemap имеет массив текстур, составляющий его лицо, и в соответствии с документами функция, о которой вы говорите, со своим третьим параметром, запрашивает у вас индекс в этот массив текстур, и общая функция вернет это:

Возвращает в * pArray массив, через который можно получить доступ к подресурсу ресурса ресурсов графического ресурса, который соответствует массиву indexIndex массива и mipLap уровня mipmap. Значение, установленное в * pArray, может меняться каждый раз при отображении ресурса.

Подробнее: http://docs.nvidia.com/cuda/cuda-driver-api/index.html#ixzz4icKF1c00Следуйте за нами: @GPUCcomputing on Twitter | NVIDIA на Facebook

Надеюсь, это поможет ответить на ваш вопрос относительно использования третьего параметра в функции, которую вы пытаетесь использовать из своего API.

Edit

ОП задал при передаче этого перечисления CU_CUBEMAP_FACE_POSITIVE_X третьему параметру вышеупомянутого вызова функции, он будет возвращать только ту грань карты куба, которая оказывается текстурой. При просмотре их документации об этом перечисляемом значении или типе, найденном здесь: enum CUarray_cubemap_face

enum CUarray_cubemap_face - Индексы массивов для граней куба

Значения

• CU_CUBEMAP_FACE_POSITIVE_X = 0x00
  • Положительное X-лицо куба-карты
• CU_CUBEMAP_FACE_NEGATIVE_X = 0x01
  • Отрицательная X-сторона куб-карты
• CU_CUBEMAP_FACE_POSITIVE_Y = 0x02
  • Положительное Y лицо куба map
• CU_CUBEMAP_FACE_NEGATIVE_Y = 0x03
  • Отрицательное Y-лицо cubemap
• CU_CUBEMAP_FACE_POSITIVE_Z = 0x04
  • Положительная Z-поверхность cubemap
• CU_CUBEMAP_FACE_NEGATIVE_Z = 0x05
  • Отрицательная Z-поверхность cubemap

Подробнее: http://docs.nvidia.com/cuda/cuda-driver-api/index.html#ixzz4idOT67USСледуйте за нами: @GPUCcomputing on Twitter | NVIDIA на Facebook

Мне кажется, что при использовании этого метода для запроса или получения информации текстуры, которая хранится в массиве карты куба, требование третьего параметра является этим перечисляемым значением; это не что иное, как 0-index в этот массив. Поэтому, проходя в CU_CUBEMAP_FACE_POSITIVE_X, поскольку третий параметр для меня не обязательно означает, что вы вернетесь только к той текстуре лица. Мне кажется, что поскольку это 0th index, он вернет весь массив текстур. Старый стиль C, проходящий вокруг массивов, как если бы они были указателями.