Программирование

 

Мы уверены, что вы имеете опыт работы с коммерчески доступными x86/GPU/ARM вычислительными системами на языках С/С++. Скорее всего, вы программировали GPU на специализированных языках типа OpenCL или CUDA. Очень вероятно, что как для универсальных, так и для графических процессоров вы использовали специальные, ориентированные на конкретное “железо”, приемы для повышения производительности работы вашего кода. И, как любому программисту, вам бы хотелось, с одной стороны, “поменьше копаться в железе” и использовать уже изученные, привычные для вас, языки программирования, с другой - “выжать из оборудования все, на что оно способно”, и получить действительно быстрый и/или, в зависимости от требования заказчика, энергоэффективный код.

 

Учитывая вышесказанное, отлично понимая трудности работы с любой новой архитектурой, мы собираемся предложить для программирования всех процессоров, построенных на базе архитектуры MALT, универсальный подход, который позволяет начать работать на MALT с C/C++ кодом, либо уже имеющимся у вас, либо написанным на основе предлагаемых нами примеров. Далее, оптимизируя его в соответствии с рекомендациями из нашего руководства, выжать из оборудования все, на что оно способно. Для удобной отладки пользовательского кода мы предоставляем полнофункциональный эмулятор. Для выявления узких мест и оптимизации в комплекте с эмулятором идет профилировщик.

 

Программист, в зависимости от своих предпочтений и требований к оптимизации прикладного кода, может выбрать один из следующих подходов:

 

C++ для MALT. Наиболее простой путь начала работы с MALT. Для программирования скалярных ядер поддерживается стандарт С++17. Работа с потоками осуществляется через конструкции C++ std::thread, std::mutex и другие. MALT выступает в роли классического многоядерного процессора, поэтому портировать существующее ПО и библиотеки легко.

 

OpenCL для MALT. Реализация стандарта OpenCL для работы со скалярными и векторными ядрами MALT. Реализация OpenCL поставляется вместе с библиотекой проблемно-ориентированных алгоритмов, оптимизированных под MALT. Теперь перейти на MALT для пользователей ускорителей AMD, NVIDIA, ARM стало проще.

 

– MALTCС. Функциональный аналог NVCC от NVIDIA. Собственный front end к базовым компиляторам для скалярных и векторных ядер. MALTCC реализует возможности архитектуры и значительно упрощает распараллеливание программ целевых классов.

 

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

 


 

C++ для MALT

 

При использовании потоков C++ решения MALT "выглядят" как обычные многоядерные процессоры. Поддерживаются стандарты языка вплоть до С++17. Функции для использования аппаратных расширений содержатся в пространстве имён malt::

 

Пример программы на С++ для MALT:

 


#include <vector> 
#include <thread>  
#include <functional>  

int main() {
     const int size_x = 1920;
     const int size_y = 1080;
     uint8_t frame[size_x * size_y], buffer[size_x * size_y];
     load_grayscale_image(frame, size_x, size_y);
     memset(buffer, 0, size_x * size_y * sizeof(uint8_t));

     auto f = [](uint8_t *in, uint8_t *out,
                 int start, int end, int shift) {
         out[start] = in[start];
         for (int offset = start + shift; offset < end; offset++)
             out[offset] = in[offset - 1] / 4 + in[offset] / 2 + in[offset - 1] / 4;
         out[end] = in[end];
     };

   std::vector<std::thread>threads;
   threads.resize(size_y);
   for (int y = 0; y < threads.size(); y++)
       threads[y] = std::thread(f, frame, buffer,
                                y * size_x, (y - 1) * size_x - 1, 1);

   for (auto &t : threads) t.join();

   threads.resize(size_x); 
   for (int x = 0; x < threads.size(); x++) 
      threads[x] = std::thread(f, buffer, frame x, x + (size_y - 1) * size_x, size_x); 
   for (auto &t : threads) t.join(); 
   return 0; }

 


 

MALTCC

 

