Android prelink in ICS

在 Android ICS 中已經將 apriori 及 soslim 都丟掉
言下之意就是 prelink 在 Android 中不再出現

根據 Commit 裡面的機制就是說硬體已經進步
所以不缺那點時間跟省下來的空間XD...

而且不是 gcc 那串 toolchain 的東西, 要分開維護很麻煩

並且對於 Address-Space-Layout Randomization 的有效性大大降低

libelf for arm

對於 arm 的 object file 使用 elfutils-libelf 時

取用 Symbol 需要注意 Symbol address 的部份

在 ELF for the ARM®  Architecture 的第 4.6.3 的部份有提到

Symbol Values 

• In addition to the normal rules for symbol values the following rules shall also apply to symbols of type STT_FUNC: If the symbol addresses an ARM instruction, its value is the address of the instruction (in a relocatable object, the offset of the instruction from the start of the section containing it). 
• If the symbol addresses a Thumb instruction, its value is the address of the instruction with bit zero set (in a relocatable object, the section offset with bit zero set). 
• For the purposes of relocation the value used shall be the address of the instruction (st_value & ~1).

ARM 跟 thumb 的 Symbol 區別方式是看 Symbol Table 中 Symbol Address 的最低位元來判斷, 所以取用請記得 st_value & ~1, 免得取到的 Address 差一.

cuda-memcheck : CUDA Debug 好幫手

在撰寫 CUDA 程式的時候最煩惱的就是 Debug 困難

即使現在有了 cuda-gdb 也是相當的困難

而在 CUDA toolkit 裡面有附贈一隻叫 cuda-memcheck 的小程式

可以幫助你檢查程式中有無下列兩種錯誤：


1. Out of boundary memory access
2. Misalignment

那接下來先了解一下上面兩個錯誤是啥

第一個錯誤 Out of boundary memory access 很好理解

比如說你的陣列長度只有 10 但是你卻存取到了 10 以上

int a[10];
a[11] = 1111; // Out of boundary memory access !!

這件事情在傳統 C 語言上面很好察覺,

但在 CUDA 中通常 index 都是跟 threadIdx 以及 blockIdx 加起來之類的

一不小心就會寫錯

而第二個錯誤 Misalignment 則是一般在 x86 PC 上寫程式幾乎不會遇到的事情

原因在於 x86 硬體允許你作這件事, 在嵌入式系統的世界比較容易遇到這種事

alignment 這件事簡單來講就是記憶體存取時位址必須為 x 的倍數

例如 4 - byte alignment 的話, 位址必須能被 4 整除

合法的 4 - byte alignment  位址如 : 0x00ffff, 0x0, 0x8 等等
不合法的 4 - byte alignment  位址如 : 0x01, 0xff06, 0xa 等等

但通常這種情只會發生在你有對指標作些特殊處理時才會發生

例如

char p[10];
int *a = (int*)p+1;

之類的運算, 通常在寫影像處理相關的程式時
比較有可能寫出這類的程式碼
在寫 CUDA 時請盡量避開

在了解 cuda-memchech 的功能後來看看如何使用

假設你有一個 CUDA 程式 a.out

那你只要下

cuda-memcheck a.out

就可以開始檢查

如果有錯誤的話會吐出類似下面的訊息


========= CUDA-MEMCHECK
========= Invalid __global__ read of size 4
=========     at 0x00000038 in f
=========     by thread (0,0,0) in block (0,0)
=========     Address 0xfb00000228 is out of bounds
=========
========= ERROR SUMMARY: 1 error


在這個時候你會發現雖然他告訴你 out of bounds
但是根本不知道在那一行出錯阿！！！

這時候你重新編譯你的程式

nvcc -G oob.cu

加個 -G 然後

cuda-memcheck a.out

重新檢查


========= CUDA-MEMCHECK
========= Invalid __global__ read of size 4
=========     at 0x000000b0 in oob.cu:3:f
=========     by thread (0,0,0) in block (0,0)
=========     Address 0xfb00000228 is out of bounds
=========
========= ERROR SUMMARY: 1 error

就會報告錯誤的行數了！

另外有個 --continue flag

cuda-memcheck --continue ./a.out

這時候他檢測到錯誤的時候就會試著繼續跑跑看了

Misalignment 錯誤訊息差不多也是這樣就不浪費版面了

Add new instruction to SimpleScalar (1) - Modify Simple Scalar

這份教學主要包含三個部份

