AI/rocm-systems
2
0
Форкнуть 0
Код Задачи Запросы на слияние Действия Пакеты Проекты Релизы Вики Активность

hipHostRegister and hipHostMalloc refactor.

Просмотр исходного кода 
Note hipHostMalloc (not hipHostAlloc or hipMallocHost).
 -  the hipHost* is used for all HIP APIs dealing with Host memory.
    (including hipHostMalloc, hipHostFree, hipHostRegister,
hipHostUnregister, hipHostGetFlags, hipHostGetDevicePointer).
  - hipMallocHost is consistent with "hipMalloc" for allocating device
    memory.  Enumerations hipHostMalloc* also used as optional
    flags parm to hipHostMalloc.


[ROCm/clr commit: 2d0fade1f7]
Этот коммит содержится в:
Ben Sander
2016-03-22 02:30:10 -05:00
родитель db8bff1235
Коммит 37a02661a6
17 изменённых файлов: 95 добавлений и 70 удалений
Скачать Patch файл Скачать Diff файл Показать все файлы Свернуть все файлы
projects/clr/hipamd/RELEASE.md
+1 -1
Просмотреть файл
@@ -15,7 +15,7 @@ Stay tuned - the work for many of these features is already in-flight.
Next:
- Deprecate hipDeviceGetProp, replace with hipGetDeviceProp
- Deprecate hipMallocHost (Replace with hipHostAlloc with hipHostAllocDefault as last parameter).
- Deprecate hipMallocHost (Replace with hipHostMalloc)
- Deprecate hipFreeHost (Replace with hipHostFree).
projects/clr/hipamd/bin/hipify
+6 -6
Просмотреть файл
@@ -283,15 +283,15 @@ while (@ARGV) {
#--------
# Memory management:
$ft{'mem'} += s/\bcudaMalloc\b/hipMalloc/g;
$ft{'mem'} += s/\bcudaMallocHost\b/hipMallocHost/g;
$ft{'mem'} += s/\bcudaMallocHost\b/hipHostMalloc/g; # note conversion to standard hipHost* naming convention
$ft{'mem'} += s/\bcudaFree\b/hipFree/g;
$ft{'mem'} += s/\bcudaFreeHost\b/hipFreeHost/g;
$ft{'mem'} += s/\bcudaFreeHost\b/hipHostFree/g; # note conversion to standard hipHost* naming convention
$ft{'mem'} += s/\bcudaHostAlloc\b/hipHostAlloc/g;
$ft{'mem'} += s/\bcudaHostGetDevicePointer\b/hipHostGetDevicePointer/g;
$ft{'mem'} += s/\bcudaHostAllocDefault\b/hipHostAllocDefault/g;
$ft{'mem'} += s/\bcudaHostAllocPortable\b/hipHostAllocPortable/g;
$ft{'mem'} += s/\bcudaHostAllocMapped\b/hipHostAllocMapped/g;
$ft{'mem'} += s/\bcudaHostAllocWriteCombined\b/hipHostAllocWriteCombined/g;
$ft{'mem'} += s/\bcudaHostAllocDefault\b/hipHostMallocDefault/g;
$ft{'mem'} += s/\bcudaHostAllocPortable\b/hipHostMallocPortable/g;
$ft{'mem'} += s/\bcudaHostAllocMapped\b/hipHostMallocMapped/g;
$ft{'mem'} += s/\bcudaHostAllocWriteCombined\b/hipHostMallocWriteCombined/g;
#--------
projects/clr/hipamd/include/hcc_detail/hip_runtime_api.h
+9 -8
Просмотреть файл
@@ -56,10 +56,10 @@ extern "C" {
#define hipEventInterprocess 0x4 ///< Event can support IPC. @warning - not supported in HIP.
#define hipHostAllocDefault 0x0
#define hipHostAllocPortable 0x1
#define hipHostAllocMapped 0x2
#define hipHostAllocWriteCombined 0x4
#define hipHostMallocDefault 0x0
#define hipHostMallocPortable 0x1
#define hipHostMallocMapped 0x2
#define hipHostMallocWriteCombined 0x4
#define hipHostRegisterDefault 0x0
#define hipHostRegisterPortable 0x1
@@ -670,7 +670,7 @@ hipError_t hipMalloc(void** ptr, size_t size) ;
* @param[in] size Requested memory size
* @return Error code
*/
hipError_t hipMallocHost(void** ptr, size_t size) __attribute__((deprecated("use hipHostAlloc instead"))) ;
hipError_t hipMallocHost(void** ptr, size_t size) __attribute__((deprecated("use hipHostMalloc instead"))) ;
/**
* @brief Allocate device accessible page locked host memory
@@ -680,7 +680,8 @@ hipError_t hipMallocHost(void** ptr, size_t size) __attribute__((deprecated("use
* @param[in] flags Type of host memory allocation
* @return Error code
*/
hipError_t hipHostAlloc(void** ptr, size_t size, unsigned int flags) ;
hipError_t hipHostMalloc(void** ptr, size_t size, unsigned int flags) ;
hipError_t hipHostAlloc(void** ptr, size_t size, unsigned int flags) __attribute__((deprecated("use hipHostMalloc instead"))) ;;
/**
* @brief Get Device pointer from Host Pointer allocated through hipHostAlloc
@@ -696,7 +697,7 @@ hipError_t hipHostGetDevicePointer(void** devPtr, void* hstPtr, size_t size) ;
* @brief Get flags associated with host pointer
*
* @param[out] flagsPtr Memory location to store flags
* @param[in] hostPtr Host Pointer allocated through hipHostAlloc
* @param[in] hostPtr Host Pointer allocated through hipHostMalloc
* @return Error code
*/
hipError_t hipHostGetFlags(unsigned int* flagsPtr, void* hostPtr) ;
@@ -786,7 +787,7 @@ hipError_t hipMemcpyToSymbol(const char* symbolName, const void *src, size_t siz
/**
* @brief Copy data from src to dst asynchronously.
*
* @warning If host or dest are not pinned, the memory copy will be performed synchronously. For best performance, use hipHostAlloc to
* @warning If host or dest are not pinned, the memory copy will be performed synchronously. For best performance, use hipHostMalloc to
* allocate host memory that is transferred asynchronously.
*
* @param[out] dst Data being copy to
projects/clr/hipamd/include/hip_runtime_api.h
+2 -2
Просмотреть файл
@@ -220,8 +220,8 @@ static inline hipError_t hipMalloc ( T** devPtr, size_t size)
// Provide an override to automatically typecast the pointer type from void**, and also provide a default for the flags.
template<class T>
static inline hipError_t hipHostAlloc( T** ptr, size_t size, unsigned int flags = hipHostAllocDefault)
static inline hipError_t hipHostMalloc( T** ptr, size_t size, unsigned int flags = hipHostMallocDefault)
{
return hipHostAlloc((void**)ptr, size, flags);
return hipHostMalloc((void**)ptr, size, flags);
}
#endif
projects/clr/hipamd/include/nvcc_detail/hip_runtime_api.h
+2 -2
Просмотреть файл
@@ -119,13 +119,13 @@ inline static hipError_t hipFree(void* ptr) {
return hipCUDAErrorTohipError(cudaFree(ptr));
}
inline static hipError_t hipMallocHost(void** ptr, size_t size) __attribute__((deprecated("use hipHostAlloc instead")));
inline static hipError_t hipMallocHost(void** ptr, size_t size) __attribute__((deprecated("use hipHostMalloc instead")));
inline static hipError_t hipMallocHost(void** ptr, size_t size) {
return hipCUDAErrorTohipError(cudaMallocHost(ptr, size));
}
inline static hipError_t hipHostAlloc(void** ptr, size_t size, unsigned int flags){
inline static hipError_t hipHostMalloc(void** ptr, size_t size, unsigned int flags){
return hipCUDAErrorTohipError(cudaHostAlloc(ptr, size, flags));
}
projects/clr/hipamd/samples/1_Utils/hipBusBandwidth/hipBusBandwidth.cpp
+10 -10
Просмотреть файл
@@ -108,7 +108,7 @@ void RunBenchmark_H2D(ResultDatabase &resultDB)
float *hostMem = NULL;
if (p_pinned)
{
hipHostAlloc((void**)&hostMem, sizeof(float) * numMaxFloats, hipHostAllocDefault);
hipHostMalloc((void**)&hostMem, sizeof(float) * numMaxFloats);
while (hipGetLastError() != hipSuccess)
{
// drop the size and try again
@@ -120,7 +120,7 @@ void RunBenchmark_H2D(ResultDatabase &resultDB)
return;
}
numMaxFloats = 1024 * (sizes[nSizes-1]) / 4;
hipHostAlloc((void**)&hostMem, sizeof(float) * numMaxFloats, hipHostAllocDefault);
hipHostMalloc((void**)&hostMem, sizeof(float) * numMaxFloats);
}
}
else
@@ -232,7 +232,7 @@ void RunBenchmark_H2D(ResultDatabase &resultDB)
CHECK_HIP_ERROR();
if (p_pinned)
{
hipFreeHost((void*)hostMem);
hipHostFree((void*)hostMem);
CHECK_HIP_ERROR();
}
else
@@ -259,15 +259,15 @@ void RunBenchmark_D2H(ResultDatabase &resultDB)
float *hostMem2;
if (p_pinned)
{
hipHostAlloc((void**)&hostMem1, sizeof(float)*numMaxFloats, hipHostAllocDefault);
hipHostMalloc((void**)&hostMem1, sizeof(float)*numMaxFloats);
hipError_t err1 = hipGetLastError();
hipHostAlloc((void**)&hostMem2, sizeof(float)*numMaxFloats, hipHostAllocDefault);
hipHostMalloc((void**)&hostMem2, sizeof(float)*numMaxFloats);
hipError_t err2 = hipGetLastError();
while (err1 != hipSuccess || err2 != hipSuccess)
{
// free the first buffer if only the second failed
if (err1 == hipSuccess)
hipFreeHost((void*)hostMem1);
hipHostFree((void*)hostMem1);
// drop the size and try again
if (p_verbose) std::cout << " - dropping size allocating pinned mem\n";
@@ -383,9 +383,9 @@ void RunBenchmark_D2H(ResultDatabase &resultDB)
CHECK_HIP_ERROR();
if (p_pinned)
{
hipFreeHost((void*)hostMem1);
hipHostFree((void*)hostMem1);
CHECK_HIP_ERROR();
hipFreeHost((void*)hostMem2);
hipHostFree((void*)hostMem2);
CHECK_HIP_ERROR();
}
else
@@ -523,8 +523,8 @@ void RunBenchmark_Bidir(ResultDatabase &resultDB)
CHECK_HIP_ERROR();
if (p_pinned)
{
hipFreeHost((void*)hostMem[0]);
hipFreeHost((void*)hostMem[1]);
hipHostFree((void*)hostMem[0]);
hipHostFree((void*)hostMem[1]);
CHECK_HIP_ERROR();
}
else
projects/clr/hipamd/src/hip_hcc.cpp
+11 -3
Просмотреть файл
@@ -2085,7 +2085,7 @@ hipError_t hipMallocHost(void** ptr, size_t sizeBytes)
}
hipError_t hipHostAlloc(void** ptr, size_t sizeBytes, unsigned int flags){
hipError_t hipHostMalloc(void** ptr, size_t sizeBytes, unsigned int flags){
std::call_once(hip_initialized, ihipInit);
hipError_t hip_status = hipSuccess;
@@ -2093,7 +2093,7 @@ hipError_t hipHostAlloc(void** ptr, size_t sizeBytes, unsigned int flags){
auto device = ihipGetTlsDefaultDevice();
if(device){
if(flags == hipHostAllocDefault){
if(flags == hipHostMallocDefault){
*ptr = hc::am_alloc(sizeBytes, device->_acc, amHostPinned);
if(sizeBytes && (*ptr == NULL)){
hip_status = hipErrorMemoryAllocation;
@@ -2101,7 +2101,7 @@ hipError_t hipHostAlloc(void** ptr, size_t sizeBytes, unsigned int flags){
hc::am_memtracker_update(*ptr, device->_device_index, 0);
}
tprintf(DB_MEM, " %s: pinned ptr=%p\n", __func__, *ptr);
} else if(flags & hipHostAllocMapped){
} else if(flags & hipHostMallocMapped){
*ptr = hc::am_alloc(sizeBytes, device->_acc, amHostPinned);
if(sizeBytes && (*ptr == NULL)){
hip_status = hipErrorMemoryAllocation;
@@ -2115,6 +2115,14 @@ hipError_t hipHostAlloc(void** ptr, size_t sizeBytes, unsigned int flags){
}
// TODO - remove me, this is deprecated.
hipError_t hipHostAlloc(void** ptr, size_t sizeBytes, unsigned int flags)
{
return hipHostMalloc(ptr, sizeBytes, flags);
};
hipError_t hipHostGetDevicePointer(void** devPtr, void* hstPtr, size_t size){
std::call_once(hip_initialized, ihipInit);
projects/clr/hipamd/tests/src/hipHostAlloc.cpp
+3 -3
Просмотреть файл
@@ -42,9 +42,9 @@ if(prop.canMapHostMemory != 1){
//std::cout<<"Exiting..."<<std::endl;
}
HIPCHECK(hipHostAlloc((void**)&A, SIZE, hipHostAllocWriteCombined | hipHostAllocMapped));
HIPCHECK(hipHostAlloc((void**)&B, SIZE, hipHostAllocDefault));
HIPCHECK(hipHostAlloc((void**)&C, SIZE, hipHostAllocMapped));
HIPCHECK(hipHostMalloc((void**)&A, SIZE, hipHostMallocWriteCombined | hipHostMallocMapped));
HIPCHECK(hipHostMalloc((void**)&B, SIZE, hipHostMallocDefault));
HIPCHECK(hipHostMalloc((void**)&C, SIZE, hipHostMallocMapped));
HIPCHECK(hipHostGetDevicePointer((void**)&Ad, A, 0));
HIPCHECK(hipHostGetDevicePointer((void**)&Cd, C, 0));
projects/clr/hipamd/tests/src/hipHostGetFlags.cpp
+7 -7
Просмотреть файл
@@ -35,9 +35,9 @@ int main(){
float *A, *B, *C, *D;
float *Ad, *Bd, *Cd, *Dd;
unsigned int FlagA, FlagB, FlagC;
FlagA = hipHostAllocWriteCombined | hipHostAllocMapped;
FlagB = hipHostAllocWriteCombined | hipHostAllocMapped;
FlagC = hipHostAllocMapped;
FlagA = hipHostMallocWriteCombined | hipHostMallocMapped;
FlagB = hipHostMallocWriteCombined | hipHostMallocMapped;
FlagC = hipHostMallocMapped;
hipDeviceProp_t prop;
int device;
HIPCHECK(hipGetDevice(&device));
@@ -45,11 +45,11 @@ HIPCHECK(hipGetDeviceProperties(&prop, device));
if(prop.canMapHostMemory != 1){
std::cout<<"Exiting..."<<std::endl;
}
HIPCHECK(hipHostAlloc((void**)&A, SIZE, hipHostAllocWriteCombined | hipHostAllocMapped));
HIPCHECK(hipHostAlloc((void**)&B, SIZE, hipHostAllocWriteCombined | hipHostAllocMapped));
HIPCHECK(hipHostAlloc((void**)&C, SIZE, hipHostAllocMapped));
HIPCHECK(hipHostMalloc((void**)&A, SIZE, hipHostMallocWriteCombined | hipHostMallocMapped));
HIPCHECK(hipHostMalloc((void**)&B, SIZE, hipHostMallocWriteCombined | hipHostMallocMapped));
HIPCHECK(hipHostMalloc((void**)&C, SIZE, hipHostMallocMapped));
HIPCHECK(hipHostAlloc((void**)&D, SIZE, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&D, SIZE, hipHostMallocDefault));
unsigned int flagA, flagB, flagC;
projects/clr/hipamd/tests/src/hipMemcpy.cpp
+3 -3
Просмотреть файл
@@ -36,7 +36,7 @@ void printSep()
// The subroutine allocates memory , copies to device, runs a vector add kernel, copies back, and checks the result.
//
// IN: numElements controls the number of elements used for allocations.
// IN: usePinnedHost : If true, allocate host with hipHostAlloc and is pinned ; else allocate host memory with malloc.
// IN: usePinnedHost : If true, allocate host with hipHostMalloc and is pinned ; else allocate host memory with malloc.
// IN: useHostToHost : If true, add an extra host-to-host copy.
// IN: useDeviceToDevice : If true, add an extra deviceto-device copy after result is produced.
// IN: useMemkindDefault : If true, use memkinddefault (runtime figures out direction). if false, use explicit memcpy direction.
@@ -67,8 +67,8 @@ void memcpytest2(size_t numElements, bool usePinnedHost, bool useHostToHost, boo
if (useHostToHost) {
if (usePinnedHost) {
HIPCHECK ( hipHostAlloc((void**)&A_hh, sizeElements, hipHostAllocDefault) );
HIPCHECK ( hipHostAlloc((void**)&B_hh, sizeElements, hipHostAllocDefault) );
HIPCHECK ( hipHostMalloc((void**)&A_hh, sizeElements, hipHostMallocDefault) );
HIPCHECK ( hipHostMalloc((void**)&B_hh, sizeElements, hipHostMallocDefault) );
} else {
A_hh = (T*)malloc(sizeElements);
B_hh = (T*)malloc(sizeElements);
projects/clr/hipamd/tests/src/hipMemcpyAsync.cpp
+4 -4
Просмотреть файл
@@ -33,7 +33,7 @@ void simpleNegTest()
size_t Nbytes = N*sizeof(float);
A_malloc = (float*)malloc(Nbytes);
HIPCHECK(hipHostAlloc((void**)&A_pinned, Nbytes, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&A_pinned, Nbytes, hipHostMallocDefault));
HIPCHECK(hipMalloc(&A_d, Nbytes));
@@ -61,7 +61,7 @@ struct HostTraits<Pinned>
static void *Alloc(size_t sizeBytes) {
void *p;
HIPCHECK(hipHostAlloc((void**)&p, sizeBytes, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&p, sizeBytes, hipHostMallocDefault));
return p;
};
};
@@ -181,8 +181,8 @@ void test_manyInflightCopies(hipStream_t stream, int numElements, int numCopies,
T *A_d;
T *A_h1, *A_h2;
HIPCHECK(hipHostAlloc((void**)&A_h1, Nbytes, hipHostAllocDefault));
HIPCHECK(hipHostAlloc((void**)&A_h2, Nbytes, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&A_h1, Nbytes, hipHostMallocDefault));
HIPCHECK(hipHostMalloc((void**)&A_h2, Nbytes, hipHostMallocDefault));
HIPCHECK(hipMalloc(&A_d, Nbytes));
for (int i=0; i<numElements; i++) {
projects/clr/hipamd/tests/src/hipMemcpy_simple.cpp
+2 -2
Просмотреть файл
@@ -68,8 +68,8 @@ void simpleTest2(size_t numElements, bool usePinnedHost)
T *A_d, *A_h1, *A_h2;
if (usePinnedHost) {
HIPCHECK ( hipHostAlloc((void**)&A_h1, sizeElements, hipHostAllocDefault) );
HIPCHECK ( hipHostAlloc((void**)&A_h2, sizeElements, hipHostAllocDefault) );
HIPCHECK ( hipHostMalloc((void**)&A_h1, sizeElements, hipHostMallocDefault) );
HIPCHECK ( hipHostMalloc((void**)&A_h2, sizeElements, hipHostMallocDefault) );
} else {
A_h1 = (T*)aligned_alloc(alignment, sizeElements);
HIPASSERT(A_h1);
projects/clr/hipamd/tests/src/hipMultiThreadStreams2.cpp
+9 -9
Просмотреть файл
@@ -34,11 +34,11 @@ Array[tx] = Array[tx] + T(1);
void run1(size_t size, hipStream_t stream){
float *Ah, *Bh, *Cd, *Dd, *Eh;
HIPCHECK(hipHostAlloc((void**)&Ah, size, hipHostAllocDefault));
HIPCHECK(hipHostAlloc((void**)&Bh, size, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&Ah, size, hipHostMallocDefault));
HIPCHECK(hipHostMalloc((void**)&Bh, size, hipHostMallocDefault));
HIPCHECK(hipMalloc(&Cd, size));
HIPCHECK(hipMalloc(&Dd, size));
HIPCHECK(hipHostAlloc((void**)&Eh, size, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&Eh, size, hipHostMallocDefault));
for(int i=0;i<N;i++){
Ah[i] = 1.0f;
@@ -58,16 +58,16 @@ void run(size_t size, hipStream_t stream1, hipStream_t stream2){
float *Ah, *Bh, *Cd, *Dd, *Eh;
float *Ahh, *Bhh, *Cdd, *Ddd, *Ehh;
HIPCHECK(hipHostAlloc((void**)&Ah, size, hipHostAllocDefault));
HIPCHECK(hipHostAlloc((void**)&Bh, size, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&Ah, size, hipHostMallocDefault));
HIPCHECK(hipHostMalloc((void**)&Bh, size, hipHostMallocDefault));
HIPCHECK(hipMalloc(&Cd, size));
HIPCHECK(hipMalloc(&Dd, size));
HIPCHECK(hipHostAlloc((void**)&Eh, size, hipHostAllocDefault));
HIPCHECK(hipHostAlloc((void**)&Ahh, size, hipHostAllocDefault));
HIPCHECK(hipHostAlloc((void**)&Bhh, size, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&Eh, size, hipHostMallocDefault));
HIPCHECK(hipHostMalloc((void**)&Ahh, size, hipHostMallocDefault));
HIPCHECK(hipHostMalloc((void**)&Bhh, size, hipHostMallocDefault));
HIPCHECK(hipMalloc(&Cdd, size));
HIPCHECK(hipMalloc(&Ddd, size));
HIPCHECK(hipHostAlloc((void**)&Ehh, size, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&Ehh, size, hipHostMallocDefault));
HIPCHECK(hipMemcpyAsync(Bh, Ah, size, hipMemcpyHostToHost, stream1));
HIPCHECK(hipMemcpyAsync(Bhh, Ahh, size, hipMemcpyHostToHost, stream2));
projects/clr/hipamd/tests/src/hipPointerAttrib.cpp
+3 -3
Просмотреть файл
@@ -115,7 +115,7 @@ void testSimple()
hipError_t e;
HIPCHECK ( hipMalloc(&A_d, Nbytes) );
HIPCHECK ( hipHostAlloc((void**)&A_Pinned_h, Nbytes, hipHostAllocDefault) );
HIPCHECK ( hipHostMalloc((void**)&A_Pinned_h, Nbytes, hipHostMallocDefault) );
A_OSAlloc_h = (char*)malloc(Nbytes);
size_t free, total;
@@ -168,7 +168,7 @@ void testSimple()
// Device-visible host memory
printf ("\nDevice-visible host memory (hipHostAlloc)\n");
printf ("\nDevice-visible host memory (hipHostMalloc)\n");
HIPCHECK( hipPointerGetAttributes(&attribs, A_Pinned_h));
printf("getAttr:%-20s", "A_pinned_h"); printAttribs(&attribs);
@@ -283,7 +283,7 @@ void clusterAllocs(int numAllocs, size_t minSize, size_t maxSize)
reference[i]._attrib.hostPointer = NULL;
reference[i]._attrib.allocationFlags = 0; // TODO-randomize these.
} else {
HIPCHECK(hipHostAlloc((void**)&ptr, reference[i]._sizeBytes, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&ptr, reference[i]._sizeBytes, hipHostMallocDefault));
reference[i]._attrib.memoryType = hipMemoryTypeHost;
reference[i]._attrib.devicePointer = ptr;
reference[i]._attrib.hostPointer = ptr;
projects/clr/hipamd/tests/src/hipStream.h
+2 -2
Просмотреть файл
@@ -85,7 +85,7 @@ void initArrays(T **Ad, T **Ah,
HIPCHECK( hipMalloc(Ad, NBytes));
}
if(usePinnedHost){
HIPCHECK( hipHostAlloc((void**)Ah, NBytes, hipHostAllocDefault));
HIPCHECK( hipHostMalloc((void**)Ah, NBytes, hipHostMallocDefault));
}
else{
*Ah = new T[N];
@@ -102,7 +102,7 @@ void initArrays(T **Ad, size_t N,
HIPCHECK( hipMalloc(Ad, NBytes));
}else{
if(usePinnedHost){
HIPCHECK(hipHostAlloc((void**)Ad, NBytes, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)Ad, NBytes, hipHostMallocDefault));
}else{
*Ad = new T[N];
HIPASSERT(*Ad != NULL);
projects/clr/hipamd/tests/src/test_common.h
+4 -4
Просмотреть файл
@@ -141,13 +141,13 @@ void initArrays(T **A_d, T **B_d, T **C_d,
if (usePinnedHost) {
if (A_h) {
HIPCHECK ( hipHostAlloc((void**)A_h, Nbytes, hipHostAllocDefault) );
HIPCHECK ( hipHostMalloc((void**)A_h, Nbytes) );
}
if (B_h) {
HIPCHECK ( hipHostAlloc((void**)B_h, Nbytes, hipHostAllocDefault) );
HIPCHECK ( hipHostMalloc((void**)B_h, Nbytes) );
}
if (C_h) {
HIPCHECK ( hipHostAlloc((void**)C_h, Nbytes, hipHostAllocDefault) );
HIPCHECK ( hipHostMalloc((void**)C_h, Nbytes) );
}
} else {
if (A_h) {
@@ -258,7 +258,7 @@ struct Pinned {
static void *Alloc(size_t sizeBytes)
{
void *p;
HIPCHECK(hipHostAlloc((void**)&p, sizeBytes, hipHostAllocDefault));
HIPCHECK(hipHostMalloc((void**)&p, sizeBytes));
return p;
};
};
projects/clr/hipamd/util/vim/hip.vim
+17 -1
Просмотреть файл
@@ -103,7 +103,13 @@ syn keyword hipFunctionName hipEventRecord
syn keyword hipFunctionName hipEventSynchronize
syn keyword hipFunctionName hipFree
syn keyword hipFunctionName hipFreeArray
syn keyword hipFunctionName hipFreeHost
syn keyword hipFunctionName hipHostMalloc
syn keyword hipFunctionName hipHostFree
syn keyword hipFunctionName hipHostGetDevicePointer
syn keyword hipFunctionName hipHostGetFlags
syn keyword hipFunctionName hipHostRegister
syn keyword hipFunctionName hipHostUnregister
syn keyword hipFunctionName hipGetChannelDesc
syn keyword hipFunctionName hipGetDevice
syn keyword hipFunctionName hipGetDeviceCount
@@ -165,6 +171,16 @@ syn keyword hipFlags hipReadModeElementType
syn keyword hipFlags hipSuccess
syn keyword hipFlags hipTextureType1D
syn keyword hipFlags hipHostMallocDefault
syn keyword hipFlags hipHostMallocPortable
syn keyword hipFlags hipHostMallocMapped
syn keyword hipFlags hipHostMallocWriteCombined
syn keyword hipFlags hipHostRegisterDefault
syn keyword hipFlags hipHostRegisterPortable
syn keyword hipFlags hipHostRegisterMapped
syn keyword hipFlags hipHostRegisterIoMemory
" Define the default highlighting.
" For version 5.7 and earlier: only when not done already
" For version 5.8 and later: only when an item doesn't have highlighting yet