ELECO '2012 Elektrik - Elektronik ve Bilgisayar Mühendisliği Sempozyumu, 29 Kasım - 01 Aralık 2012, Bursa

CuEEG: EEG Verilerinin Hızlı İşlenmesi için GPU Tabanlı Bir Yaklaşım CuEEG: A GPU-Based Approach for Fast Processing of EEG Data Musa PEKER1, Baha ŞEN2 1

Bilgisayar Mühendisliği Karabük Üniversitesi [email protected]

2

Bilgisayar Mühendisliği Yıldırım Beyazıt Üniversitesi [email protected]

a data base used for diagnosis of epileptic attacks. Data show that CUDA programming has positive effects on the results. It is believed that this approach can be used for the building the infrastructure for a speedy system for real time applications by using graphic processors for the classification of not only EEG data but also of other biometric signals.

Özet Bu çalışmada EEG verilerinin sınıflandırılmasına yönelik yeni bir yaklaşım sunulmaktadır. Bu yaklaşımın dikkate aldığı kriter hız parametresidir. Şimdiye dek gerek epilepsi teşhisi gerekse EEG verilerinin sınıflandırılmasına yönelik yapılmış hemen hemen bütün çalışmalar sınıflandırma doğruluğunu iyileştirmeye yönelik olarak gerçekleştirilmiştir. Bu çalışmada özellikle gerçek zamanlı uygulamalarda bir vazgeçilmez olan hız parametresi dikkate alınmakta ve EEG verilerinin hızlı bir şekilde sınıflandırılması için yeni bir yaklaşım sunulmaktadır. Bunun için özellikle son yıllarda popülaritesi gittikçe artan grafik işlemciler üzerinde programlama yapmamızı sağlayan CUDA programlamadan yararlanılmıştır. Grafik işlemcilerden sınıflandırma aşamasında yararlanılmıştır. Sınıflandırma algoritması olarak EEG verilerinin sınıflandırılmasında yüksek doğruluk oranı veren geri yayılımlı sinir ağı tercih edilmiştir. Bu çalışmada epileptik nöbet tespiti için kullanılan bir veri tabanı üzerinde deneyler gerçekleştirilmiştir. Sonuçlar incelendiğinde CUDA programlamanın sonuçlar üzerinde olumlu etkilerinin olduğu görülmüştür. Bu şekilde sadece EEG verilerine yönelik değil diğer biyomedikal işaretlerin sınıflandırılmasında da grafik işlemciler kullanılarak gerçek zamanlı uygulamalara yönelik hızlı bir sistemin altyapısı oluşturulabilmektedir.

1. Giriş GPU’ lar (Graphics Processing Unit – Grafik İşleme Birimi), esas olarak grafiksel işlemlerde CPU (Central Processing Unit – Merkezi İşlemci Birimi) üzerindeki yükü azaltmak ve daha iyi performans sağlamak için özelleşmiş yapılardır. Günümüzde, GPU’lar da CPU’lar gibi çok çekirdekli yapılara sahip hale gelmiştir. Paralel çok çekirdekli, genel matematik işlemlerini grafik işleme birimi üzerinde yapmaya olanak sağlayan GPU’ lar, grafik kartlarının günümüzde özellikle yüksek performans gerektiren sinyal işleme algoritmalarının gerçeklemesinde önemli bir yere gelmesini sağlamıştır [1]. Sinyal işleme yazılımlarında kullanılan algoritmaların karmaşıklığı arttıkça ve verilerin boyutları büyüdükçe büyük verileri kısa sürelerde işlemeye yönelik bir arayış ortaya çıkmaya başlamıştır. GPU’ ların kabiliyetlerini arttırmaya yönelik olarak 2007 yılında serbest bir platform olan CUDA’ nın NVIDIA tarafından piyasaya sürülmesinin ardından, bu alana ilgi artmış, grafik kartları üzerinde de genel amaçlı matematiksel işlemlerin yapılabileceği görülmüş ve ilerleyen süreçlerde bu alanda çalışmalar hızlanmıştır [1]. Bu çalışmada CUDA Programlama ile EEG verilerinin daha hızlı sınıflandırılması amaçlanmıştır. Sınıflandırma aşamasında, genellikle EEG verilerinin sınıflandırılmasında yüksek doğruluk değerleri veren geri yayılımlı sinir ağı tercih edilmiştir. Yapay sinir ağlarının eğitimi, optimal performansa ulaşmak için gerekli ağırlık güncellemeleri ve yüksek iterasyon sayısından dolayı zaman alıcı olmaktadır [2]. Hesaplama maliyetini azaltmak veya yakınsama hızını yükseltmek için çaba gösterilen çalışmalar mevcuttur [3, 4]. CUDA üzerinde geri yayılımlı sinir ağı uygulaması gerçekleştirilirken BLAS kütüphanesinin bir CUDA uygulaması olan CUBLAS kullanılarak süreç basitleştirilmiştir. Bununla beraber CUBLAS gerekli bütün kütüphanelere sahip olmadığı için çekirdek (kernel) fonksiyon kullanımına ihtiyaç duyulmuştur.

