From ac124b317943c38d8fd772687d5961a05c5ce719 Mon Sep 17 00:00:00 2001
From: Rahul Garg <rahul.garg@amd.com>
Date: Mon, 13 Nov 2017 08:56:39 +0530
Subject: [PATCH] Fix module_api sample

---
 samples/0_Intro/module_api/Makefile            |   6 +++---
 samples/0_Intro/module_api/launchKernelHcc.cpp |   1 +
 samples/0_Intro/module_api/runKernel.cpp       |   1 +
 samples/0_Intro/module_api/test.cl             |  12 ------------
 samples/0_Intro/module_api/test.co             | Bin 9824 -> 0 bytes
 5 files changed, 5 insertions(+), 15 deletions(-)
 delete mode 100644 samples/0_Intro/module_api/test.cl
 delete mode 100755 samples/0_Intro/module_api/test.co

diff --git a/samples/0_Intro/module_api/Makefile b/samples/0_Intro/module_api/Makefile
index 38bd00a6a6..270d4c1211 100644
--- a/samples/0_Intro/module_api/Makefile
+++ b/samples/0_Intro/module_api/Makefile
@@ -5,7 +5,7 @@ endif
 HIPCC=$(HIP_PATH)/bin/hipcc
 HIP_PLATFORM=$(shell $(HIP_PATH)/bin/hipconfig --compiler)
 
-all: vcpy_kernel.code runKernel.hip.out defaultDriver.hip.out
+all: vcpy_kernel.code runKernel.hip.out launchKernelHcc.hip.out
 
 runKernel.hip.out: runKernel.cpp
 	$(HIPCC) $(HIPCC_FLAGS) $< -o $@
@@ -13,8 +13,8 @@ runKernel.hip.out: runKernel.cpp
 launchKernelHcc.hip.out: launchKernelHcc.cpp
 	$(HIPCC) $(HIPCC_FLAGS) $< -o $@
 
-defaultDriver.hip.out: defaultDriver.cpp
-	$(HIPCC)  $(HIPCC_FLAGS) $< -o $@
+#defaultDriver.hip.out: defaultDriver.cpp
+#	$(HIPCC)  $(HIPCC_FLAGS) $< -o $@
 
 vcpy_kernel.code: vcpy_kernel.cpp
 	$(HIPCC)  --genco  $(GENCO_FLAGS) $^ -o $@