1. 加指令到 SimpleScalar  (實驗中採用該版本 http://www4.ncsu.edu/~hhashem/mase-alphalinux.htm)
2. 加指令到 gas (Binutils 2.20)
3. 加指令到 gcc (GCC 4.4.3)

不過著重在前兩項, 第三項沒背景知識一時半刻也弄不來

教學範例是為 Alpha 加上整數除法!

(Alpha真他媽的畸形 有浮點數除法沒整數除法)


為你的SimpleScalar增加新的指令!

雖然這鬼東西已經夠老了, 不過可怕的是中文資源幾乎都只教怎麼安裝

沒教怎麼改

而增加指令只有首頁那該死的 def file format 可以參考

去年面對的時候選擇了逃避

沒想到其實沒有想像中的難...

就進入正題吧

http://www.simplescalar.com/docs/README-def.txt

有空還是請記得讀一下官方文件



1. 加指令到 SimpleScalar

為了達到增加指令的裡面，我們必須要先了解 SimpleScalar 是怎麼 decode 的

首先 SimpleScalar 採用一種名為 Decode Tree 的鬼東西，大致上來說就是需要在新增指令的時候大概按照他的樹狀結構編排

Decode Tree 的開端來自於 MD_TOP_OP(INST) 這個 Macro


這個是來自 ARM 的

#define MD_TOP_OP(INST) (((INST) >> 24) & 0x0f)

ARM 使用 23~27 bit 當主要的 Opcode
28~31被拿去當 Condition Code

這個是來自 Alpha 的

#define MD_TOP_OP(INST)         (((INST) >> 26) & 0x3f)

Alpha 則是前 7 bit 都抓去當 Opcode

所以有了 MD_TOP_OP 後我們可以開始建構 Decode Tree 了

以Alpha 為例

LDA 的 OPCODE 是 0x08

那麼我們就直接定義


DEFINST(LDA,                    0x08,
        "lda",                  "a,o(b)",
        IntALU,                 F_ICOMP,
        DGPR(RA), DNA,          DNA, DGPR(RB), DNA)

第一個參數是拿來辨別的
第二個參數是OPCODE
第三個參數是拿來 disasm 用的資訊
第四個參數也是拿來 disasm 用的資訊
第五個參數是會用到哪個 Functional Unit, 這個在 sim-outorder 下會有影響
第六個參數則是分類該指令, 同樣的也是在 sim-outorder 下會有影響
接下來六七個是 Output Register , 拿來控制 Dependency 用的
八九十則是 Input Register , 也是拿來控制 Dependency 用的
所以後五個參數, 請直接跟同類型的指令拷貝

注意一下 ARM 的參數個數有點不太一樣, 主要差在後面不止五個

接著請記得在這個 Marco 的前面定義一個

LDA_IMPL 的 Macro


#define LDA_IMPL                    \
{                                   \
  SET_GPR(RA, GPR(RB) + SEXT(OFS)); \
}

這個是用來實作該指令的功能的


看到這邊應該正常來講已經可以做到基本的增加指令了

但是通常事情沒那麼簡單

常常有 OPCODE 外還會有 Function Code , Alpha 中Operate Instruction 就是這樣, 在 MIPS 中也有

那麼我們就不能直接像 LDA 般那麼歡樂的加指令

這時候套個 SimpleScalar 說明文件的術語

我們需要 Sub Decode Tree !

意思是說我們用 OPCODE 分了某類後想要再依照某些欄位分類的話就要 Sub Decode Tree

阿 Sub Decode Tree 到底是什麼 ?

簡單的說就是類似再一次的設定 MD_TOP_OP 的那種感覺

不過已經不是 TOP 而是 SECOND 了, 喔 這不是重點

接下來我們需要 DEFLINK 這個東西來定義 Sub Decode Tree Root

例如以 Alpha add, sub 等等的 OPCODE 0x10 為例

我們只要這樣定義

DEFLINK(INTA, 0x10, "inta", 5, 0x7f)

第一個欄位是該  Sub Decode Tree 的名稱
第二個欄位是OPCODE
第三個欄位是註解
第四個欄位是新的OPCODE的偏移量
第五個欄位是新的OPCODE的MASK

所以這個 DEFLINK 的效用大概會等於

MD_SECOND_OP(INST) ((INST >> 5) & 0x7f)

喔, 不過沒有 MD_SECOND_OP 這麼 Macro , 別真的去宣告


就會出現一顆 Sub Decode Tree 了


接下來很重要的是我們需要 CONNECT 兩棵 Decode Tree

CONNECT(INTA)

然後就只要符合 MD_TOP_OP(INST) = 0x10 的都會掉進這顆叫 INTA 的 Sub Decode Tree 了~

接著我們只要開心的 DEFINST 就可以了~

不過其實很多情況下我們並不只想要兩層 Decode Tree 就解決, 有時候可能會需要三層甚至四層, 這時候該怎麼辦呢?

方法就是在該 Sub Decode Tree 的 CONNECT 後再 DEFLINK

那麼這個 DEFLINK 就會是 Sub Sub Decode Tree 了

在 Alpha 中就用這種無聊的技巧來分別第三個 Operand 是 Imm 還是 Reg

對, 真的很無聊...明明就用個 if 就解決了, 還要靠這該死的 Decode Tree

如果無聊去看實作的話會發現這樣做只是比較慢而且免費讓 DEF 膨脹到快兩倍而已






下集待續吧~

Qt Creator for CentOS 5.x

用 CentOS 最 g8 的就是一堆比較新的套件都不太能用

要碼自己編不然就放棄
(而且還會遇到一堆套件版本相依性問題!

最近要用到 qt creator 簡直是麻煩到爆炸了

還好有個好心的老外有提供

http://joseph.freivald.com/linux/


安裝方法

1.

rpm -ivh http://software.freivald.com/centos/software.freivald.com-1.0.0-1.noarch.rpm

2.

然後就可以開心的用 yum 安裝了

yum install qt-creator

歐~人生真是美好~

DebugStream for C++

千年發文

其實這是剛才心血來潮寫的小 helper class

主要用途是來控制輸出資訊用的

並且可以在輸出時指定輸出等級, 例如 lv 3 以上才輸出之類的

廢話不多說先來段使用範例

#include "DebugStream.hpp" 
// 使用前只要 include header 就可以了
#include <iomanip>  
// iomanip for std::setprecision
int main(){
  DebugStream ds(10);       
  // 開一個 DebugStream, debug level 設為 10
  ds[3] << "abc\n";   
  // verbose level 3, 小於debug level 所以會被印出來
  ds[11] << "XD\n";   
  // verbose level 11, 大於debug level 所以不會被印出來
  ds[8] << 123 << std::endl;
  ds.precision(2);          
  // 跟平常使用 cout 習慣一樣喔~
  ds[5] << 0.54321 << std::endl; 
  ds[5] << std::setprecision(4) << 0.12354 << std::endl; 
  // 當然也可以採用這種方式~
  return 0;
}

預計未來可能應該大概會加入 printf 的功能來達到 ds[3].printf("%d", 3); 之類的功能

code 按這裡取得完整程式碼

ARM System call in SimpleScalar

對ARM的認知大致上都是從課堂跟GCC的md檔來的

但是對於這些OS底層實作相關的知識還真是薄弱

說正確一點是弱到不行

SimpleScalar 的 system call實作上大致上是在

target-arm/syscall.c裡面

有時候遇到

warning: invalid/unimplemented syscall xxx, PC=0x0001c454, winging it

的時候可以直接略過沒關係，但是...事情總不會那麼順利

而且SimpleScalar早就沒在更新了，也大概不用期待下一版會解決這種事發生

以rename來當例子好了，忘記是哪個benchmark需要用到，少了就不能正常運作

rename 的 system call 編號是 38

所以其實要補system call的第一步是先查出缺什麼，並且它是不是必要的，
(必要的定義在於沒補齊就會跑不出數據...

接著把編號拿去syscall.c裡面查查看是啥

以38為例

#define ARM_SYS_rename 38

由此可知編號38是rename，接著丟到Google大神去查一下各參數的功能是啥

int rename ( const char * oldname, const char * newname );

rename就如他名字一般，就是拿來改檔案的

所以我們其實只要把參數在丟到對應的system call就好

第一個參數會放在R0

第二個則在R1

所以直覺上我們這樣寫

regs->regs_R[MD_REG_R0] =
rename(regs->regs_R[MD_REG_R0], regs->regs_R[MD_REG_R1]);

不過會發現一件事，regs->regs_R[MD_REG_R0]所存的指標不是host所用的address

直接傳進去可能會segment fault，或是發生其他不可預期的事

參考一下其他system call的實作可以發現有一個函數可以用

char buf[MAXBUFSIZE];
mem_strcpy(mem_fn, mem, Read, /*fname*/regs->regs_R[MD_REG_R0], buf);

從simulator 所用的記憶體中搬出東西來，第二個參數也照做就行了

接著補上頭跟尾
case ARM_SYS_rename:
{
char buf[MAXBUFSIZE];
char buf2[MAXBUFSIZE];
/* copy filename to host memory */
mem_strcpy(mem_fn, mem, Read, /*fname*/regs->regs_R[MD_REG_R0], buf);
mem_strcpy(mem_fn, mem, Read, /*fname*/regs->regs_R[MD_REG_R1], buf2);
/* chmod the file */
/*result*/
regs->regs_R[MD_REG_R0] =
rename(buf, buf2);

break;
}


然後補上錯誤處理部份

case ARM_SYS_rename:
{
char buf[MAXBUFSIZE];
char buf2[MAXBUFSIZE];
/* copy filename to host memory */
mem_strcpy(mem_fn, mem, Read, /*fname*/regs->regs_R[MD_REG_R0], buf);
mem_strcpy(mem_fn, mem, Read, /*fname*/regs->regs_R[MD_REG_R1], buf2);
/* chmod the file */
/*result*/
regs->regs_R[MD_REG_R0] =
rename(buf, buf2);

/* check for an error condition */
if (regs->regs_R[MD_REG_R0] == (qword_t)-1)
/* got an error, return details */
regs->regs_R[MD_REG_R0] = -errno;
break;
}

會用qword_t則是有一些歷史原因，simplescalar最原本是porting到alpha上用的，大概是因此沿用下來的慣例

俗話說的好要入境隨俗，所以留著吧

然後再把這段code插入sys_syscall的那個超大switch中的某個適合的地方就可以了



後記：

其實最讓我驚訝的還是在

system call的時候

第五個參數竟然就直接放在r4，這跟原本認知第五個參數以後要丟stack不一樣

至於第六個或更多參數呢...目前還沒看到，遇到再補完這邊文章