Abstract The study presents a new approach in the classification of EEG data. The basic criterion in this approach is the parameter of speed. Almost all up to date studies both for the diagnosis of epilepsy and for the classification of EEG data aim to improve classification accuracy. This study focuses on the speed parameter which is an indispensable part in real time applications and presents a new approach to classify EEG data in a faster manner. CUDA program has been utilized that makes it possible to make programs on Graphic Processors which is highly popular especially in the last years. Graphic processors have been used in the classification phase. As classification algorithm, back propagation neural network with a higher accuracy rate has been preferred in the classification of EEG data. The study utilized experiments on

685

ELECO '2012 Elektrik - Elektronik ve Bilgisayar Mühendisliği Sempozyumu, 29 Kasım - 01 Aralık 2012, Bursa

Uygulama epilepsi ve oluşan bir veri tabanı grafik kartı üzerinde uygulama, performans performans vermiştir.

sağlıklı insanların EEG işaretlerinden üzerinde gerçekleştirilmiştir. NVIDIA tek gizli katman ile gerçekleştirilen açısından CPU’ ya göre daha iyi bir

rastgele bir ağırlık kümesi ile başlanır, Q katmanlı ileri beslemeli bir ağ için geri yayılım algoritması [7]; 𝑞 q = 1, 2, 3, ........ , Q katman numarası, 𝑋𝑖 : q’ inci katmandaki 𝑞 i biriminin girdisi, 𝑦𝑖 : q’ inci katmandaki i biriminin girdisi, 𝑞 𝑤𝑖𝑗 : (q-1)’inci katmandaki i birimini, q’ ncu katmandaki j birimine bağlayan ağırlık olmak üzere aşağıdaki adımlarla gerçekleştirilir: 1. Adım: w’ ye reel değerli küçük rastgele sayıları başlangıç değeri olarak atanır. 2. Adım: Rastgele bir (giriş-hedef) çalışma modeli seçilir ve q katmanındaki her bir j birimi için ileri yönde çıktı değerleri hesaplanır. Böylece çıkış,

2. Veri Seti Bu çalışmada Bonn Üniversitesi epileptoloji kliniğinde hazırlanan EEG veri tabanı kullanılmıştır [5]. Bu veri tabanında A, B, C, D, E olmak üzere, 5 farklı EEG veri kümesi bulunmaktadır. Bu çalışmada A ve E veri kümesi kullanılmıştır. A kümesi, sağlıklı kişilerden, gözler açık biçimde, yüzey elektrotları ile kaydedilmiş EEG verisini içerir. E kümesi ise hastalıklı kişilerden, epilepsi krizi anında, girişimsel olarak kaydedilmiş EEG verisini içerir. Sinyaller 12 bit analog-sayısal dönüştürücü ile dönüştürüldükten sonra bilgisayar ortamına 173.61 Hz örnekleme frekansı ile aktarılmıştır. Epileptik özellikler kendini 30-40 Hz altındaki frekans bantlarında gösterdiğinden spektral aralığı 0.5-85 Hz olan sinyallere 0.53-40 Hz bant geçiren filtre uygulanmıştır. Örnek EEG görüntüleri Şekil 1’ de görülmektedir.

𝑞

0

1000

2000

3000

4000

𝑞−1

𝛿𝑖

5000

2000 0

-2000

0

1000

2000

3000

4000

𝑞−1

= 𝑓 ′ �𝑋𝑖

𝑞

𝑞

� � 𝛿𝑖 𝑤𝑖𝑗

(3)

𝑖

