Skip to content

Commit fbcf1ff

Browse files
YdrMasterYdrMaster
committed
fix: 所有 unsigned long int 替换为 uint64_t
Signed-off-by: YdrMaster <ydrml@hotmail.com>
1 parent 15e2474 commit fbcf1ff

File tree

19 files changed

+119
-120
lines changed

19 files changed

+119
-120
lines changed

‎src/ops/causal_softmax/bang/causal_softmax_bang.h‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -25,7 +25,7 @@ infiniopStatus_t bangGetCausalSoftmaxWorkspaceSize(CausalSoftmaxBangDescriptor_t
2525

2626
infiniopStatus_t bangCausalSoftmax(CausalSoftmaxBangDescriptor_t desc,
2727
void *workspace,
28-
unsigned long int workspace_size,
28+
uint64_t workspace_size,
2929
void *data,
3030
void *stream);
3131

‎src/ops/causal_softmax/bang/causal_softmax_bang.mlu‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -787,7 +787,7 @@ void causal_softmax_bang_f16(CausalSoftmaxBangDescriptor_t desc, void *workspace
787787

788788
infiniopStatus_t bangCausalSoftmax(CausalSoftmaxBangDescriptor_t desc,
789789
void *workspace,
790-
unsigned long int workspace_size,
790+
uint64_t workspace_size,
791791
void *data,
792792
void *stream) {
793793
if (cnrtSetDevice(desc->device_id) != cnrtSuccess) {
@@ -798,4 +798,4 @@ infiniopStatus_t bangCausalSoftmax(CausalSoftmaxBangDescriptor_t desc,
798798
return STATUS_SUCCESS;
799799
}
800800
return STATUS_BAD_TENSOR_DTYPE;
801-
}
801+
}

‎src/ops/causal_softmax/bang/causal_softmax_cnnl.cc‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -52,7 +52,7 @@ infiniopStatus_t cnnlDestroyCausalSoftmaxDescriptor(CausalSoftmaxCnnlDescriptor_
5252

5353
infiniopStatus_t cnnlCausalSoftmax(CausalSoftmaxCnnlDescriptor_t desc,
5454
void *workspace,
55-
unsigned long int workspace_size,
55+
uint64_t workspace_size,
5656
void *data,
5757
void *stream) {
5858
if (cnrtSetDevice(desc->device_id) != cnrtSuccess) {

‎src/ops/causal_softmax/bang/causal_softmax_cnnl.h‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -26,7 +26,7 @@ infiniopStatus_t cnnlGetCausalSoftmaxWorkspaceSize(CausalSoftmaxCnnlDescriptor_t
2626

2727
infiniopStatus_t cnnlCausalSoftmax(CausalSoftmaxCnnlDescriptor_t desc,
2828
void *workspace,
29-
unsigned long int workspace_size,
29+
uint64_t workspace_size,
3030
void *data,
3131
void *stream);
3232

‎src/ops/causal_softmax/cuda/causal_softmax.cc‎

Lines changed: 7 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -5,20 +5,20 @@
55
infiniopStatus_t cudaCreateCausalSoftmaxDescriptor(CudaHandle_t handle,
66
CausalSoftmaxCudaDescriptor_t *desc_ptr,
77
infiniopTensorDescriptor_t y) {
8-
unsigned long int ndim = y->ndim;
8+
uint64_t ndim = y->ndim;
99
// TODO: only support 2d or 3d tensor
1010
if (ndim != 2 && ndim != 3) {
1111
return STATUS_BAD_TENSOR_SHAPE;
1212
}
1313
if (!dtype_eq(y->dt, F16)) {
1414
return STATUS_BAD_TENSOR_DTYPE;
1515
}
16-
unsigned long int total_seq_len = y->shape[ndim - 1];
17-
unsigned long int seq_len = y->shape[ndim - 2];
18-
unsigned long int batch_size = 1;
19-
unsigned long int stride_b = 0;
20-
unsigned long int stride_i = y->strides[ndim - 2];
21-
unsigned long int stride_j = y->strides[ndim - 1];
16+
uint64_t total_seq_len = y->shape[ndim - 1];
17+
uint64_t seq_len = y->shape[ndim - 2];
18+
uint64_t batch_size = 1;
19+
uint64_t stride_b = 0;
20+
uint64_t stride_i = y->strides[ndim - 2];
21+
uint64_t stride_j = y->strides[ndim - 1];
2222
if (stride_j != 1) {
2323
return STATUS_BAD_TENSOR_STRIDES;
2424
}

‎src/ops/causal_softmax/cuda/causal_softmax.cu‎

Lines changed: 12 additions & 12 deletions
Original file line numberDiff line numberDiff line change
@@ -218,17 +218,17 @@ __global__ void fused_softmax_standard(
218218
}
219219

220220

221-
void causal_softmax_nv_gpu_f16(CausalSoftmaxCudaDescriptor_t desc, void* y, void *stream) {
222-
unsigned long int total_seq_len = desc->total_seq_len;
223-
unsigned long int seq_len = desc->seq_len;
224-
unsigned long int batch_size = desc->batch_size;
225-
unsigned long int stride_x = desc->stride_b;
226-
unsigned long int stride_y = desc->stride_i;
227-
unsigned long int stride_z = desc->stride_j;// covert byte strides to element strides
221+
void causal_softmax_nv_gpu_f16(CausalSoftmaxCudaDescriptor_t desc, void *y, void *stream) {
222+
uint64_t total_seq_len = desc->total_seq_len;
223+
uint64_t seq_len = desc->seq_len;
224+
uint64_t batch_size = desc->batch_size;
225+
uint64_t stride_x = desc->stride_b;
226+
uint64_t stride_y = desc->stride_i;
227+
uint64_t stride_z = desc->stride_j;// covert byte strides to element strides
228228
unsigned int max_items_per_thread = desc->max_items_per_thread;
229229

230230
dim3 grid(batch_size, seq_len);
231-
231+
232232
if (max_items_per_thread == 1) {
233233
fused_softmax_padding<MAX_THREADS_PER_BLOCK>
234234
<<<grid, total_seq_len, 0, (cudaStream_t) stream>>>((half *) (y), stride_x, stride_y, stride_z);
@@ -243,13 +243,13 @@ void causal_softmax_nv_gpu_f16(CausalSoftmaxCudaDescriptor_t desc, void* y, void
243243

244244
infiniopStatus_t cudaCausalSoftmax(CausalSoftmaxCudaDescriptor_t desc,
245245
void *workspace,
246-
unsigned long int workspace_size,
246+
uint64_t workspace_size,
247247
void *data,
248-
void *stream){
249-
if(cudaSetDevice(desc->device_id) != cudaSuccess){
248+
void *stream) {
249+
if (cudaSetDevice(desc->device_id) != cudaSuccess) {
250250
return STATUS_BAD_DEVICE;
251251
}
252-
if (dtype_eq(desc->dtype, F16)){
252+
if (dtype_eq(desc->dtype, F16)) {
253253
causal_softmax_nv_gpu_f16(desc, data, stream);
254254
return STATUS_SUCCESS;
255255
}

‎src/ops/causal_softmax/cuda/causal_softmax.cuh‎

Lines changed: 7 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -8,12 +8,12 @@ struct CausalSoftmaxCudaDescriptor {
88
Device device;
99
int device_id;
1010
DT dtype;
11-
unsigned long int batch_size;
12-
unsigned long int stride_b;
13-
unsigned long int seq_len;
14-
unsigned long int stride_i;
15-
unsigned long int total_seq_len;
16-
unsigned long int stride_j;
11+
uint64_t batch_size;
12+
uint64_t stride_b;
13+
uint64_t seq_len;
14+
uint64_t stride_i;
15+
uint64_t total_seq_len;
16+
uint64_t stride_j;
1717
unsigned int max_items_per_thread;
1818
};
1919

@@ -27,7 +27,7 @@ infiniopStatus_t cudaGetCausalSoftmaxWorkspaceSize(CausalSoftmaxCudaDescriptor_t
2727

2828
infiniopStatus_t cudaCausalSoftmax(CausalSoftmaxCudaDescriptor_t desc,
2929
void *workspace,
30-
unsigned long int workspace_size,
30+
uint64_t workspace_size,
3131
void *data,
3232
void *stream);
3333

‎src/ops/matmul/ascend/matmul_aclnn.cc‎

Lines changed: 10 additions & 11 deletions
Original file line numberDiff line numberDiff line change
@@ -69,13 +69,12 @@ infiniopStatus_t aclnnCreateMatmulDescriptor(AscendHandle_t handle,
6969
// aclnnGemm support C = alpha * A @ B + beta * C
7070
// see https://www.hiascend.com/document/detail/zh/CANNCommunityEdition/80RC3alpha003/apiref/aolapi/context/aclnnGemm.md
7171
ret = aclnnGemmGetWorkspaceSize(ta, tb, tc, (*desc_ptr)->alpha, (*desc_ptr)->beta, transA, transB, tc,
72-
(*desc_ptr)->mt, &workspaceSize, &executor);
72+
(*desc_ptr)->mt, &workspaceSize, &executor);
7373
CHECK_RET(ret == ACL_SUCCESS,
74-
LOG_PRINT("aclnnGemmGetWorkspaceSize failed. ERROR: %d\n", ret);
75-
return STATUS_EXECUTION_FAILED);
74+
LOG_PRINT("aclnnGemmGetWorkspaceSize failed. ERROR: %d\n", ret);
75+
return STATUS_EXECUTION_FAILED);
7676
aclSetAclOpExecutorRepeatable(executor);
7777

78-
7978
return STATUS_SUCCESS;
8079
}
8180

@@ -109,14 +108,14 @@ infiniopStatus_t aclnnMatmul(MatmulAclnnDescriptor_t desc,
109108
aclrtSetDevice(desc->device_id);
110109

111110
for (int i = 0; i < batch; i++) {
112-
AclSetTensorAddr(executor, 0, ta, (char *)(a) + i * desc->info->a_matrix.stride * desc->dtype.size);
113-
AclSetTensorAddr(executor, 1, tb, (char *)(b) + i * desc->info->b_matrix.stride * desc->dtype.size);
114-
AclSetTensorAddr(executor, 2, tc, (char *)(c) + i * desc->info->c_matrix.stride * desc->dtype.size);
115-
AclSetTensorAddr(executor, 3, tc, (char *)(c) + i * desc->info->c_matrix.stride * desc->dtype.size);
111+
AclSetTensorAddr(executor, 0, ta, (char *) (a) + i * desc->info->a_matrix.stride * desc->dtype.size);
112+
AclSetTensorAddr(executor, 1, tb, (char *) (b) + i * desc->info->b_matrix.stride * desc->dtype.size);
113+
AclSetTensorAddr(executor, 2, tc, (char *) (c) + i * desc->info->c_matrix.stride * desc->dtype.size);
114+
AclSetTensorAddr(executor, 3, tc, (char *) (c) + i * desc->info->c_matrix.stride * desc->dtype.size);
116115
aclnnStatus ret = aclnnGemm(workspace,
117-
workspaceSize,
118-
executor,
119-
stream);
116+
workspaceSize,
117+
executor,
118+
stream);
120119
CHECK_RET(ret == ACL_SUCCESS,
121120
LOG_PRINT("aclnnGemm failed. ERROR: %d\n", ret);
122121
return STATUS_EXECUTION_FAILED);

‎src/ops/random_sample/bang/random_sample_bang.h‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -24,7 +24,7 @@ infiniopStatus_t bangGetRandomSampleWorkspaceSize(RandomSampleBangDescriptor_t d
2424

2525
infiniopStatus_t bangRandomSample(RandomSampleBangDescriptor_t desc,
2626
void *workspace,
27-
unsigned long int workspace_size,
27+
uint64_t workspace_size,
2828
void *result,
2929
void const *probs,
3030
float random_val,

0 commit comments

Comments
 (0)