MALTCC - front end, в значительной мере автоматизирующий раздачу заданий для скалярных и векторных ядер в гетерогенной системе MALT. При помощи готовых шаблонов MALTCC позволяет значительно упростить разработку задач целевых классов, скрыв от программиста трудности синхронизации потоков в асинхронной гетерогенной системе. MALTCC - наиболее простой путь разработки под MALT для программистов, уже знакомых с SDK NVIDIA® CUDA® или подобными.

 

Пример программы на MALTCC: 

 


// Подбор паролей по хешу MD5
#include "maltcc.h"
#include "md5.c"

atomic_t found;
void cpu_check_md5(const char* s) {
  Hash h = HashFn(MD5_Id,s);
  for (int i = 0; i < N_HASHES; i++)        // Проверка результата
    if (!memcmp(h.bytes(),hashes[i],16)) {
      printf("\tFound! %s\n",s); 
      found++;
    }
}

__simd __kernel check_words(u32 w[W_LEN]) { // Исполняется на ускорителях
  u32 i, j, r;
  u32 hs[N_HASHES], abc[ABC_SZ];                  
  __slave on_init() {
    for (int i=0; i < N_HASHES; i++)        // Первые слова от хэшей
      hs[i] = cpu_to_le(hashes[i][0]);      // для экспресс-проверки
    for (unsigned i = 0; i < ABC_SZ; i++)   
      abc[i] = alphabet[i];
  }
  j = 0; REPEAT(ABC_SZ) {            
    w[0] = (w[0] & 0xFFFF) | abc[j] << 16;  j++;
    MD5(w,r); 
    i = 0; REPEAT(N_HASHES) {
        __success |= r==hs[i]; i++;
    }
  }
  __slave on_success() {                    // Callback при успехе
    u32 ws = le_to_cpu(w[0]);
    char *s = (char*)&ws; s[W_SZ]=0;
    printf("Found? [w=%s]\n",s);
    for (int i = 0; i < ABC_SZ; i++) {      // Просмотр всех окончаний
      s[W_SZ-1] = alphabet[i];
      cpu_check_md5(s);
    }
  }
} 

__slave dict_fragment(int d) {              // Исполняется на скалярных ядрах
  for (u32 i=0; i<ABC_SZ; i++) {            
    u32 w = (u32)(alphabet[d]) | (u32)(alphabet[i]) << 8;
    check_words(&w);                        // Проверяем серию слов на векторных  ускорителях
  }
}

int main(int argc, char *argv[]) {
  for (int i=0; i < ABC_SZ; i++)
    dict_fragment(i);                       // Запуск ф-ии на всех скалярных ядрах
  slave_sync();
  return found;  
}

 

NVIDIA® и CUDA® являются товарными знаками компании NVIDIA Corporation

 


 

OpenCL для MALT 

 

Реализация стандарта OpenCL для работы со скалярными и векторными ядрами MALT. Реализация OpenCL поставляется вместе с библиотекой проблемно-ориентированных алгоритмов, оптимизированных под MALT. Теперь перейти на MALT для пользователей ускорителей AMD, NVIDIA, ARM стало проще.

 

Пример программы на OpenCL для MALT:

 


// Примерный код OpenCL-ядра для выделения особых точек изображения

#define DIM_X   128
#define DIM_Y   128

#define M       9
#define MASK(n) (masks+(n)*M*M)
#define HESS_L  100

int D(__global uchar* img, int x, int y, __constant uchar *mask) {
  x -= M/2; y -= M/2; int v = 0;
  for (int dx=0; dx<M; dx++)
    for (int dy=0; dy<M; dy++) 
      v += mask[dx+M*dy] * img[(x+dx)+DIM_X*(y+dy)];
  return v;    
}

__kernel void mark_ipoints(__global uchar *img, __constant uchar* masks) {
  for (int x=M/2; x<DIM_X-M/2; x++) { 
    for (int y=M/2; y<DIM_Y-M/2; y++) {
      int Dxx = D(img,x,y,MASK(0)), 
          Dyy = D(img,x,y,MASK(1)), 
          Dxy = D(img,x,y,MASK(2));
      img[x+DIM_X*y] |= (abs(5*Dxx*Dyy - 4*Dxy*Dxy) > HESS_L) << 7; 
    } 
  }
}