5. Adım: Ağırlık değerleri aşağıdaki denklemlere göre güncellenir. 𝑦𝑒𝑛𝑖 𝑞 𝑒𝑠𝑘𝑖 (4) 𝑤𝑖𝑗 = 𝑤𝑖𝑗 + ∆𝑤𝑖𝑗 Öğrenme katsayısı η kullanılarak, öğrenme aşamasında 𝑞 ağırlıkta yapılan değişim ∆𝑤𝑖𝑗 , aşağıdaki eşitliğe göre olmalıdır. 𝑞 𝑞 𝑞−1 ∆𝑤𝑖𝑗 = 𝜂𝛿𝑖 𝑦𝑖 (5) 6. Adım: 2. adıma dönüp, toplam hata kabul edilebilir bir düzeye gelene kadar her bir p modeli için işlemler tekrarlanır. Bu çalışmada Q=3 için işlemler yapılmıştır.

a)Sağlıklı kişiden alınan EEG kaydı Genlik (μV)

(1)

3. Adım: Çıkış birimleri için hata terimleri (𝛿) hesaplanır: 𝑄 𝑞 𝑝 𝑄 (2) 𝛿𝑖 = �𝑦𝑖 − 𝑦𝑖 �𝑓 ′ �𝑋𝑖 � 4. Adım: q = Q, Q −1, ... , 2 katmanlarındaki tüm i birimleri için geriye yayılımla gizli katman birimleri için hata terimleri hesaplanır.

0

-200

𝑞

𝑤𝑖𝑗 �

𝑖

200

Genlik (μV)

𝑞−1

𝑦𝑖 = 𝑓 �� 𝑦𝑖

5000

b) Epilepsi hastasından alınan EEG kaydı

4.2. Cuda Programlama

Şekil 1: A ve E veri setlerinden alınan örnekler

CUDA, GPU üzerindeki yüzlerce çekirdeği kullanarak genel amaçlı matematik işlemleri yapmaya olanak sağlayan bir GPU mimarisi, yazılımı ve programlama platformudur [1]. CUDA’ nın başarılı sonuçlar üretmesinin altında yatan neden, CUDA mimarisinin, grafik işleme gibi işlem yoğunluğu olan ve yüksek derecelerde paralellik gerektiren işlemler için geliştirilmiş olmasından kaynaklanmaktadır.

3. Öznitelik Çıkarma Bu aşamada EEG sinyallerinden anlamlı bilginin keşfi için 9 öznitelik çıkarma algoritması kullanılmıştır. 4096 noktadan oluşan her EEG sinyali bu öznitelik çıkarma aşamasından geçtikten sonra 9 nokta ile temsil edilmiştir. Bu öznitelikler istatiksel ve enerji tabanlı özniteliklerdir. Bu çalışmada tercih edilen öznitelik algoritmaları sırasıyla minimum değer, maksimum değer, standart sapma, aritmetik ortalama, varyans, ortanca, sıfır geçiş sayısı, ortalama teager enerjisi ve ortalama enerji değerleridir.

4. Sınıflandırma Bu bölümde ilk aşamada sınıflandırma aşamasında kullanılan geri yayılım sinir ağı hakkında kısa bir bilgi verilmiştir. İkinci aşamada CUDA programlamadan bahsedilecektir.

Sekil 2: GPU veri işleme için daha fazla transistor tahsis eder.

CUDA programlama ile ilgili geniş bir bilgi [8] numaralı kaynaktan alınabilir.

4.1. İleri Beslemeli Geri Yayılımlı Sinir Ağı

5. Uygulama ve Değerlendirme

Danışmanlı öğrenen, ileri beslemeli geri yayılım ağları, adını hatayı yayma biçiminden almıştır. Hata azaltma işlemini hatayı tüm ağa yayarak gerçekleştirdiği için bu ağa “Hatayı Geriye Yayma” da denmektedir. Bu algoritmada elde edilen çıktı ile olması gereken çıktı arasındaki fark yani hata değeri, tüm ağırlıklara yansıtılarak dereceli olarak minimum düzeye indirilmeye çalışılır [6]. Geri yayılım algoritmasında eğitime

Bu çalışmada EEG verilerinin sınıflandırılması için önerilen yapıya ait blok gösterim Şekil 3’ de verilmektedir. Bu yapıda verilerin özniteliklerinin çıkarılabilmesi için hem A veri seti (sağlıklı), hem de E veri seti (epileptik aktivite durumu) için 100*4096 boyutunda iki ayrı vektör matris oluşturulmuştur. Bu vektörlerden, 100*4096 boyutundaki A ve E veri setinin her bir sütunu için 9 adet öznitelik parametresi elde edilerek geri yayılımlı sinir ağı ile sınıflandırılmıştır.

