Компьютерная графика

Информация о пользователе

Привет, Гость! Войдите или зарегистрируйтесь.

Вы здесь » Компьютерная графика » Программирование графики и GPU » Быстрое копирование памяти в ядре.

Быстрое копирование памяти в ядре.

Страница: 1

Сообщений 1 страница 5 из 5

1

  • Автор: VirusVFV
  • Новичок
  • Зарегистрирован: 2011-04-10
  • Приглашений: 0
  • Сообщений: 5
  • Уважение: [+0/-0]
  • Позитив: [+0/-0]
  • Провел на форуме:
    1 час 7 минут
  • Последний визит:
    2011-04-24 12:48:52

Всем доброе время суток.
Хочу задать вопрос.
Допустим есть массив:
__device__ __constant__ unsigned char constArray[64];

в функции ядра его необходимо скопировать в локальную или глобальную переменную  например  unsigned char cudaArray[64];
Можно это сделать обычным циклом :   for(i=0;i<64;i++) cudaArray[i]=constArray[i];
Но, насколько я понимаю, копировать из константной памяти в глобальную посимвольно - это довольно таки медленно.

Как это можно сделать более быстро, если можно ?
есть ли аналог memcpy для функций ядра ?

0

2

  • Автор: VirusVFV
  • Новичок
  • Зарегистрирован: 2011-04-10
  • Приглашений: 0
  • Сообщений: 5
  • Уважение: [+0/-0]
  • Позитив: [+0/-0]
  • Провел на форуме:
    1 час 7 минут
  • Последний визит:
    2011-04-24 12:48:52

Фокус типа:

Код:
 uint64_t *sstr;
 uint64_t *dstr;

 sstr=(uint64_t*)constArray;
 dstr=(uint64_t*)cudaArray;

 dstr[0]=sstr[0];
 dstr[1]=sstr[1];
...
 dstr[8]=sstr[8];

Не проходит. Компилятор выдает ошибку:
Error: Unaligned memory accesses not supported

0

3

  • Автор: Steps3D
  • Администратор
  • Зарегистрирован: 2008-07-29
  • Приглашений: 0
  • Сообщений: 303
  • Уважение: [+8/-1]
  • Позитив: [+4/-0]
  • Провел на форуме:
    3 дня 12 часов
  • Последний визит:
    2019-07-02 19:07:48

А может копировать 32-битовыми словами ?
И в чем смысл копирования константной памяти в глобальную в ядре, когда это можно сделать на хосте проще и быстрее.
Весь смысл константной памяти - в скорости доступа.

0

4

  • Автор: VirusVFV
  • Новичок
  • Зарегистрирован: 2011-04-10
  • Приглашений: 0
  • Сообщений: 5
  • Уважение: [+0/-0]
  • Позитив: [+0/-0]
  • Провел на форуме:
    1 час 7 минут
  • Последний визит:
    2011-04-24 12:48:52

Дак вот я и пытаюсь копировать 32(64)-битными словами:
sstr=(uint64_t*)constArray;
dstr=(uint64_t*)cudaArray;
dstr[0]=sstr[0];

но и в том и в другом случае: Error: Unaligned memory accesses not supported.
Кстати, Cuda 3.1 ошибки не выдает, а Cuda 3.2 и 4.0 выдают ошибку.

По поводу необходимости:
cudaArray - массив, состоящий из общей для всех нитей строки (соли) и индивидуальной для каждой нити части (пароля)
Затем этот массив (строка) передаётся __device__ функции для  обсчета MD5 и т.д.
Если не составлять строку в ядре - то это чревато дополнительным количеством ветвлений, что например на gtx480 еще  больше снижает быстродействие нежели составление строки.

0

5

  • Автор: VirusVFV
  • Новичок
  • Зарегистрирован: 2011-04-10
  • Приглашений: 0
  • Сообщений: 5
  • Уважение: [+0/-0]
  • Позитив: [+0/-0]
  • Провел на форуме:
    1 час 7 минут
  • Последний визит:
    2011-04-24 12:48:52

Дополнение:
Если cudaArray объявить как
__device__ __constant__ __align__(8) unsigned char constArray[64];
то ошибка исчезает.  8-)

Помимо копирования 32(64)-битными словами, есть ли еще какие способы, более быстрые по скорости ?

0

Страница: 1

Вы здесь » Компьютерная графика » Программирование графики и GPU » Быстрое копирование памяти в ядре.

создать форум