summary | shortlog | log | commit | commitdiff | tree
raw | patch | inline | side by side (parent: 66a1477)
raw | patch | inline | side by side (parent: 66a1477)
author | Jianzhong Xu <xuj@ti.com> | |
Mon, 4 Jan 2016 16:45:49 +0000 (16:45 +0000) | ||
committer | Jianzhong Xu <xuj@ti.com> | |
Mon, 4 Jan 2016 16:45:49 +0000 (16:45 +0000) |
159 files changed:
diff --git a/Makefile b/Makefile
index a42fdba8394caebf2b088e5534ef1fd950de3129..b7e5e3d9907f0a061ef8b5a1949d40b1fbe174c9 100644 (file)
--- a/Makefile
+++ b/Makefile
cd ../$(LINALG_BLIS_DIR); ./configure -p install/$(BLIS_CFG) c66x; make -j8 MEM_MODEL=$(MEM_MODEL) TARGET=$(TARGET) LIBOS=$(LIBOS); make install; \
cd ../$(LINALG_TICBLAS_DIR)/src; make MEM_MODEL=$(MEM_MODEL) TARGET=$(TARGET) LIBOS=$(LIBOS); cd ../lib; \
echo "combining BLIS, CBLAS, and TICBLAS libraries into one: libcblas.ae66"; \
- mkdir -p objs; cd objs; rm *; ar x ../../../blis/install/c66xMedium/lib/libblis.ae66; mmv 'cblas*.o' 'blis_cblas#1.o'; \
+ mkdir -p objs; cd objs; rm *; ar x ../../../blis/install/$(BLIS_CFG)/lib/libblis.ae66; mmv 'cblas*.o' 'blis_cblas#1.o'; \
ar -x ../../../cblas/lib/C66/libcblas.ae66; ar -x ../libticblas.ae66; chmod +rw *;cd ../../..; \
- mkdir -p lib; cd lib; rm *; ar -cr libcblas.ae66 ../ticblas/lib/objs/*;
+ mkdir -p lib; cd lib; rm *; ar -cr libcblas.ae66 ../ticblas/lib/objs/*; cd ..
-
-ARMplusDSP:
+ARMlibs:
cd $(LINALG_CBLAS_DIR); make arch=ARM alllib; \
cd ../$(LINALG_BLIS_DIR); ./configure -p install/arm cortex-a15; make -j8; make install; \
- cd ../$(LINALG_BLASACC_DIR)/src; make MEM_MODEL=$(MEM_MODEL) TARGET=$(TARGET); cd ..; \
- cd ../$(LINALG_CLAPACK_DIR); make f2clib; make cblaswrap; cd SRC; make -j8
+ cd ../$(LINALG_CLAPACK_DIR); make f2clib; make cblaswrap; cd SRC; make -j8; cd ..
+ARMplusDSP: DSPlibs ARMlibs
+ cd $(LINALG_BLASACC_DIR)/src; make debug MEM_MODEL=$(MEM_MODEL) TARGET=$(TARGET); cd ../..; \
+ cp $(LINALG_BLASACC_DIR)/lib/libcblas_armplusdsp.a ./lib; \
+ cp $(LINALG_BLIS_DIR)/install/arm/lib/libblis.a ./lib; \
+ cp $(LINALG_CLAPACK_DIR)/lapack_ARM.a ./lib/liblapack.a; \
+ cp $(LINALG_CLAPACK_DIR)/libcblaswr_ARM.a ./lib/libcblaswr.a; \
+ cp $(LINALG_CLAPACK_DIR)/F2CLIBS/libf2c_ARM.a ./lib/libf2c.a
cleanDSPlibs:
- cd $(LINALG_CBLAS_DIR); make clean; \
+ cd $(LINALG_CBLAS_DIR); make arch=C66 clean; \
cd ../$(LINALG_BLIS_DIR); ./configure -p install/$(BLIS_CFG) c66x; make -j8 clean; \
- cd ../$(LINALG_TICBLAS_DIR)/src; make clean; cd ../lib/objs; rm *; cd ../../..;
+ cd ../$(LINALG_TICBLAS_DIR)/src; make clean; cd ../lib/objs; rm *; cd ../; rm libcblas.ae66; rm -r objs; cd ../..;
+
+cleanARMlibs:
+ cd $(LINALG_CBLAS_DIR); make arch=ARM clean; \
+ cd ../$(LINALG_BLIS_DIR); ./configure -p install/arm cortex-a15; make clean; \
+ cd ../$(LINALG_BLASACC_DIR); make clean; \
+ cd ../$(LINALG_BLIS_DIR)/testsuite; make clean; \
+ cd ../../$(LINALG_CLAPACK_DIR); make clean
#DSPonly:
BLIStest:
+ cd ../$(LINALG_BLIS_DIR); ./configure -p install/arm cortex-a15; \
cd $(LINALG_BLIS_DIR)/testsuite; make lib=OpenCLCBLAS -j8
BLAStest:
cd $(LINALG_CLAPACK_DIR)/TESTING/EIG; make
-cleanARMplusDSP:
- cd $(LINALG_CBLAS_DIR); make arch=ARM clean; make arch=C66 clean; \
- cd ../$(LINALG_TICBLAS_DIR)/src; make clean; cd ..; \
- cd ../$(LINALG_BLIS_DIR); ./configure -p install/Large c66x; make clean; \
- ./configure -p install/arm cortex-a15; make clean; \
- cd ../$(LINALG_BLASACC_DIR); make clean; \
- cd ../$(LINALG_BLIS_DIR)/testsuite; make clean; \
- cd ../../$(LINALG_CLAPACK_DIR); make clean
-
-cleanARMplusDSP_AM57x:
- cd $(LINALG_CBLAS_DIR); make arch=ARM clean; make arch=C66 clean; \
- cd ../$(LINALG_TICBLAS_DIR)/src; make clean; cd ..; \
- cd ../$(LINALG_BLIS_DIR); ./configure -p install/am57x am57x; make clean; \
- ./configure -p install/arm cortex-a15; make clean; \
- cd ../$(LINALG_BLASACC_DIR); make clean; \
- cd ../$(LINALG_BLIS_DIR)/testsuite; make clean; \
- cd ../../$(LINALG_CLAPACK_DIR); make clean
+cleanARMplusDSP: cleanDSPlibs cleanARMlibs
-cleanShannon:
- cd $(LINALG_CBLAS_DIR); make arch=ARM clean; make arch=C66 clean; \
- cd ../$(LINALG_TICBLAS_DIR)/src; make clean; cd ..; \
- cd ../$(LINALG_BLIS_DIR); ./configure -p install/shannon shannon; make clean; \
- ./configure -p install/arm cortex-a15; make clean; \
- cd ../$(LINALG_BLASACC_DIR); make clean; \
- cd ../$(LINALG_BLIS_DIR)/testsuite; make clean; \
- cd ../../$(LINALG_CLAPACK_DIR); make clean
clean:
cd $(LINALG_CBLAS_DIR)/src; make arch=ARM clean; \
cd ../$(LINALG_BLASACC_DIR)/src; make -f Makefile.ARM cleanARM; \
cd ../../$(LINALG_CLAPACK_DIR); make clean
-DSPportion:
- cd $(LINALG_CBLAS_DIR); make arch=C66 alllib; \
- cd ../$(LINALG_TICBLAS_DIR)/src; make; cd ..; \
- cd ../$(LINALG_BLIS_DIR); ./configure -p install/c66x c66x; make -j8; make install; \
- cd ../$(LINALG_BLASACC_DIR)/src; make ti_cblas_kernel.dsp_h
-
install:
install -m 755 -d ${DESTDIR}/usr/include
install -m 755 -d ${DESTDIR}/usr/lib
cp $(CBLAS_HEADERS) ${DESTDIR}/include
cp ./lib/libcblas.ae66 ${DESTDIR}/lib
cp -r docs ${DESTDIR}
+
+installARMplusDSPlib:
+ install -m 755 -d ${DESTDIR}/include
+ install -m 755 -d ${DESTDIR}/lib
+ cp $(CBLAS_HEADERS) ${DESTDIR}/include
+ cp $(CLAPACK_HEADERS) ${DESTDIR}/include
+ cp ./lib/libblis.a ${DESTDIR}/lib
+ cp ./lib/libcblas_armplusdsp.a ${DESTDIR}/lib
+ cp ./lib/liblapack.a ${DESTDIR}/lib
+ cp ./lib/libcblaswr.a ${DESTDIR}/lib
+ cp ./lib/libf2c.a ${DESTDIR}/lib
+ cp -r docs ${DESTDIR}
+
index 0ce5362734499937cbf72736f2bf9b77f796f630..0d606b338600ea77eea9ce0f73790e10eafcfd89 100644 (file)
--- a/blasblisacc/src/Makefile
+++ b/blasblisacc/src/Makefile
INCDIR += -I$(XDC_DIR)/packages
INCDIR += -I$(BIOS_DIR)/packages
INCDIR += -I$(XDAIS_DIR)/packages
-INCDIR += -I$(LIBARCH_DIR)
+INCDIR += -I$(LIBARCH_DIR)/include
INCDIR += -I$(PDK_DIR)/packages
INCDIR += -I$(TI_OCL_INSTALL_DIR)
INCS = -I. -I$(strip $(subst ;, -I,$(subst $(space),$(space),$(INCDIR))))
OBJS = ti_cblas_initfini.o
+# CBLAS and BLIS directories
+CBLAS_DSP_LIB = ../../cblas/lib/C66/libcblas_C66.ae66
+TICBLAS_DSP_LIB = ../../ticblas/lib/libticblas.ae66
+CBLAS_ARM_LIB = ../../cblas/lib/ARM/libcblas_ARM.a
+LIBARCH_LIB = $(LIBARCH_DIR)/lib/libArch.ae66
+
+ifeq ($(MEM_MODEL),Large)
+BLIS_DSP_LIB = ../../blis/install/c66xLarge/lib/libblis.ae66
+else ifeq ($(MEM_MODEL),Medium)
+BLIS_DSP_LIB = ../../blis/install/c66xMedium/lib/libblis.ae66
+else ifeq ($(MEM_MODEL),Small)
+BLIS_DSP_LIB = ../../blis/install/c66xSmall/lib/libblis.ae66
+#else ifeq ($(MEM_MODEL),Tiny)
+endif
CPP_DEBUG = -g
CPP_FLAGS = -D_LITTLE_ENDIAN -D__ARMv7 -D$(TARGET) -I../../cblas/include -I../../blis/install/arm/include/blis/ -I$(TI_OCL_INSTALL_DIR)/include -fopenmp
index 9e0437243be0b063a9e2bd18132e935f93d6b3cd..a4e8255c1be7b6fd9da86d6bd9b7ead1405007a9 100644 (file)
OBJCOPY_ARGS=
ARM_PLUS_DSP_LIB= $(ARM_PLUS_DSP_LIB_DIR)/libcblas_armplusdsp.a
-# CBLAS and BLIS directories
-CBLAS_DSP_LIB = ../../cblas/lib/C66/libcblas_C66.ae66
-BLIS_DSP_LIB = ../../blis/install/c66x/lib/libblis.ae66
-CBLAS_ARM_LIB = ../../cblas/lib/ARM/libcblas_ARM.a
-
OCL_BIN = ti_cblas_kernel.out
ifeq ($(TI_CBLAS_FAT_BINARY), 1)
index 0900fd7991afd6e0090b9257aa89aaf18496cdc9..fdea549af1fc8564b93fe57071c73b12b391c815 100644 (file)
#ifdef __cplusplus
-extern Kernel* ti_cblas_get_kernel(const char *fname);
+extern Kernel* ti_cblas_get_kernel(int idx, const char *fname);
int ti_cblas_delete_kernel(Kernel* K);
#if 0
extern Context ti_cblas_ocl_context;
extern std::vector<Device>* ti_cblas_ocl_devices;
extern CommandQueue* ti_cblas_ocl_Q;
extern Program::Binaries* ti_cblas_ocl_binary;
-//extern Program* ti_cblas_ocl_program;
-extern Program ti_cblas_ocl_program;
+extern Program* ti_cblas_ocl_program;
#endif
#else
extern cl_kernel ti_cblas_get_kernel(int idx, const char *fname);
index 8011b2b2b828bc4a4ed9419aea1957a1f45a5977..39daf708f3665a6dbc3a3b8f36e29e49d8fe4b05 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_caxpy");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CAXPY_IDX, "ocl_cblas_caxpy");
#ifdef __cplusplus
try
#else
index 8a8fbc11350d48623dd8f21214dfc35040c79c68..bf5b16d0cf2e88d088f1ca470a6db33a7715654c 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ccopy");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CCOPY_IDX, "ocl_cblas_ccopy");
#ifdef __cplusplus
try
#else
diff --git a/blasblisacc/src/ti_cblas_cblas_cdotc_sub.c b/blasblisacc/src/ti_cblas_cblas_cdotc_sub.c
index 233d143908e60061ffc8d11848e364ca6fc13e7e..9cfc775883c857f2d934a622ec5e87f03a48b4e5 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cdotc_sub");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CDOTC_SUB_IDX, "ocl_cblas_cdotc_sub");
#ifdef __cplusplus
try
#else
diff --git a/blasblisacc/src/ti_cblas_cblas_cdotu_sub.c b/blasblisacc/src/ti_cblas_cblas_cdotu_sub.c
index 20ad6d0a87a8611e81447ee7ade1862050431248..68d4684a1f3c8c8fc98c3dbbdaffc1a928f858d8 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cdotu_sub");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CDOTU_SUB_IDX, "ocl_cblas_cdotu_sub");
#ifdef __cplusplus
try
#else
index 9f2a60470db81b7739e62f34355d40e1de4dd185..097792c134ec24685a50ddada1299d4db27116b9 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cgbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CGBMV_IDX, "ocl_cblas_cgbmv");
#ifdef __cplusplus
try
#else
index 947f84cf0bf945225777dc7a469b4284ed9e7cc7..ffd982072e7e077bbfc186ae7790be1924e3b991 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cgemm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CGEMM_IDX, "ocl_cblas_cgemm");
#ifdef __cplusplus
try
#else
index 6f807ff768429839e5cb6272ee28f66f66d77edc..37962f7f3c59c8c5a2e7c6a41ca031cae843ff96 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cgemv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CGEMV_IDX, "ocl_cblas_cgemv");
#ifdef __cplusplus
try
#else
index 4da0a7708f7bb37d7a47980db4a864bc4d58c60e..e8feac03ac0126dda9c8148cedc56b5803b19d91 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cgerc");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CGERC_IDX, "ocl_cblas_cgerc");
#ifdef __cplusplus
try
#else
index fea0731c7f39268b8e59e2b609c2e6154936ac3a..22f8c5046a527a67d8416d5a3af12cb6264efb93 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cgeru");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CGERU_IDX, "ocl_cblas_cgeru");
#ifdef __cplusplus
try
#else
index 66a36bcbd96737486329dc569c84a880b4ab7d2d..ad7e51bdec8af637d8bd560db330cb99e1aeeafd 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_chbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHBMV_IDX, "ocl_cblas_chbmv");
#ifdef __cplusplus
try
#else
index c61eb41bb7c9072c6af634e1d0b9e3fcfd215172..b51fbc2953d5f3d445173f2d40f7d25fa29d9bba 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_chemm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHEMM_IDX, "ocl_cblas_chemm");
#ifdef __cplusplus
try
#else
index 97bf4a25f613b40add4cb1b0a8aac9eba9f1243e..8e076486e7069d395ba15c860ff9917db4051fdb 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_chemv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHEMV_IDX, "ocl_cblas_chemv");
#ifdef __cplusplus
try
#else
index 2183e42b910c4a69f01d18edd384ed0f6c25c84d..b3ff8c31897c073178be5e85ac9c054a3317081d 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cher");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHER_IDX, "ocl_cblas_cher");
#ifdef __cplusplus
try
#else
index a340d3d12516245f3e4c690f5e58bb940a0eb422..eced71f9e068f36310ff13125f347bd154d08c26 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cher2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHER2_IDX, "ocl_cblas_cher2");
#ifdef __cplusplus
try
#else
index 0bebac137b5a6317196bbc29c9b146223c69dd6d..709f036129cad062b6d46434b78fcdadfe3e56ce 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cher2k");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHER2K_IDX, "ocl_cblas_cher2k");
#ifdef __cplusplus
try
#else
index 2a3a5f46ae1668ac22cbe70af16421efcd51f528..d7db80b6a597cae2391e8fcd4495ea3bf1a14591 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cherk");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHERK_IDX, "ocl_cblas_cherk");
#ifdef __cplusplus
try
#else
index c48d2415f7cb3272a86868daf9a4be1c60b3007e..c4d1d624df19ef648ac0bbf0b8b772bab84b1d80 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_chpmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHPMV_IDX, "ocl_cblas_chpmv");
#ifdef __cplusplus
try
#else
index d52a6261722c7399021607f89f604d07b90e384a..670016e0fd00928f84d311257bea466c629b0814 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_chpr");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHPR_IDX, "ocl_cblas_chpr");
#ifdef __cplusplus
try
#else
index 75132be99884b3c412e1f04bbdffda51f7e724b9..50d29f732e25eca733394abcd054e34dbf907618 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_chpr2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CHPR2_IDX, "ocl_cblas_chpr2");
#ifdef __cplusplus
try
#else
index af7fa3907bc26784723110b83d690bded7ba8fc6..c31ba615e8be56b3463efe8bb00678c3505c5acd 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_crotg");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CROTG_IDX, "ocl_cblas_crotg");
#ifdef __cplusplus
try
#else
index 9a7b1340ddd4a9198872ab99fafafcdd593fae2a..40d2452480b253c43d3c1a60ff12e77b3a126ed8 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cscal");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CSCAL_IDX, "ocl_cblas_cscal");
#ifdef __cplusplus
try
#else
index 3f9b51d35ca2a2cffa465ad04128860b697b97cd..aea68b46d1e5e28d9e266a7d500aea7b528cba77 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_csscal");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CSSCAL_IDX, "ocl_cblas_csscal");
#ifdef __cplusplus
try
#else
index 7e5039187efdfe2dcb16c53df2b5894e8f0612a1..67249d32b0156aa9aeecfc6a3d62c43a8716a624 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_cswap");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CSWAP_IDX, "ocl_cblas_cswap");
#ifdef __cplusplus
try
#else
index 494b499352cf37d186dc2b17890fc10db72b5b95..6050de720b21b53d987867bbff0c1bbbc2367bc9 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_csymm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CSYMM_IDX, "ocl_cblas_csymm");
#ifdef __cplusplus
try
#else
index de62fc1ace0d31693a7a86f3a037ca951ac7864f..11feff3f8ef46768d433addd1dd61f499f72b6d8 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_csyr2k");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CSYR2K_IDX, "ocl_cblas_csyr2k");
#ifdef __cplusplus
try
#else
index 1dd65d4c114dbb2eda9944b72351a820fe7d5a78..4840abdfc8efb414bd75151e7214fdb479d913a2 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_csyrk");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CSYRK_IDX, "ocl_cblas_csyrk");
#ifdef __cplusplus
try
#else
index e17c4711e27b780e0bf1a6be1271018daff95ed4..018b07030dd14955cf541b6d1c35166008f06e4f 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ctbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CTBMV_IDX, "ocl_cblas_ctbmv");
#ifdef __cplusplus
try
#else
index 7d76fd8671415bfcf6361e75d82bed956cdeba77..6f34e3ae8c2e72642a732000e49845bb774f659f 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ctbsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CTBSV_IDX, "ocl_cblas_ctbsv");
#ifdef __cplusplus
try
#else
index abc37d0957bbbedb9465f0ea35b382191b95c265..b2db324062eeb3f301dce6e89f8590c78ac0d142 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ctpmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CTPMV_IDX, "ocl_cblas_ctpmv");
#ifdef __cplusplus
try
#else
index f1242bc5764967dca48ac544a58e3bd34f77a574..45969d33cd990209ad89cd6d0ce64ce8d329ae6a 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ctpsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CTPSV_IDX, "ocl_cblas_ctpsv");
#ifdef __cplusplus
try
#else
index 33ba1462c30730081750ee0657046291ee295e66..54f291f3ada4bb76599a2e80e7b9145d67cb1d77 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ctrmm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CTRMM_IDX, "ocl_cblas_ctrmm");
#ifdef __cplusplus
try
#else
index 4b22df49c0ef0d6adb8ace36166053715fb155ca..40054f941ce48333e91fa9af8267adee3d5a974e 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ctrmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CTRMV_IDX, "ocl_cblas_ctrmv");
#ifdef __cplusplus
try
#else
index 05882ea46c030e4d91007cb5f8f164253211810d..7804059c0351970f512ab4f496f721fc3faee424 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ctrsm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CTRSM_IDX, "ocl_cblas_ctrsm");
#ifdef __cplusplus
try
#else
index 4f636566c82ba65c9628bc2559eb4c78200cad0b..42d5a05d2ecf7c7099ab595206cba546270f8dd4 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ctrsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CTRSV_IDX, "ocl_cblas_ctrsv");
#ifdef __cplusplus
try
#else
index 80ba1428076a93bfcb05235ed2a1b704c873eb21..ac2c2af1905e9862621ec3c14007021e10283869 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dasum");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DASUM_IDX, "ocl_cblas_dasum");
#ifdef __cplusplus
try
#else
index 7c561bb21875719b3015fd4bdf031807f5963ebd..fee688edb32d0aa4cd9050c42644b26076cacfea 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_daxpy");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DAXPY_IDX, "ocl_cblas_daxpy");
#ifdef __cplusplus
try
#else
index 403fedbc7de5cb8e45c71cb4698b6b5755ab1c0e..58769ff7228ed96b15365979b6fe02f303876ff5 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dcopy");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DCOPY_IDX, "ocl_cblas_dcopy");
#ifdef __cplusplus
try
#else
index 24469f64c35b5a529b6b4d8fe226301bc0d597a2..5bfbb81b112d1ba164fd7d6361eb42c570627b36 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ddot");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DDOT_IDX, "ocl_cblas_ddot");
#ifdef __cplusplus
try
#else
index c3371a00c2dd5a4705f2ae0676536f7d55cd7e03..2e03fd2241bb5a78b5798c7e475abae99e7dcdc6 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dgbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DGBMV_IDX, "ocl_cblas_dgbmv");
#ifdef __cplusplus
try
#else
index f670d9b7916635ddac8962e453183e798e0fa678..6fa53241eb386e9229f7c1261df1b2a931ca3ea5 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dgemm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DGEMM_IDX, "ocl_cblas_dgemm");
#ifdef __cplusplus
try
#else
index 461afe198af0b22cc150ef152d1b504eb74700d4..acb7123ff30729f4f966d080dcc8d6c28c3703fe 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dgemv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DGEMV_IDX, "ocl_cblas_dgemv");
#ifdef __cplusplus
try
#else
index 5a8f94254fc6b19a3b9a5665ecd5157fa796e47e..c035efed70e6150e651419c058d8402616b3ae47 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dger");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DGER_IDX, "ocl_cblas_dger");
#ifdef __cplusplus
try
#else
index 70489bd6409f146a25f6d6e60b7a242f4e86ff59..0be91b895ee4701bff3e183c13da99a63203b15b 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dnrm2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DNRM2_IDX, "ocl_cblas_dnrm2");
#ifdef __cplusplus
try
#else
index 0c5cfa15292eab479df109e31494f8fa34472605..bb4cdc07730ba5aa5dd0c901b36207725d6c2974 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_drot");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DROT_IDX, "ocl_cblas_drot");
#ifdef __cplusplus
try
#else
index f375f4360c026a86e75a304084fa896ae3ce501b..2c20ae308e8754752271ccfd84f3e27d6a9802fe 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_drotg");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DROTG_IDX, "ocl_cblas_drotg");
#ifdef __cplusplus
try
#else
index 39c6a80137ebe952fc1fde0706976bdda2232b36..1e4a374ea085398c56ccd279b9cb5fe02999a472 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_drotm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DROTM_IDX, "ocl_cblas_drotm");
#ifdef __cplusplus
try
#else
index c7d101d7d74fb4a72feaabdd3559dba792765e40..ef4e979a576a1754a26a4a629bb3599c1383efd2 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_drotmg");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DROTMG_IDX, "ocl_cblas_drotmg");
#ifdef __cplusplus
try
#else
index bf4c8e6498b6c648be69080a7c8a510162bc8158..d3e1988c0837a5dd87065b12660350a6b78da736 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dsbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSBMV_IDX, "ocl_cblas_dsbmv");
#ifdef __cplusplus
try
#else
index 3e1bc031675da44de6345e5880919e42402c2e2c..44c1510a00a0b7f831be31e3a06e5b495e66d70e 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dscal");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSCAL_IDX, "ocl_cblas_dscal");
#ifdef __cplusplus
try
#else
index fba396937344656a8fed5fbac478b1c0ec84f32f..7e7ff095c7f714b7e75e9c23f4b8226f555adf11 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dsdot");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSDOT_IDX, "ocl_cblas_dsdot");
#ifdef __cplusplus
try
#else
index 8f64ca20db4b4d6dec0be822fad4e9d694ca5a9e..5239c73b879d30a40851a5f46ae1a04e33faadde 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dspmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSPMV_IDX, "ocl_cblas_dspmv");
#ifdef __cplusplus
try
#else
index 0c7de8123d8bed73e8a183ecdc4e19c8e1acca4b..d8c25dcf49eabb891204b2ab0be2378d6c88655a 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dspr");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSPR_IDX, "ocl_cblas_dspr");
#ifdef __cplusplus
try
#else
index e8c0289ccd5cc18f321abcd7761f9843fae3d7f5..4d6b2c049debc0b14516beb402ec8e1f3d0593c6 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dspr2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSPR2_IDX, "ocl_cblas_dspr2");
#ifdef __cplusplus
try
#else
index 9036f3187f471da2d4ebd0d3cf219a532b4b6152..63cd15ada38181b6b5f14825263bb73857119fc4 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dswap");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSWAP_IDX, "ocl_cblas_dswap");
#ifdef __cplusplus
try
#else
index e74a029be068afed928029258e6c269e64c0d240..a6223ee92349bc7f723adac969d7e6ee376d978f 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dsymm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSYMM_IDX, "ocl_cblas_dsymm");
#ifdef __cplusplus
try
#else
index 9f60c6fee84c1152d0a5499e67fd19f8637a1c0d..aca46f23ecadab0c668bac4510662baa53d1bddf 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dsymv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSYMV_IDX, "ocl_cblas_dsymv");
#ifdef __cplusplus
try
#else
index b85a76a3d9a8a3a43f1b26ea86fbc931b5cb18d2..a421bcde70a15430b00f875af77980e9f38c55cb 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dsyr");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSYR_IDX, "ocl_cblas_dsyr");
#ifdef __cplusplus
try
#else
index 4a0143adbffff72927c02ff62012610f8123285c..5c64af929d45692dcaf68c44065e18404802bdde 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dsyr2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSYR2_IDX, "ocl_cblas_dsyr2");
#ifdef __cplusplus
try
#else
index 043b00ce7a5400aa5f683997d9b2ffd4690ef65c..5e126a700094fd74be18c69c66c1ba50d9ca3a60 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dsyr2k");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSYR2K_IDX, "ocl_cblas_dsyr2k");
#ifdef __cplusplus
try
#else
index b152fbcd511f40c1ebf2143922c46811b8220761..a5b7d87ef9aa0d5f2d4086bbc0eb0806cd617995 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dsyrk");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DSYRK_IDX, "ocl_cblas_dsyrk");
#ifdef __cplusplus
try
#else
index b14bf8b4821de9308dc8617495d61ff4909cc7f1..8e1149847c15a490d37eb72c30e2cacd0ca4d89d 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dtbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DTBMV_IDX, "ocl_cblas_dtbmv");
#ifdef __cplusplus
try
#else
index b81f39430128bf4a61199c8e725186f20ff59ccf..095284708aea09ac3eaeb5c3a8e1d6a102301f5e 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dtbsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DTBSV_IDX, "ocl_cblas_dtbsv");
#ifdef __cplusplus
try
#else
index c0839c7066a109c8b0827d301d56be7592076029..f8fa7fcde306160ef9e6620d8d2fd2a912ea12a4 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dtpmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DTPMV_IDX, "ocl_cblas_dtpmv");
#ifdef __cplusplus
try
#else
index 4a07438f7c04161b99ef9f7dbea3b7e63497cc2e..6856032f5357329ac0a27df39ac90dd24b561714 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dtpsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DTPSV_IDX, "ocl_cblas_dtpsv");
#ifdef __cplusplus
try
#else
index 910a928db1840bc7fcf5d9bb624c96c6c1505bd9..fd70c5dc4b706e5293b03e5920f4033319752ffb 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dtrmm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DTRMM_IDX, "ocl_cblas_dtrmm");
#ifdef __cplusplus
try
#else
index 0da3b25a4b8afac3731b99074d8130c1f468c954..54755ec2ee5b645002cb2dce5ac2e7f5da8069ea 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dtrmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DTRMV_IDX, "ocl_cblas_dtrmv");
#ifdef __cplusplus
try
#else
index d74cad1cfa6863291a93e599a41a0c2fdec07b79..798a88e7db5ae6895105b8877d2809c8ff3eaaea 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dtrsm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DTRSM_IDX, "ocl_cblas_dtrsm");
#ifdef __cplusplus
try
#else
index 53f642eab5c73e1ea487639f69dabad5a76e924f..15d492b19ef6d54d96ad3cd9040e2c840d5120ef 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dtrsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DTRSV_IDX, "ocl_cblas_dtrsv");
#ifdef __cplusplus
try
#else
index 956ae211b4e5a904538fc6be34bbd702056a013b..fd4874ce4c76eb0d6fbdab0ea5b4defb17b9bdb2 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dzasum");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DZASUM_IDX, "ocl_cblas_dzasum");
#ifdef __cplusplus
try
#else
index 3b25d2f1d5f452dcfa17cd6788fb1c3395154e78..96147f96cfc07c8bd203d7188e4540b8d413e4a3 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_dznrm2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_DZNRM2_IDX, "ocl_cblas_dznrm2");
#ifdef __cplusplus
try
#else
index 616a92d1cd342172b4c14378f0671254d9806838..c5569ad9e9a73e65cdd22399c849f2fb11e96ef8 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_icamax");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ICAMAX_IDX, "ocl_cblas_icamax");
#ifdef __cplusplus
try
#else
index 073906e7e058598e6956db5dd15c854cfa3b1fdb..3ed7745801849fcb57d95545c06b9593633c0563 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_idamax");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_IDAMAX_IDX, "ocl_cblas_idamax");
#ifdef __cplusplus
try
#else
index bc0cb8dc49a996fae44214611d161e23a3106a9e..19bea6ab6ddd0bd78d74e478b220f1b8c00d7c11 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_isamax");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ISAMAX_IDX, "ocl_cblas_isamax");
#ifdef __cplusplus
try
#else
index cb9c43457b54596fea6977d37fc0083826b725a7..77141cf81d0e2fb1d2ee61287eebbd0a2d1f7aed 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_izamax");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_IZAMAX_IDX, "ocl_cblas_izamax");
#ifdef __cplusplus
try
#else
index b78d76e330e8d3634ebc95a756120dcb2448b82b..7892bb7221f69ac12fbeb16eaa005d4954fb8af8 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sasum");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SASUM_IDX, "ocl_cblas_sasum");
#ifdef __cplusplus
try
#else
index 9ab8bb8a4a2dc21ecbd2d969eee333366ab156c5..3862933e632a4fd52a7255eebf01593ed6c3c88f 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_saxpy");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SAXPY_IDX, "ocl_cblas_saxpy");
#ifdef __cplusplus
try
#else
index 26a02b2523cd24e52623d295f3c9c9b307c38a73..23c6f4f38c27cc687ed0892dd73449ef80b718bd 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_scasum");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SCASUM_IDX, "ocl_cblas_scasum");
#ifdef __cplusplus
try
#else
index b9b52856d6ab0cb27fdfe6e72cb46f744f55dc01..a966e9ee6950f86c5e80f883be581ca7c3a1fc8f 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_scnrm2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SCNRM2_IDX, "ocl_cblas_scnrm2");
#ifdef __cplusplus
try
#else
index 9dad1cefe055e6b95db2729bb20eea73b8266228..dd6cab58259f0aff629480c062babebdaaad0ea3 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_scopy");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SCOPY_IDX, "ocl_cblas_scopy");
#ifdef __cplusplus
try
#else
index f7d0ef1c4e936de0b62fd14cae2d50af475320d5..69fd6d44bb21eb4c8da40cf7f4c35e76dfb19e74 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sdot");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SDOT_IDX, "ocl_cblas_sdot");
#ifdef __cplusplus
try
#else
index 14fb7872aceb0caf39a508b90c3cb5d58b04dbe4..d5445ea42bd48a1e64bc44e8e22f300d7fe89b9e 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sdsdot");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SDSDOT_IDX, "ocl_cblas_sdsdot");
#ifdef __cplusplus
try
#else
index 4f45e281b0b1db509e4ea2028b266c29e6f1e0da..f4f2826d101e2f19550cc062d299bc315d784ddb 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sgbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SGBMV_IDX, "ocl_cblas_sgbmv");
#ifdef __cplusplus
try
#else
index ec4c413e9b81863b340de7f2318f660035e894ae..72e3af3b58ab693ccfc9b31001d4f5d554b8db58 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sgemm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SGEMM_IDX, "ocl_cblas_sgemm");
#ifdef __cplusplus
try
#else
index 49d85bef5607cd7bdc2f840e5793433682ff2e4f..c6e0c1daf7faf8f5b7166676f291dadbeb1fa048 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sgemv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SGEMV_IDX, "ocl_cblas_sgemv");
#ifdef __cplusplus
try
#else
index 54be54a410eab493334b7c2c502a2500742fe02c..1ee795f876353a1adbbfc44a0be97e0f50e1c92c 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sger");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SGER_IDX, "ocl_cblas_sger");
#ifdef __cplusplus
try
#else
index 6fc9b4289d90453e094d88f24c0e53b6e90357e3..79a7dcbfb6da15956a44b1bb3eabbf0ca453d710 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_snrm2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SNRM2_IDX, "ocl_cblas_snrm2");
#ifdef __cplusplus
try
#else
index da897ccdb84e09472824553535d1f0cd68d7e6e6..051e99a68f1e988d49bf33412e31a652ff75826e 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_srot");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SROT_IDX, "ocl_cblas_srot");
#ifdef __cplusplus
try
#else
index 4fb2a93475e451e25aaf65b75cbcafad07c8fb4f..2b7a071bb9a248c351b5b0ee5f4397d599c3d94a 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_srotg");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SROTG_IDX, "ocl_cblas_srotg");
#ifdef __cplusplus
try
#else
index eb48711bf7fc4e0e9fbf6d6e1820303bc57b49d0..e67142c90213ae74fc2522d00b4a4381099d6b68 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_srotm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SROTM_IDX, "ocl_cblas_srotm");
#ifdef __cplusplus
try
#else
index f33cbf29537559482d6bcbf0a37dbb1f08a6c494..008e2d0a35a5878f61dd8cef1cede20cc1b58c63 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_srotmg");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SROTMG_IDX, "ocl_cblas_srotmg");
#ifdef __cplusplus
try
#else
index b1aa3b07cf82ea03670c49ba0f3bb4268fe92901..08d9dd7479b36b0172ba457a90ca089824ca9cc5 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ssbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSBMV_IDX, "ocl_cblas_ssbmv");
#ifdef __cplusplus
try
#else
index baeffe56e24dd4d5c955edd7c7bba5c52ba5437c..d3afdb9af57c433e646f69fb153bc69b3d2d4169 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sscal");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSCAL_IDX, "ocl_cblas_sscal");
#ifdef __cplusplus
try
#else
index 3fd769adb98206c0709bc56bae841a0efc90aa40..78f82ddba2e69a90b0e49dd5ad1083ad8bd22ca0 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sspmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSPMV_IDX, "ocl_cblas_sspmv");
#ifdef __cplusplus
try
#else
index fd767b2e86fb70e59f7b55e390cae1e2bb7cbf4d..707533f1203af0ff8e44dd7565d2540094f4f931 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sspr");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSPR_IDX, "ocl_cblas_sspr");
#ifdef __cplusplus
try
#else
index aa7875ad33d3046c8a7454b0f10769d9114426a9..dac6d8538f1a8aa11aab90484f408d19893fa50e 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sspr2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSPR2_IDX, "ocl_cblas_sspr2");
#ifdef __cplusplus
try
#else
index cfb6cae569995d21495a060ff8a562e6a510012d..241d2136298fbee4b1b8061244d8a306bc031112 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_sswap");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSWAP_IDX, "ocl_cblas_sswap");
#ifdef __cplusplus
try
#else
index 68b45bf71224eaa2bb9c1c9ae7198f672b68e75a..c090bed21bdefa2027be33820dfda2ab4d1586b3 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ssymm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSYMM_IDX, "ocl_cblas_ssymm");
#ifdef __cplusplus
try
#else
index 39cdcee292dbdd988736ec932685ec9734c531b4..e92214f0d9b4b9f67c938c86df235519614774b8 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ssymv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSYMV_IDX, "ocl_cblas_ssymv");
#ifdef __cplusplus
try
#else
index efc228b697dcdb80fd1a33b3fbd9b36a93cb91c6..5c45b02b42ccab1327a9fdcabae28d5d7eb0e8a5 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ssyr");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSYR_IDX, "ocl_cblas_ssyr");
#ifdef __cplusplus
try
#else
index 9218c3b8fad77f7b563ed561968ad1ab24531915..ea04df9a24f991f193fdd6f7b9cb887f2341c97d 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ssyr2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSYR2_IDX, "ocl_cblas_ssyr2");
#ifdef __cplusplus
try
#else
index 9613eb5eb594b47bec739a2dc689216e4979084d..97b6b443b80c33ae0988494f948b559cd83c8647 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ssyr2k");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSYR2K_IDX, "ocl_cblas_ssyr2k");
#ifdef __cplusplus
try
#else
index d7296d24c0768513b12059f05a87018129cbefae..17c8bbddeeaee3871f6cb208fc49eb4d73a5a18f 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ssyrk");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_SSYRK_IDX, "ocl_cblas_ssyrk");
#ifdef __cplusplus
try
#else
index f7041aaf3323f5898e10afdb5067e590c7c74835..97b7922d647442cdc1f94eaa8d67aba92cf2cb2d 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_stbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_STBMV_IDX, "ocl_cblas_stbmv");
#ifdef __cplusplus
try
#else
index a7e586819a667bb91babb4a68385e453611a8297..47002da6eaa00838607cb2a56bac48e161455d86 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_stbsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_STBSV_IDX, "ocl_cblas_stbsv");
#ifdef __cplusplus
try
#else
index a9cf8363538bd8597e9b23a498fed79179d0bc17..a5ea0271c4383d0167bb98c8dd5c7a70bbd8a847 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_stpmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_STPMV_IDX, "ocl_cblas_stpmv");
#ifdef __cplusplus
try
#else
index cc419808b2d301ec5430e442133edd6fd6b0fe31..0b23825a29b71d1927177aff2dee0b1b14bc7b06 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_stpsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_STPSV_IDX, "ocl_cblas_stpsv");
#ifdef __cplusplus
try
#else
index 14384df78dd16a72e1200e93ba0edda3ca41b1ca..fc1d9d82400c27ccc8adb3423f5576cd573e6bc5 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_strmm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_STRMM_IDX, "ocl_cblas_strmm");
#ifdef __cplusplus
try
#else
index b9bbd6bd43bd7c8ee987598f206e7513e9234f2a..933bd345d9e178504416da29d1234f01d227607f 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_strmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_STRMV_IDX, "ocl_cblas_strmv");
#ifdef __cplusplus
try
#else
index 8e63aeeef5121a6ae84df8419ab0312cbc13227a..56ad072c622be0bb7b1489001e296e77c4d18676 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_strsm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_STRSM_IDX, "ocl_cblas_strsm");
#ifdef __cplusplus
try
#else
index 4855e1c1974ace9311b02cd1583b8c1869961b3b..97aed05220eae5f502c1c6f462e67caca929906f 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_strsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_STRSV_IDX, "ocl_cblas_strsv");
#ifdef __cplusplus
try
#else
index 87a37da1d4894efbdf8bffb371f11d910c9f4b65..2ff97c8e9002195a31690106b57c6741a9bc578a 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_xerbla");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_XERBLA_IDX, "ocl_cblas_xerbla");
#ifdef __cplusplus
try
#else
index 134e612f92f8fdbc8efd1499367689fddcb6f741..8450fae6dddcad9b73926122c1d31c307607af9b 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zaxpy");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZAXPY_IDX, "ocl_cblas_zaxpy");
#ifdef __cplusplus
try
#else
index 2b2bb9d007edcd14d6873e5cd738ec69d31b9240..60edee0b1a1dcd0898b4e4e537e0f1881ad22f77 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zcopy");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZCOPY_IDX, "ocl_cblas_zcopy");
#ifdef __cplusplus
try
#else
diff --git a/blasblisacc/src/ti_cblas_cblas_zdotc_sub.c b/blasblisacc/src/ti_cblas_cblas_zdotc_sub.c
index 98aa5505058bd1efc2f5f5d9580f4ab21afa128c..80cd4d8d19fa29ed864e90aff61fb3bf55901008 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zdotc_sub");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZDOTC_SUB_IDX, "ocl_cblas_zdotc_sub");
#ifdef __cplusplus
try
#else
diff --git a/blasblisacc/src/ti_cblas_cblas_zdotu_sub.c b/blasblisacc/src/ti_cblas_cblas_zdotu_sub.c
index 5058183de23498a6ea056d90ddd516ef9c8b9b50..841cf7103056e3085d99ca635e23c0ac9ab41396 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zdotu_sub");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZDOTU_SUB_IDX, "ocl_cblas_zdotu_sub");
#ifdef __cplusplus
try
#else
index f35b69d92f24df99c572833b587b89b3856064fe..d2c7eca168abcbea0f24d50e003b9cd0ad42de62 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zdscal");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZDSCAL_IDX, "ocl_cblas_zdscal");
#ifdef __cplusplus
try
#else
index bc6b0d9ba6a0a3db4db578362217c146c04dc707..0a2a0740e8ce2892c5a71309b72157bcc707ac89 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zgbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZGBMV_IDX, "ocl_cblas_zgbmv");
#ifdef __cplusplus
try
#else
index eb8ce004f2b64bd739e80fa0e7b1ec0c064d7e39..77e5bf914df46ef5c0d22898d3b5988180284cd8 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zgemm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZGEMM_IDX, "ocl_cblas_zgemm");
#ifdef __cplusplus
try
#else
index 63dc74fbaeaeee368d0570e5195c7655ec46c729..a195377e20dbc6f6f0257e60ef63d0ba9797b168 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zgemv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZGEMV_IDX, "ocl_cblas_zgemv");
#ifdef __cplusplus
try
#else
index a51a0016ae1fa63f799b0cb9911d082478b240b1..341c8f1fe67c11eb94930b8a0fada68781c109de 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zgerc");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZGERC_IDX, "ocl_cblas_zgerc");
#ifdef __cplusplus
try
#else
index 53aa086df4f9e6c496578e493d517810883afbfc..94cddea2813c7f4cb1fcbfa312f41b51984b9b2c 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zgeru");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZGERU_IDX, "ocl_cblas_zgeru");
#ifdef __cplusplus
try
#else
index 8402a6fdb5cb1028408b229849b7a0937b11f143..36fa157152bb4caad5c34d72e0c86987e848f27a 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zhbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHBMV_IDX, "ocl_cblas_zhbmv");
#ifdef __cplusplus
try
#else
index 1a1cb8b40c97b8a1631dcea08dbc3ce16d571fd8..d206e1361fdcb13fe6151a4ac8b82dfc7384a434 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zhemm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHEMM_IDX, "ocl_cblas_zhemm");
#ifdef __cplusplus
try
#else
index b86bd760c75e00f40a56f6f71e3f73e56bcb44ac..6e87053d010dd9a82dc474db8a40b02f62c2a1d8 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zhemv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHEMV_IDX, "ocl_cblas_zhemv");
#ifdef __cplusplus
try
#else
index b26953c7c7704e88d1481eaec09f6af6eb898158..a3dcd1b25557b440cc52078b8ce54920806b8be6 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zher");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHER_IDX, "ocl_cblas_zher");
#ifdef __cplusplus
try
#else
index 884cc561f075ee5e9d3d39802d2559f371422a98..146d17ff455202d1746c5b1c112724192aff2c37 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zher2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHER2_IDX, "ocl_cblas_zher2");
#ifdef __cplusplus
try
#else
index 80f76dcc4518ccb80b6476bbec23aead7dfb3f3d..6f52d3ae5d260957a04446aee24309c2ca037cfd 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zher2k");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHER2K_IDX, "ocl_cblas_zher2k");
#ifdef __cplusplus
try
#else
index aad6fdbf2654856d9e222b47404008e8f35cae6b..e5437115432891ca5fa07b00195a43f301729402 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zherk");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHERK_IDX, "ocl_cblas_zherk");
#ifdef __cplusplus
try
#else
index 85d73c88a1477dde1805728dfbfe5f82b5dde3da..9a5dd8fc42da0346c0cb73bb11b8a59dc3d11b98 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zhpmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHPMV_IDX, "ocl_cblas_zhpmv");
#ifdef __cplusplus
try
#else
index 5632ee733ddb4c0c136d8bd456105a77f785bfbe..d479109e494b12efc0b9afcd9f963d6cb9f40687 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zhpr");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHPR_IDX, "ocl_cblas_zhpr");
#ifdef __cplusplus
try
#else
index 6063f828c0982ff6206ca3d28208009dd46ddab0..e55ec9f3bcf193671f735c27c7e6434eb0694065 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zhpr2");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZHPR2_IDX, "ocl_cblas_zhpr2");
#ifdef __cplusplus
try
#else
index ef70e02461b83eea7d3f48b20269628e6b5d085e..41d2041dd63bd8727e59b8ead3fe98b8c39c6f07 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zrotg");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZROTG_IDX, "ocl_cblas_zrotg");
#ifdef __cplusplus
try
#else
index fd6a16359b6ac422ce8c7512a22aaf85391e2cb3..c08248fa6b7bb3457d89d3fa723d7ceb3dd43018 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zscal");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZSCAL_IDX, "ocl_cblas_zscal");
#ifdef __cplusplus
try
#else
index f8b98c7debb8cdd7a9a9e9afa4467c7e518e7397..a966c27943faa13cd8cda00a678db1646a2ee556 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zswap");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZSWAP_IDX, "ocl_cblas_zswap");
#ifdef __cplusplus
try
#else
index e6f0df99ded8c6c3c17888eeeaa34827c78377e9..deda9fd7c9995cd58cb5b0f7f602228acdd4125c 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zsymm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZSYMM_IDX, "ocl_cblas_zsymm");
#ifdef __cplusplus
try
#else
index 15ab279a0305be36860d3edece24d0d75cbf4b3f..62de2c1216ee1e8b84370c748851aa67baf099d7 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zsyr2k");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZSYR2K_IDX, "ocl_cblas_zsyr2k");
#ifdef __cplusplus
try
#else
index 32767e6e06cd993a3cb9907ecd39468bd7d298f7..5271187201085378fb052d7f2ef8071ff0b31caa 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_zsyrk");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZSYRK_IDX, "ocl_cblas_zsyrk");
#ifdef __cplusplus
try
#else
index a160fe79eb03598db3855c65e25a160f19ab1e02..5b7dc34149d73f267e505e6b2a9bcf50e45b0282 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ztbmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZTBMV_IDX, "ocl_cblas_ztbmv");
#ifdef __cplusplus
try
#else
index bb9f6916d1f9b1de28db005b53ee4058a86344b9..fd9ba055b3443d816665ce5f420ea50b4206b7f3 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ztbsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZTBSV_IDX, "ocl_cblas_ztbsv");
#ifdef __cplusplus
try
#else
index f631ce67bd3548f6f80b78982173375d4d50e31c..469a4996d0cea55b2c59c4dd90816391dc6254ce 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ztpmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZTPMV_IDX, "ocl_cblas_ztpmv");
#ifdef __cplusplus
try
#else
index 5c00274d80d49b6c306a4a28d9f0dc6bf63f5482..4def0efd93f10316fc3993786f16584f4c7f2cce 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ztpsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZTPSV_IDX, "ocl_cblas_ztpsv");
#ifdef __cplusplus
try
#else
index 859a87bccd24a06944aaf557f81a2dd21eb28e2f..803bc4bc43c0fd03d6af68b031d8cd64a74129bd 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ztrmm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZTRMM_IDX, "ocl_cblas_ztrmm");
#ifdef __cplusplus
try
#else
index d979d62ac2912aa4db704cbd890c146766405d69..a0d12f44c92e89a983365953a0bd51ad19e539b3 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ztrmv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZTRMV_IDX, "ocl_cblas_ztrmv");
#ifdef __cplusplus
try
#else
index cba160981ca2e0a882c0bb50aa75e39bdec02232..8a2411a5c917f88e3b78197918ffa6bebca23964 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ztrsm");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZTRSM_IDX, "ocl_cblas_ztrsm");
#ifdef __cplusplus
try
#else
index 46329f850506fb49408da40afb9e04d0546cd1cd..02d2797cc74948d66524f55e0f86282d8a5137b9 100644 (file)
#else
cl_kernel __K;
#endif
- __K = ti_cblas_get_kernel("ocl_cblas_ztrsv");
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_ZTRSV_IDX, "ocl_cblas_ztrsv");
#ifdef __cplusplus
try
#else
index e7e2c910eaf049ed2c06075a2f6f9fd0c9d315a5..f2dd549b161d11bab0d5df6b8cac9fcd3c724d7e 100644 (file)
#endif
/* Global variables */
-Context* ti_cblas_ocl_context = NULL;
-std::vector<Device>* ti_cblas_ocl_devices = NULL;
-CommandQueue* ti_cblas_ocl_Q = NULL;
-Program::Binaries* ti_cblas_ocl_binary = NULL;
-//Program* ti_cblas_ocl_program = NULL;
-Program ti_cblas_ocl_program;
+#ifdef __cplusplus
+
+#if 0
+Context ti_cblas_ocl_context;
+std::vector<Device> ti_cblas_ocl_devices;
+CommandQueue ti_cblas_ocl_Q;
+Program::Binaries ti_cblas_ocl_binary;
+Program ti_cblas_ocl_program;
+Kernel* ti_cblas_ocl_kernels[TI_CBLAS_NUM_KERNELS];
+#else
+Context* ti_cblas_ocl_context = NULL;
+std::vector<Device>* ti_cblas_ocl_devices = NULL;
+CommandQueue* ti_cblas_ocl_Q = NULL;
+Program::Binaries* ti_cblas_ocl_binary = NULL;
+Program* ti_cblas_ocl_program = NULL;
+#endif
+#else
+cl_context ti_cblas_ocl_context;
+cl_command_queue ti_cblas_ocl_Q;
+cl_program ti_cblas_ocl_program;
+cl_kernel ti_cblas_ocl_kernels[TI_CBLAS_NUM_KERNELS];
+#endif
int ti_cblas_init_done = 0; /* flag to check if init is complete */
int ti_cblas_disable_debug = 0; /* runtime toggle to disable debug */
int ti_cblas_offload = TI_CBLAS_OFFLOAD_SIZE;
fprintf(stderr, "ERROR: (%s,%d)\n", msg, code);
}
+#ifdef __cplusplus
extern "C"
+#endif
int ti_blis_init(void)
{
- int r_val = 1;
- TI_CBLAS_DEBUG_PRINT("Initializing BLIS ARM\n");
- bli_init();
- TI_CBLAS_DEBUG_PRINT("BLIS ARM initialized\n");
-
- Event e;
- Kernel* __K;
-
- __K = ti_cblas_get_kernel("ocl_bli_init");
-
- try
- {
- void *msmc_ptr;
- TI_CBLAS_DEBUG_PRINT("Initializing BLIS DSP\n");
- msmc_ptr = ti_cblas_mem_alloc(MSMC_BUF_SIZE);
- Buffer buf_MSMC(*ti_cblas_ocl_context, CL_MEM_READ_WRITE|CL_MEM_USE_HOST_PTR, MSMC_BUF_SIZE, (void *)msmc_ptr);
- __K->setArg(0, buf_MSMC);
-
- __K->setArg(1, __local(L2_BUF_SIZE));
+ int r_val = 1;
+ TI_CBLAS_DEBUG_PRINT("Initializing BLIS ARM\n");
+ bli_init();
+ TI_CBLAS_DEBUG_PRINT("BLIS ARM initialized\n");
+
+#ifdef __cplusplus
+ Event e;
+ Kernel* __K;
+#else
+ cl_kernel __K;
+#endif
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CGEMM_IDX, "ocl_bli_init");
+#ifdef __cplusplus
+ try
+#else
+ cl_int err = CL_SUCCESS;
+#endif
+ {
+ void *msmc_ptr;
+ TI_CBLAS_DEBUG_PRINT("Initializing BLIS DSP\n");
+ msmc_ptr = ti_cblas_mem_alloc(MSMC_BUF_SIZE);
+#ifdef __cplusplus
+ Buffer buf_MSMC(*ti_cblas_ocl_context, CL_MEM_READ_WRITE|CL_MEM_USE_HOST_PTR, MSMC_BUF_SIZE, (void *)msmc_ptr);
+ __K->setArg(0, buf_MSMC);
- ti_cblas_ocl_Q->enqueueTask(*__K, 0, &e);
- e.wait();
+#else
+ cl_mem buf_MSMC = clCreateBuffer(ti_cblas_ocl_context, CL_MEM_READ_WRITE|CL_MEM_USE_HOST_PTR, MSMC_BUF_SIZE, (void *)msmc_ptr, &err);
+ TI_CBLAS_OCL_CHKERROR("clCreateBuffer",err);
+ err |= clSetKernelArg(__K, 0, sizeof(buf_MSMC), &buf_MSMC);
+ TI_CBLAS_OCL_CHKERROR("clSetKernelArg",err);
+#endif
- ti_cblas_mem_free(msmc_ptr);
- ti_cblas_delete_kernel(__K);
- TI_CBLAS_DEBUG_PRINT("BLIS DSP initialized\n");
+#ifdef __cplusplus
+ __K->setArg(1, __local(L2_BUF_SIZE));
+#else
+ err |= clSetKernelArg(__K, 1, L2_BUF_SIZE, NULL);
+#endif
- }
+#ifdef __cplusplus
+ ti_cblas_ocl_Q->enqueueTask(*__K, 0, &e);
+ e.wait();
+#else
+ cl_event e;
+ err |= clEnqueueTask(ti_cblas_ocl_Q, __K, 0, 0, &e);
+ TI_CBLAS_OCL_CHKERROR("clEnqueueTask",err);
+ err |= clWaitForEvents(1, &e);
+ TI_CBLAS_OCL_CHKERROR("clWaitForEvents",err);
+ err |= clReleaseEvent(e);
+ TI_CBLAS_OCL_CHKERROR("clReleaseEvent",err);
- catch (Error err)
- {
- ti_cblas_error(err.what(),err.err());
- r_val = 1;
- return r_val;
- }
+#endif
+ ti_cblas_mem_free(msmc_ptr);
+ ti_cblas_delete_kernel(__K);
+ TI_CBLAS_DEBUG_PRINT("BLIS DSP initialized\n");
+
+ }
+#ifdef __cplusplus
+ catch (Error err)
+ {
+ ti_cblas_error(err.what(),err.err());
+ r_val = 1;
+ return r_val;
+ }
+#endif
}
+#ifdef __cplusplus
extern "C"
+#endif
int ti_blis_finalize(void)
{
- int r_val = 1;
- bli_finalize();
+ int r_val = 1;
+ bli_finalize();
- Event e;
- Kernel* __K;
- __K = ti_cblas_get_kernel("ocl_bli_finalize");
-
- try
- {
- ti_cblas_ocl_Q->enqueueTask(*__K, 0, &e);
- e.wait();
- ti_cblas_delete_kernel(__K);
- }
+#ifdef __cplusplus
+ Event e;
+ Kernel* __K;
+#else
+ cl_kernel __K;
+#endif
+ __K = ti_cblas_get_kernel(TI_CBLAS_CBLAS_CGEMM_IDX, "ocl_bli_finalize");
+#ifdef __cplusplus
+ try
+#else
+ cl_int err = CL_SUCCESS;
+#endif
+ {
+#ifdef __cplusplus
+ ti_cblas_ocl_Q->enqueueTask(*__K, 0, &e);
+ e.wait();
+#else
+ cl_event e;
+ err |= clEnqueueTask(ti_cblas_ocl_Q, __K, 0, 0, &e);
+ TI_CBLAS_OCL_CHKERROR("clEnqueueTask",err);
+ err |= clWaitForEvents(1, &e);
+ TI_CBLAS_OCL_CHKERROR("clWaitForEvents",err);
+ err |= clReleaseEvent(e);
+ TI_CBLAS_OCL_CHKERROR("clReleaseEvent",err);
- catch (Error err)
- {
- ti_cblas_error(err.what(),err.err());
- r_val = 1;
- return r_val;
- }
+#endif
+ ti_cblas_delete_kernel(__K);
+ }
+#ifdef __cplusplus
+ catch (Error err)
+ {
+ ti_cblas_error(err.what(),err.err());
+ r_val = 1;
+ return r_val;
+ }
+#endif
}
+#ifdef __cplusplus
extern "C"
+#endif
int ti_cblas_finalize(void)
{
- int r_val = 1;
- //printf("ti_cblas_finalize\n");
+ int r_val = 1;
+ //printf("ti_cblas_finalize\n");
- /* If ti_cblas_init_done is equal to 0,
- * then we know that ti_cblas_init was not called,
- * and so we can return early.
- */
- if(ti_cblas_init_done == 0)
- return 0;
+ /* If ti_cblas_init_done is equal to 0,
+ * then we know that ti_cblas_init was not called,
+ * and so we can return early.
+ */
+ if(ti_cblas_init_done == 0)
+ return 0;
- //r_val = ti_blis_finalize();
- /*Using same name as ti_cblas_init critical region. See notes in bli_init*/
+ //r_val = ti_blis_finalize();
+ /*Using same name as ti_cblas_init critical region. See notes in bli_init*/
#pragma omp critical (ti_cblas_init_critical)
- {
- if (ti_cblas_init_done == 1)
- {
- // Destroy Pthread
- pthread_mutex_destroy(&MUTEX);
- pthread_cond_destroy (&CV);
-
- //destroy Command queue, program, devices and context.
- if(ti_cblas_ocl_Q != NULL)
- {
- delete(ti_cblas_ocl_Q);
- ti_cblas_ocl_Q = NULL;
- }
-/* if(ti_cblas_ocl_program != NULL)
- {
- delete(ti_cblas_ocl_program);
- ti_cblas_ocl_program = NULL;
- }*/
- if(ti_cblas_ocl_binary != NULL)
- {
- delete(ti_cblas_ocl_binary);
- ti_cblas_ocl_binary = NULL;
- }
- if(ti_cblas_ocl_devices != NULL)
- {
- delete(ti_cblas_ocl_devices);
- ti_cblas_ocl_devices = NULL;
- }
- if(ti_cblas_ocl_context != NULL)
- {
- delete(ti_cblas_ocl_context);
- ti_cblas_ocl_context = NULL;
- }
- ti_cblas_init_done = 0;
- r_val = 0;
- }
- }
- return r_val;
+ {
+ if (ti_cblas_init_done == 1)
+ {
+ // Destroy Pthread
+ pthread_mutex_destroy(&MUTEX);
+ pthread_cond_destroy (&CV);
+
+ //destroy Command queue, program, devices and context.
+ if(ti_cblas_ocl_Q != NULL)
+ {
+ delete(ti_cblas_ocl_Q);
+ ti_cblas_ocl_Q = NULL;
+ }
+ if(ti_cblas_ocl_program != NULL)
+ {
+ delete(ti_cblas_ocl_program);
+ ti_cblas_ocl_program = NULL;
+ }
+ if(ti_cblas_ocl_binary != NULL)
+ {
+ delete(ti_cblas_ocl_binary);
+ ti_cblas_ocl_binary = NULL;
+ }
+ if(ti_cblas_ocl_devices != NULL)
+ {
+ delete(ti_cblas_ocl_devices);
+ ti_cblas_ocl_devices = NULL;
+ }
+ if(ti_cblas_ocl_context != NULL)
+ {
+ delete(ti_cblas_ocl_context);
+ ti_cblas_ocl_context = NULL;
+ }
+ ti_cblas_init_done = 0;
+ r_val = 0;
+ }
+ }
+ return r_val;
}
void ti_cblas_auto_finalize(void)
}
}
- /* 3-digit value: 012
- * Left-most digit => L1 (0)
- * Middle-digit => L2 (1)
- * Right-most => L3 (2)
- */
- TI_CBLAS_L1_OFFLOAD = ti_cblas_offload / 100;
- int tmp_offload = ti_cblas_offload % 100;
- TI_CBLAS_L2_OFFLOAD = tmp_offload / 10;
- TI_CBLAS_L3_OFFLOAD = tmp_offload % 10;
- TI_CBLAS_DEBUG_PRINT("BLAS Offload values: L1=%d, L2=%d, L3=%d\n",
- TI_CBLAS_L1_OFFLOAD, TI_CBLAS_L2_OFFLOAD, TI_CBLAS_L3_OFFLOAD);
- if ((TI_CBLAS_L1_OFFLOAD == TI_CBLAS_OFFLOAD_SIZE)) {
- TI_CBLAS_ERROR_EXIT("Size-based offload NOT supported for BLAS Level 1 yet.\n");
- }
- if ((TI_CBLAS_L2_OFFLOAD == TI_CBLAS_OFFLOAD_SIZE)) {
- TI_CBLAS_ERROR_EXIT("Size-based offload NOT supported for BLAS Level 2 yet.\n");
- }
+ /* 3-digit value: 012
+ * Left-most digit => L1 (0)
+ * Middle-digit => L2 (1)
+ * Right-most => L3 (2)
+ */
+ TI_CBLAS_L1_OFFLOAD = ti_cblas_offload / 100;
+ int tmp_offload = ti_cblas_offload % 100;
+ TI_CBLAS_L2_OFFLOAD = tmp_offload / 10;
+ TI_CBLAS_L3_OFFLOAD = tmp_offload % 10;
+ TI_CBLAS_DEBUG_PRINT("BLAS Offload values: L1=%d, L2=%d, L3=%d\n",
+ TI_CBLAS_L1_OFFLOAD, TI_CBLAS_L2_OFFLOAD, TI_CBLAS_L3_OFFLOAD);
+ if ((TI_CBLAS_L1_OFFLOAD == TI_CBLAS_OFFLOAD_SIZE)) {
+ TI_CBLAS_ERROR_EXIT("Size-based offload NOT supported for BLAS Level 1 yet.\n");
+ }
+ if ((TI_CBLAS_L2_OFFLOAD == TI_CBLAS_OFFLOAD_SIZE)) {
+ TI_CBLAS_ERROR_EXIT("Size-based offload NOT supported for BLAS Level 2 yet.\n");
+ }
/*------------------------------------------------------------------------
- * Read the offline compiled kernel module
- *-----------------------------------------------------------------------*/
- TI_CBLAS_DEBUG_PRINT("Reading Kernels\n");
- const unsigned char* bin;
+ * Read the offline compiled kernel module
+ *-----------------------------------------------------------------------*/
+ TI_CBLAS_DEBUG_PRINT("Reading Kernels\n");
+ const unsigned char* bin;
#ifdef TI_CBLAS_FAT_BINARY
bin = (unsigned char *)ti_cblas_kernel_dsp_bin;
const size_t bin_length = ti_cblas_kernel_dsp_bin_len;
#else
const char binary[] = "./ti_cblas_kernel.out";
unsigned int bin_length;
-
+#ifdef __cplusplus
bin_length = ocl_read_binary(binary, (char*&)bin);
+#else
+ FILE *fp = fopen(binary, "r");
+ if (!fp) {
+ TI_CBLAS_ERROR_EXIT("Could not open OpenCL pre-compiled binary %s for reading\n", binary);
+ }
+ struct stat fileinfo;
+ stat(binary, &fileinfo);
+ bin_length = fileinfo.st_size;
+ bin = (char *)malloc(bin_length);
+ if (!bin) {
+ TI_CBLAS_ERROR_EXIT("Could not malloc of size %d for reading OpenCL binary\n", bin_length);
+ }
+ if (fread((char *)bin, bin_length, 1, fp) != 1) {
+ TI_CBLAS_ERROR_EXIT("Could not read %d bytes of OpenCL binary\n", bin_length);
+ }
+ fclose(fp);
+#endif /* cplusplus */
#endif /* FAT_BINARY */
/* OpenCL init */
TI_CBLAS_DEBUG_PRINT("Initializing OpenCL\n");
+#ifdef __cplusplus
+ ti_cblas_ocl_context = new Context(CL_DEVICE_TYPE_ACCELERATOR);
+ ti_cblas_ocl_devices = new std::vector<Device> (ti_cblas_ocl_context->getInfo<CL_CONTEXT_DEVICES>());
+ ti_cblas_ocl_binary = new Program::Binaries(1, std::make_pair(bin, bin_length));
+ ti_cblas_ocl_program = new Program(*ti_cblas_ocl_context, *ti_cblas_ocl_devices, *ti_cblas_ocl_binary);
+ ti_cblas_ocl_program->build(*ti_cblas_ocl_devices);
+ ti_cblas_ocl_Q = new CommandQueue(*ti_cblas_ocl_context, ti_cblas_ocl_devices[0][0], CL_QUEUE_PROFILING_ENABLE);
+#else
+ cl_int err;
+ cl_device_id device;
+ /* Create an in-order command queue by default*/
+ int queue_flags = 0;
+#ifdef TI_CBLAS_PROFILE
+ queue_flags |= CL_QUEUE_PROFILING_ENABLE;
+#endif
- ti_cblas_ocl_context = new Context(CL_DEVICE_TYPE_ACCELERATOR);
- ti_cblas_ocl_devices = new std::vector<Device> (ti_cblas_ocl_context->getInfo<CL_CONTEXT_DEVICES>());
- ti_cblas_ocl_binary = new Program::Binaries(1, std::make_pair(bin, bin_length));
- //ti_cblas_ocl_program = new Program(*ti_cblas_ocl_context, *ti_cblas_ocl_devices, *ti_cblas_ocl_binary);
- ti_cblas_ocl_program = Program(*ti_cblas_ocl_context, *ti_cblas_ocl_devices, *ti_cblas_ocl_binary);
- //ti_cblas_ocl_program->build(*ti_cblas_ocl_devices);
- ti_cblas_ocl_program.build(*ti_cblas_ocl_devices);
- ti_cblas_ocl_Q = new CommandQueue(*ti_cblas_ocl_context, ti_cblas_ocl_devices[0][0], CL_QUEUE_PROFILING_ENABLE);
+ ti_cblas_ocl_context = clCreateContextFromType(0,CL_DEVICE_TYPE_ACCELERATOR,0,0,&err);
+ TI_CBLAS_OCL_CHKERROR("clCreateContextFromType",err);
+ err = clGetDeviceIDs(0,CL_DEVICE_TYPE_ACCELERATOR,1,&device,0);
+ TI_CBLAS_OCL_CHKERROR("clGetDeviceIDs",err);
+ ti_cblas_ocl_Q = clCreateCommandQueue(ti_cblas_ocl_context, device, queue_flags, &err);
+ TI_CBLAS_OCL_CHKERROR("clCreateCommandQueue",err);
+ ti_cblas_ocl_program = clCreateProgramWithBinary(ti_cblas_ocl_context, 1, &device, &bin_length, &bin, NULL, &err);
+ TI_CBLAS_OCL_CHKERROR("clCreateProgramWithBinary",err);
+ const char *compile_options = "";
+ err = clBuildProgram(ti_cblas_ocl_program, 1, &device, compile_options, 0, 0);
+ TI_CBLAS_OCL_CHKERROR("clBuildProgram",err);
+
+#endif
#ifndef TI_CBLAS_FAT_BINARY
+#ifdef __cplusplus
delete [] bin;
+#else
+ free((char*)bin);
+#endif
#endif /* FAT_BINARY */
-
TI_CBLAS_DEBUG_PRINT("OpenCL initialized\n");
-
+
TI_CBLAS_DEBUG_PRINT("Initializing Pthreads\n");
-
- /* Initializing pthreads */
- pthread_cond_init (&CV, 0);
- pthread_mutex_init(&MUTEX, 0);
- TI_CBLAS_DEBUG_PRINT("Pthreads initialized\n");
-
- TI_CBLAS_DEBUG_PRINT("Initializing BLIS\n");
- ti_blis_init();
- TI_CBLAS_DEBUG_PRINT("BLIS initialized\n");
-
- atexit(ti_cblas_auto_finalize);
-
- TI_CBLAS_PROFILE_REPORT(" Initialization took %8.2f us\n", (float) clock_diff);
- ti_cblas_init_done = 1;
- TI_CBLAS_DEBUG_PRINT("ti_cblas_init: Finished OpenCL initialization\n");
- } //end of !ti_cblas_init_done
-
- } // End of critical section
-
- return;
+ /* Initializing pthreads */
+ pthread_cond_init (&CV, 0);
+ pthread_mutex_init(&MUTEX, 0);
+ TI_CBLAS_DEBUG_PRINT("Pthreads initialized\n");
+
+ TI_CBLAS_DEBUG_PRINT("Initializing BLIS\n");
+ ti_blis_init();
+ TI_CBLAS_DEBUG_PRINT("BLIS initialized\n");
+
+ atexit(ti_cblas_auto_finalize);
+
+ TI_CBLAS_PROFILE_REPORT(" Initialization took %8.2f us\n", (float) clock_diff);
+ ti_cblas_init_done = 1;
+ TI_CBLAS_DEBUG_PRINT("ti_cblas_init: Finished OpenCL initialization\n");
+ } //end of !ti_cblas_init_done
+ } // End of critical section
+ return;
}
void ti_cblas_mem_free(void *ptr)
{
- pthread_mutex_lock(&MUTEX);
- __free_msmc(ptr);
- pthread_cond_broadcast(&CV);
- pthread_mutex_unlock(&MUTEX);
+ pthread_mutex_lock(&MUTEX);
+ __free_msmc(ptr);
+ pthread_cond_broadcast(&CV);
+ pthread_mutex_unlock(&MUTEX);
}
void *ti_cblas_mem_alloc(size_t size)
{
- void *ptr;
- pthread_mutex_lock(&MUTEX);
- /*-------------------------------------------------------------------------
+ void *ptr;
+ pthread_mutex_lock(&MUTEX);
+ /*-------------------------------------------------------------------------
+
+ * Loop in case of false signal after broadcast.
- * Loop in case of false signal after broadcast.
+ *------------------------------------------------------------------------*/
+ while ((ptr = __malloc_msmc(size)) == 0)
- *------------------------------------------------------------------------*/
- while ((ptr = __malloc_msmc(size)) == 0)
+ pthread_cond_wait(&CV, &MUTEX);
+ pthread_mutex_unlock(&MUTEX);
+ return ptr;
- pthread_cond_wait(&CV, &MUTEX);
- pthread_mutex_unlock(&MUTEX);
- return ptr;
}
* function with index 'idx'. Initializes the handle if it's
* not been used before, otherwise returns earlier handle
*/
-Kernel* ti_cblas_get_kernel(const char *fname)
+#ifdef __cplusplus
+Kernel*
+#else
+cl_kernel
+#endif
+ti_cblas_get_kernel(int idx, const char *fname)
{
- Kernel* __K;
-
- TI_CBLAS_DEBUG_PRINT("In ti_cblas_get_kernel: to get kernel %s.\n", fname);
- __K = new Kernel(ti_cblas_ocl_program, fname);
- TI_CBLAS_DEBUG_PRINT("ti_cblas_get_kernel: kernel %s is obtained.\n", fname);
+#if 0
+ if (!ti_cblas_kernel_valid[idx]) {
+#ifdef __cplusplus
+ ti_cblas_ocl_kernels[idx] = new Kernel(ti_cblas_ocl_program, fname);
+#else
+ cl_int err;
+ ti_cblas_ocl_kernels[idx] = clCreateKernel(ti_cblas_ocl_program,fname,&err);
+ TI_CBLAS_OCL_CHKERROR("clCreateKernel",err);
+#endif
+ ti_cblas_kernel_valid[idx] = 1;
+ }
+ return ti_cblas_ocl_kernels[idx];
+#else
+#ifdef __cplusplus
+ Kernel* __K;
+#else
+ cl_kernel __K;
+#endif
+#ifdef __cplusplus
+ __K = new Kernel(*ti_cblas_ocl_program, fname);
+#else
+ cl_int err;
+ __K = clCreateKernel(ti_cblas_ocl_program,fname,&err);
+ TI_CBLAS_OCL_CHKERROR("clCreateKernel",err);
+#endif
- return __K;
+ return __K;
+#endif
}
+#ifdef __cplusplus
int ti_cblas_delete_kernel(Kernel* K)
+#else
+int ti_cblas_delete_kernel(cl_kernel K)
+#endif
{
- if(K != NULL){
- delete(K);
- K = NULL;
- }
-
- return 0;
+#ifdef __cplusplus
+ if(K != NULL)
+ {
+ delete(K);
+ K=NULL;
+ }
+#else
+ clReleaseKernel(K);
+#endif
+ return 0;
}
+
+
index d207e037c8dfa385f7de1d60d36fdbab11459d74..c08b76ec14b133622e77be3c160e4d1b9d053ad8 100644 (file)
* THE POSSIBILITY OF SUCH DAMAGE.
*****************************************************************************/
-#include <stdio.h>
+/*#include <stdio.h>*/
#include "../../ticblas/ticblas.h"
#include <libarch.h>
@@ -49,28 +49,37 @@ int bli_l3_mem_config(double *msmc_buf, size_t msmc_buf_size, size_t *l1D_SRAM_s
||(smem_size_med > msmc_buf_size) /* provided MSMC memory */
||(smem_size_slow > BLIS_L3_DDR_SIZE_ZERO) /* DDR not used */
) {
- return(-2);
+ return(TICBLAS_INIT_ERROR);
}
/* Configure L1D if necessary */
*l1D_SRAM_size_orig = lib_get_L1D_SRAM_size(); /* get current L1D SRAM size */
l1d_cfg_err = LIB_CACHE_SUCCESS;
+/*
printf("Original L1D SRAM size is: %d\n", *l1D_SRAM_size_orig);
printf("Required L1D SRAM size is: %d\n", smem_size_vfast);
+*/
if(*l1D_SRAM_size_orig < smem_size_vfast) { /* configure L1D if needs more SRAM */
+ /*printf("Configuring L1D SRAM on all cores.\n");*/
#pragma omp parallel
{
l1d_cfg_err = lib_L1D_config_SRAM(smem_size_vfast);
}
}
-
- printf("New L1D SRAM size is: %d\n", lib_get_L1D_SRAM_size());
-
+/*
+ #pragma omp parallel
+ {
+ int core_id = lib_get_coreID();
+ printf("New L1D SRAM size from core %d is: %d\n", core_id, lib_get_L1D_SRAM_size());
+ }
+*/
/* Configure L2 if necessary */
*l2_SRAM_size_orig = lib_get_L2_SRAM_size(); /* get current L2 SRAM size */
l2_cfg_err = LIB_CACHE_SUCCESS;
+/*
printf("Original L2 SRAM size is: %d\n", *l2_SRAM_size_orig);
printf("Required L2 SRAM size is: %d\n", smem_size_fast);
+*/
if(*l2_SRAM_size_orig < smem_size_fast) { /* configure L2 if needs more SRAM */
#pragma omp parallel
{
@@ -79,15 +88,18 @@ int bli_l3_mem_config(double *msmc_buf, size_t msmc_buf_size, size_t *l1D_SRAM_s
}
if(l1d_cfg_err || l2_cfg_err) {
- return(-3);
+ return(TICBLAS_INIT_ERROR);
}
-
+/*
printf("New L2 SRAM size is: %d\n", lib_get_L2_SRAM_size());
-
+*/
/* get L1D and L2 SRAM base address */
l1d_SRAM_ptr = lib_get_L1D_SRAM_base();
l2_SRAM_ptr = lib_get_L2_SRAM_base();
-
+/*
+ printf("L1D SRAM base address is 0x%x.\n", (unsigned int)l1d_SRAM_ptr);
+ printf("L2 SRAM base address is 0x%x.\n", (unsigned int) l2_SRAM_ptr);
+*/
/* pass allocated memories for heap initialization */
return(tiCblasInit(l1d_SRAM_ptr, smem_size_vfast,
l2_SRAM_ptr, smem_size_fast,
void ti_bli_init_dsp(char *l3_buf, char *l2_buf)
{
printf("In function ti_bli_init_dsp, l3_buff is 0x%x, l2_buf is 0x%x.\n", (unsigned int)l3_buf, (unsigned int)l2_buf);
- bli_init();
+ bli_init();
}
/* This function will be removed. Function tiCblasDelete() will be used instead. */
void ti_bli_finalize_dsp(void)
{
- bli_finalize();
+ bli_finalize();
}
index df2abc1495169b6b3d8b9329812b7c2bdce751c5..8121b8d7fef96d1ffa518bfcd5df40782cea3212 100755 (executable)
#else
cl_kernel __K;
#endif
- __K = ${namespace}_get_kernel(\"ocl_$trampname\");
+ __K = ${namespace}_get_kernel($trampdef, \"ocl_$trampname\");
+
#ifdef __cplusplus
try
#else
extern void ${namespace}_error(const char* msg, int code);
extern void ${namespace}_init(void);
#ifdef __cplusplus
-extern Kernel* ${namespace}_get_kernel(const char *fname);
+extern Kernel* ${namespace}_get_kernel(int idx, const char *fname);
extern Context ${namespace}_ocl_context;
extern std::vector<Device> ${namespace}_ocl_devices;
extern CommandQueue ${namespace}_ocl_Q;
extern Program ${namespace}_ocl_program;
extern Kernel* ${namespace}_ocl_kernels[];
#else
-extern cl_kernel ${namespace}_get_kernel(const char *fname);
+extern cl_kernel ${namespace}_get_kernel(int idx, const char *fname);
extern cl_context ${namespace}_ocl_context;
extern cl_command_queue ${namespace}_ocl_Q;
extern cl_program ${namespace}_ocl_program;
index 444819e15c0cce011796805f53f6719af2ca43d7..c5a6182d580c8e14ff383ca6f21d9b6ac0231624 100755 (executable)
//#include <ti/csl/device/k2h/src/cslr_device.h>
#include <libarch.h>
-//#include <ti/csl/csl_chipAux.h> // CSL_chipReadDNUM -> to read coreID
-//#include <ti/csl/csl_cacheAux.h> // CACHE_invL1d
+#include <ti/csl/csl_chipAux.h> // CSL_chipReadDNUM -> to read coreID
+#include <ti/csl/csl_cacheAux.h> // CACHE_invL1d
// for __clock64()
//#include <dsp_c.h>
index 252f79ca219eceaafeff21488da5b29838b2139c..d012a5eaa0d5fe0163546aadc645344f25984bee 100755 (executable)
CMISCFLAGS += -I$(XDC_DIR)/packages
CMISCFLAGS += -I$(BIOS_DIR)/packages
CMISCFLAGS += -I$(XDAIS_DIR)/packages
-CMISCFLAGS += -I$(LIBARCH_DIR)
+CMISCFLAGS += -I$(LIBARCH_DIR)/include
CMISCFLAGS += -I$(CGTROOT)/include
CMISCFLAGS += -I$(PDK_DIR)/packages
diff --git a/examples/dsponly/dgemm_test/Makefile.common b/examples/dsponly/dgemm_test/Makefile.common
index 2f0b8d4e1cf39a71ff009cc1570afe1f1f330230..08deee81cecf19a68aac54e3274dbf623cedc573 100644 (file)
LNK_CMD = $(CFGDIR)/linker.cmd
OPT_CMD = $(CFGDIR)/compiler.opt
CL_OPTS = -@ $(OPT_CMD) -mv6600 --omp -I $(OMP_DIR)/packages/ti/runtime/openmp
-CL_OPTS += -I$(FC_DIR)/packages -I$(XDAIS_DIR)/packages -I$(EDMA3_DIR)/packages -I$(LIBARCH_DIR) -I$(LINALG_DIR)/include
+CL_OPTS += -I$(FC_DIR)/packages -I$(XDAIS_DIR)/packages -I$(EDMA3_DIR)/packages -I$(LIBARCH_DIR)/include -I$(LINALG_DIR)/include
CL_OPTS += -D$(TARGET) -DLIB_RTOS
LNK_OPTS = -x -c --priority -w
CL = $(CGTROOT)/bin/cl6x
diff --git a/examples/make.inc b/examples/make.inc
index 36c7df426c50952d69494fb3b4c128beaf4496ca..d1a70af2bf105b0e1ab30bc9a6a2aa0f00fb720f 100644 (file)
--- a/examples/make.inc
+++ b/examples/make.inc
# gcc ARM cross compiler will not, by default, search the host's
# /usr/include. Explicitly specify here to find dependent vendor headers
-CC = arm-linux-gnueabihf-gcc
+ CC = arm-linux-gnueabihf-gcc
else
-CC = gcc
+ CC = gcc
endif
+CFLAGS = -g -O2 -I$(TARGET_ROOTDIR)/usr/include -I$(LINALG_DIR)/include
-CFLAGS = -g -O2 -I$(TARGET_ROOTDIR)/usr/include
-
-LIB_DIR = $(TARGET_ROOTDIR)/usr/lib/
+LIB_DIR = $(LINALG_DIR)/lib/
LD_FLAGS=-L$(TARGET_ROOTDIR)/lib -L$(TARGET_ROOTDIR)/usr/lib -Wl,-rpath-link,$(TARGET_ROOTDIR)/lib -Wl,-rpath-link,$(TARGET_ROOTDIR)/usr/lib
BLASLIB = $(LIB_DIR)libcblas_armplusdsp.a $(LIB_DIR)libblis.a -lOpenCL -locl_util -lstdc++ -lrt -lm -lgomp -lpthread
LAPACKLIB = $(LIB_DIR)libcblaswr.a $(LIB_DIR)liblapack.a $(LIB_DIR)libf2c.a $(LIB_DIR)libcblas_armplusdsp.a $(LIB_DIR)libblis.a -lOpenCL -locl_util -lstdc++ -lrt -lm -lgomp -lpthread
diff --git a/make.inc b/make.inc
index 07519e8f3d8ea025447174239eac611007d4c64d..b2e1afec899fc27de0980f28c6d0d6af27e19070 100644 (file)
--- a/make.inc
+++ b/make.inc
$$(info Using $(1) = $$($(1)))
endef
-# CBLAS and BLIS directories
-CBLAS_DSP_LIB = ../../cblas/lib/C66/libcblas_C66.ae66
-TICBLAS_DSP_LIB = ../../ticblas/lib/libticblas.ae66
-CBLAS_ARM_LIB = ../../cblas/lib/ARM/libcblas_ARM.a
-LIBARCH_LIB = $(LIBARCH_DIR)/lib/libArch.a66x
-
-ifeq ($(MEM_MODEL),Large)
-BLIS_DSP_LIB = ../../blis/install/c66xLarge/lib/libblis.ae66
-else ifeq ($(MEM_MODEL),Medium)
-BLIS_DSP_LIB = ../../blis/install/c66xMedium/lib/libblis.ae66
-else ifeq ($(MEM_MODEL),Small)
-BLIS_DSP_LIB = ../../blis/install/c66xSmall/lib/libblis.ae66
-#else ifeq ($(MEM_MODEL),Tiny)
-endif
-
%.o: %.cpp
@echo Compiling $<
$(CPP) -c $(CPP_FLAGS) $<
diff --git a/setup_env_devkit.sh b/setup_env_devkit.sh
--- /dev/null
+++ b/setup_env_devkit.sh
@@ -0,0 +1,15 @@
+#!/bin/bash
+
+export TI_OCL_INSTALL_DIR="/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/cortexa15hf-vfp-neon-linux-gnueabi/usr/share/ti/opencl"
+export CGTROOT="/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/x86_64-arago-linux/usr/share/ti/cgt-c6x"
+export TI_OCL_CGT_INSTALL="/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/x86_64-arago-linux/usr/share/ti/cgt-c6x"
+export PDK_DIR="/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/cortexa15hf-vfp-neon-linux-gnueabi/usr/share/ti/ti-pdk-tree"
+export FC_DIR="/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/cortexa15hf-vfp-neon-linux-gnueabi/usr/share/ti/ti-framework-components-tree"
+export XDAIS_DIR="/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/cortexa15hf-vfp-neon-linux-gnueabi/usr/share/ti/ti-xdais-tree"
+export BIOS_DIR="/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/cortexa15hf-vfp-neon-linux-gnueabi/usr/share/ti/ti-sysbios-tree"
+export OMP_DIR="/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/cortexa15hf-vfp-neon-linux-gnueabi/usr/share/ti/ti-omp-tree"
+export LIBARCH_DIR="/home/a0869574local/proclibs/libarch_intgit/libarch"
+export TARGET_ROOTDIR="/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/cortexa15hf-vfp-neon-linux-gnueabi"
+export XDC_DIR=/home/a0869574local/ti-rtos-sdk-12-08/xdctools_3_31_02_38_core
+
+export PATH=/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/x86_64-arago-linux/usr/share/ti/cgt-c6x/bin:/home/a0869574local/ti/processor-sdk-linux-k2hk-evm-02.00.01.06/linux-devkit/sysroots/x86_64-arago-linux/usr/bin:$PATH
diff --git a/ticblas/src/Makefile b/ticblas/src/Makefile
index 60bc2d61ac51576847777a98553364c76dccc43a..11770beb015760da186c64b9dd8f7533d0a32aaa 100644 (file)
--- a/ticblas/src/Makefile
+++ b/ticblas/src/Makefile
INCDIR += -I$(XDC_DIR)/packages
INCDIR += -I$(BIOS_DIR)/packages
INCDIR += -I$(XDAIS_DIR)/packages
-INCDIR += -I$(LIBARCH_DIR)
+INCDIR += -I$(LIBARCH_DIR)/include
INCDIR += -I$(PDK_DIR)/packages
ifeq ($(LIBOS),LIB_OPENCL)