686

ELECO '2012 Elektrik - Elektronik ve Bilgisayar Mühendisliği Sempozyumu, 29 Kasım - 01 Aralık 2012, Bursa

EEG Sinyalleri

1 2

….

Global Bellek (Global Memory)

CPU 𝑊11 𝑊21 � … 𝑊𝑛1

4096 Aritmetik Ortalama Minimum Değer Ortanca Değer Maksimum Değer

İstatistiki ve enerji tabanlı Ortalama Teager Enerjisi öznitelik Varyans çıkarma Sıfır Geçiş Sayısı Ortalama Enerji Standart Sapma

GPU

İleri beslemeli geri yayılımlı sinir ağı

𝑊12 𝑊22 … 𝑊𝑛2

… … … …

𝑊1𝑛 𝑋11 𝑊2𝑛 𝑋 � × � 21 … … 𝑊𝑛𝑛 𝑋𝑛1

𝑋12 𝑋22 … 𝑋𝑛2

Grid

… 𝑋1𝑛 … 𝑋2𝑛 � … … … 𝑋𝑛𝑛

(Buradaki işlemler “cublasSgemm” fonksiyonu kullanılarak gerçekleştirildi.

Epilepsi Sağlıklı

Şekil 3: Epileptik atak tespiti için önerilen yöntem

𝑊11 ⎡ 𝑥 ⎢𝑋 ⎢ 11 ⎢ = ⎢ 𝑌11 + ⎢ ⎣ 𝑌11

𝑊12 𝑥 𝑋21 = 𝑌12

.. .. .. .. ..

𝑊1𝑛 𝑊21 𝑥 ⎤⎡ 𝑥 𝑋𝑛1 ⎥ ⎢ 𝑋11 ⎥⎢ = ⎥⎢ = 𝑌1𝑛 ⎥ ⎢ 𝑌21 + ⎥⎢ thread ⎦ ⎣ 𝑌21

𝑊2𝑛 𝑊𝑛1 𝑊𝑛2 … 𝑊𝑛𝑛 𝑥 ⎤⎡ 𝑥 𝑥 … 𝑥 ⎤ ⎥ ⎢ 𝑋𝑛1 𝑋1𝑛 𝑋2𝑛 … 𝑋𝑛𝑛 ⎥ ⎥⎢ ⎥ = ⎥⎢ = = … = ⎥ 𝑌2𝑛 ⎥ ⎢ 𝑌𝑛1 + 𝑌𝑛2 … 𝑌𝑛𝑛 ⎥ ⎥⎢ ⎥ thread ⎦ ⎣ 𝑌𝑛𝑛 thread ⎦

𝑊22 𝑥 𝑋21 = 𝑌22

… … … … …

blok Paylaşımlı Bellek (Shared Memory)

5.1. CUDA Tabanlı Sinir Ağının Gerçekleştirimi

(Buradaki işlemler için çekirdek fonksiyon yazıldı)

Sinir ağı uygulaması CUDA tarafından uygulanan bir seri matris çarpımı ve aktivasyon fonksiyonu uygulaması olarak gerçekleştirilebilir. CUDA blok başına paylaşılan hafızayı (shared memory) kullanarak matris çarpımını etkin şekilde işleyebilmektedir. Uygulamalarda paylaşımlı bellek verimi önemli oranda artıracaktır. Örnek verilirse; GPU genel işlemlerinde global hafızaya erişim için yaklaşık olarak 400600 döngü gerekliyse, CUDA’ nın hafıza ortamında paylaşılan hafızaya erişimi için sadece 4 döngü yeterli olmaktadır [9]. Bu nedenle CUDA’ nın paylaşımlı hafızası işlemlerin etkin yürütülmesi için yardımcı olmaktadır. Ayrıca matrisin eleman sayısına eşit sayıda dizin (thread) sağlayarak ve daha sonra da her bir dizinde (thread) işlemleri bağımsız olarak hesaplayarak aktivasyon fonksiyonlar da paralel olarak uygulanabilir [9]. Şekil 4’ de CUDA üzerinde gerçekleştirilen sinir ağının ileriye doğru hesaplama adımlarından bir kısmının matris işlemleri sunulmaktadır. Şekil incelendiğinde, sinir ağlarının paralel işlemeye müsait bir yapısının olduğu görülecektir. Geri yayılımlı sinir ağı algoritmasının paralel uygulaması, verilerin GPU’ da matris ve vektör işlemlerinin uygulanmasından ibarettir [10]. Bu çalışmada iki tür paralel işlem mevcuttur: vektör-matris işlemleri ve aritmetik işlemler. Vektör-matris işlemlerine örnek olarak giriş değerleri ile giriş ve gizli katman değerleri arasındaki ağırlık değerlerinin çarpımı verilebilir. Aritmetik işlemlere de örnek olarak sigmoid fonksiyonu ve ağırlık güncellemeleri verilebilir. Bu çalışmada ilk aşamada yani matris ve vektör işlemleri için CUBLAS kütüphanesi tercih edilmiştir. CUBLAS optimum paralelliğe uygun yapılandırılmış bir seri algoritma sunduğundan, bloklar ve dizinlerin tanımlanması problem teşkil etmemektedir [10]. İkinci aşamada yani aritmetik işlemler için ise çekirdek (kernel) tanımlanarak paralelliğin gerçekleştirilmesi sağlanmıştır. Burada kernel, GPU üzerinde çalışan fonksiyona verilen isimdir. Bu iki aşama ile ilgili bazı örnekler aşağıda sunulmaktadır. Matris çarpım işlemlerinde CUBLAS fonksiyonlarından “cublasSgemm” fonksiyonu kullanılmıştır. Örnek bir kullanım aşağıda sunulmaktadır. ‘cublasSgemm (char transa, char transb, int m, int n,int k, float alpha, const float *A, int lda,const float *B, int ldb, float beta,float *C, int ldc)’ A, B ve C matris değerleridir. Alpha ve beta ise skaler değerlerdir.

Sigmoid Fonksiyonu

𝑆11 𝑆21 � … 𝑆𝑛1

𝑆12 𝑆22 … 𝑆𝑛2

… 𝑆1𝑛 … 𝑆2𝑛 � … … … 𝑆𝑛𝑛

Sigmoid Fonksiyonu

Şekil 4: CUDA kullanarak geri yayılımlı sinir ağının matris çarpım işlemleri (ileri hesaplama adımlarından bir kesit)[9]

Bu fonksiyona göre A matrisi mxk, B matrisi kxn ve C matrisi de mxn boyutundadır. cublasSgemm fonksiyonu C = (alpha x A x B) + (beta x C) matris işlemini hesaplar. Alpha=1 ve beta=0 olarak alındığında işlem C=AxB’ ye indirgenir. Lda, ldb ve ldc ise sabit değerlerdir. Fonksiyondaki trans ifadesi matris transpoze işlemleri için kullanılmaktadır. Genellikle ‘n’ yapısı tercih edilir. Kullanımı aşağıda belirtilmiştir. ‘ Eğer transa=’N’ veya ‘n’ , A=A., Eğer transa=’T’,’t’, ‘C’ or ‘c’ ise A=𝐴𝑇 ′dir. transb ifadesi ise aynı amaçla B matrisi için kullanılır.’ CUBLAS kullanırken dikkat edilmesi gereken MATLAB' da olduğu gibi matrislerde verinin "column major (veriler sütun öncelikli olarak bellekte tutulmaktadır)" şeklinde işlendiği ve tutulduğudur. Belirtilen cublasSgemm fonksiyonunun dışında bu çalışmada GPU bellek ayrımı için cublasAlloc, ayrılan bellek alanını tekrar bırakmak için cublasFree, CPU’ dan GPU’ ya vektör aktarımı için CublasSetVector, GPU’dan CPU’ ya vektör aktarımı için CublasGetVector, CPU’ dan GPU’ ya matris aktarımı için CublasSetMatrix, GPU’ dan CPU’ ya matris aktarımı için CublasGetMatrix fonksiyonlarından yararlanılmıştır. Aritmetik işlemlerin paralelliği için kernel tanımlanmıştır. Aktivasyon fonksiyonu, hata fonksiyonu ve ağırlık güncellemeleri için çekirdek (kernel) fonksiyonlar geliştirilmiştir. Sigmoid aktivasyon fonksiyonu ile ilgili hesaplamalarda işlemlerin GPU üzerinde gerçekleştirilmesini sağlayan örnek bir kernel Şekil 5’de sunulmuştur. __global__ void sigmoid( float *cikis1, int maksimum) { __shared__ float u_shared[32]; //Paylaşımlı bellek ayırma int sutun = blockIdx.x * blockDim.x + threadIdx.x; //GPU dizininin (thread) işleyeceği adres bilgisi if (sutun