diff --git a/samples/0_Intro/module_api/launchKernelHcc.cpp b/samples/0_Intro/module_api/launchKernelHcc.cpp
index e86e44cb24..7c90198b6a 100644
--- a/samples/0_Intro/module_api/launchKernelHcc.cpp
+++ b/samples/0_Intro/module_api/launchKernelHcc.cpp
@@ -72,6 +72,7 @@ int main(){
 		uint32_t one = 1;
 
     struct {
+        uint32_t _hidden[6];
         void * _Ad;
         void * _Bd;
     } args;
diff --git a/samples/0_Intro/module_api/runKernel.cpp b/samples/0_Intro/module_api/runKernel.cpp
index 2759c02140..fb34f80b7b 100644
--- a/samples/0_Intro/module_api/runKernel.cpp
+++ b/samples/0_Intro/module_api/runKernel.cpp
@@ -68,6 +68,7 @@ int main(){
 		uint32_t one = 1;
 
     struct {
+        uint32_t _hidden[6];
         void * _Ad;
         void * _Bd;
     } args;
diff --git a/samples/0_Intro/module_api/test.cl b/samples/0_Intro/module_api/test.cl
deleted file mode 100644
index 81b20cab0e..0000000000
--- a/samples/0_Intro/module_api/test.cl
+++ /dev/null
@@ -1,12 +0,0 @@
-__kernel void memset(char in, __global int* out) {
-int tx = get_global_id(0);
-out[tx] = in;
-}
-
-
-__kernel void vadd(__global float *Ad, __global float *Bd, __global float *Cd, int N){
-int tx = get_global_id(0);
-if(tx < N){
-Cd[tx] = Ad[tx] + Bd[tx];
-}
-}
diff --git a/samples/0_Intro/module_api/test.co b/samples/0_Intro/module_api/test.co
deleted file mode 100755
index a3e6b991e39e55b9967ae08c86be94a583ce0dbe..0000000000000000000000000000000000000000
GIT binary patch
literal 0
HcmV?d00001

literal 9824
zcmeGi+in|0bi8q#khCh|B5|uUq7Q9R;_R+>H@ikcag#Q!>Xh`B3Zc@Bch`yO`hvZV
z8c^LpLG*zK-~;#qzJcbkyp=xZp)W{%;2VH*c4oX@lU6MRTGTy~_spEjoHOT~nZ1l}
z950_xGMPCdV1(Q!8En7E33D8g)kC5IXpY!q9IpGw1j#`^8__s_jw9f5*p3DxMB;Hl
zM>6U?3@%S_j>OU*)Eh;X@NXF38R2>oOFrtMuY}?SfJWCoUKL%5rJe%xC()DhJNIko
zMf0O#;32#!$kGn_hj143B7401*jKa_*+YjBhVX+iPL^249x?ii_Df3(ks7v<{wDk|
zoJh}-u}BWOeHzdgmoIYJ?_qo#2W6+Hrzbw6tnTHP7FPUrr`Bu~>4#LGq8hzC6){eA
z*r`UfS)^P}bT6?DzuIm#!t=FH{UWuDgmAg(vS7)tH`|+nZ<+-wA#(g;XW4JhUo6t3
z>Jqz#V<Yo8Ff}!y1>6$>3okX>R~OsOZi|nutA?ErJHNPmt`r1Kcd^YH-2fcalGZOY
zLl&Ir)={uiq_&zgQ}VpF-|4VQ;KxFm3Sq%ei4d&(#Yl1{(gw<b4c$z-khBpC(d*pG
zX}{g@15lixbXo+T)6TFu%&6|yJATMz%dD~1WotlvzvVaP%XB_TA4sh>MRl3lQhSY$
zPXVM2Ao8e!B+eH4rJ%M3brBOz*BV}tikN}G3Jbdah0PWQy03+O?y9@a+I@1#b<cKL
z0P6HQOl?=wBAc=4n7~cPRt8Lq_<~k!`f#lgzH>M>4vFQ~;}Q16q1nEP0f(G_FvEDp
zdo-Z)5Dz0CW%36LY=e0o*ZEh<E1sYzubA+N9LGWfIj#mx_Mc=qE}DU)JZ~<ET`pdn
zG2s#W!fuNb)@HvbpZiq^>oA|si)SgkcMk*J>)*g9D)BtEB_2fZULMh3;u?1Z9iIpA
zhh6!I!6n3%-$4kj^8`KZq!CCXkVYVlz%vyA+{gJ3cYN5{#XkQqCrji<au=AB@a{hd
z@^u_(ll+aj&(3DExjz8<glDLc{gP~rZe_k8iqgv{*&ZP$cOV3roRaH}DdW&z*ny~I
zZxUr2=Qu)g+Zd@W;NKwM+(w;G4ge-!-5XaF(t`FV`T3V@rlf$du8?t^knisOkc;k?
zdS+Cy32E){j~Jide%hhftNgz;Fp)|4Rxk+v$NXcSDNF+M{u<yG=33+!^9oNtx}w~=
z{xRM9sYkYM9VEAI-vk{HB_kPGBcyi|+I($54Zsr<*35{quQ#ep^l~Jtk3k(}#&UKR
z>LfEcVMCSZnf+KJ%R4w0R2tOC@(z5D$wWS6@Ug44gVlun9{K%iLY{sdMXsioGy-V^
z(g>sxNF$I&AdNs8fiwbX1fJ;#$g}1LFK{;}g}3<5gAkpy16~%d`|_=Sf5a7M&d^pR
zL(u!^<orB!12{^g8~iAknt3&^QhKDqI=)An4T>Tf5VcuNGp1clS53#5wk_3hJ#SjG
zOx18bdX(%o%#rmlY;}q=Gi$YQy<5q<&HBta=K0M=bZFVEx4NM}6KUo_glcZZa~$2Q
zT2|Gr*sf{m%=dJM`F_PTD_*6l8LDBaN9p5@90VKnXkbG(w1VbY1&>t<o?0***LImv
zowa5Sk6G@lqj|3B8GAGhpIMqyFb&r>wb`of+XcJoR8-Y44cjz*%hoJI_x&fHhV8hj
zrMX^3ovr!>rn_pjU>H`V;5eq~!FhY7QgwYtJxa>8##O&vq~-DgPia5pyEX7T16EV{
z<s_GXDHtW&L-f~COo8g8NqMZh-Oejsk@TVS3X*R$L!acm%|>UlPV(!lv(9^Ao8&|P
zS{U_Mz2=fU|3gFG?RYF?1Z%18H$VV1LskL2-U-{h#W6W{OuM2*SDg%=d|X-h;7n=h
z)O^zUztO|EVT_`$N2o6w0?Yq%APWKiw=oE)i)&h{$#pLANg<4SXbZn>A;|TA0NS_*
zNPW4^C8oebkbG%h;#Yx(=S7k)_YsNZK7wN)@H*jbydD%ZxlhS+DvA65^N}h;0~`W6
z>Y@)3&?H_JUFol!zx<v#1^l5QBaKxm6o=5T?7P56nmQyW^?w^u80wM*(37zn9Af!x
x-uqS%F~>Lu>A%ch8EE~v3t2?>h5knaG;9vrUjf=~`tM<20mLx81J_~wKLJkil%fCt